Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

fixed gfx1202 error as gfx1102 #2816

Open
wants to merge 1 commit into
base: r2.16-rocm-enhanced
Choose a base branch
from
Open
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Original file line number Diff line number Diff line change
Expand Up @@ -116,7 +116,7 @@ bool HasFastFP16Support(const DeviceProperties& props) {
absl::flat_hash_set<std::string> FP16SupportedDevices = {
{"gfx906"}, {"gfx908"}, {"gfx90a"}, {"gfx910"}, {"gfx940"}, {"gfx941"},
{"gfx942"}, {"gfx1010"}, {"gfx1012"}, {"gfx1030"}, {"gfx1100"},
{"gfx1200"},{"gfx1201"}, {"gfx1202"}
{"gfx1200"},{"gfx1201"}, {"gfx1102"}
};
std::string gcnArchName = props.environment().at("architecture");
std::vector<std::string> gpu_arch = absl::StrSplit(gcnArchName, ":");
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -75,7 +75,7 @@ inline GpuStats GetNumGPUs(const Cluster& cluster) {
compute_capability_it->second == "gfx942" ||
compute_capability_it->second == "gfx1200" ||
compute_capability_it->second == "gfx1201" ||
compute_capability_it->second == "gfx1202") && is_enabled) {
compute_capability_it->second == "gfx1102") && is_enabled) {
gpu_stats.num_voltas++;
}
#endif
Expand Down
6 changes: 3 additions & 3 deletions tensorflow/core/util/gpu_device_functions.h
Original file line number Diff line number Diff line change
Expand Up @@ -743,7 +743,7 @@ __device__ inline double GpuAtomicAdd(double* ptr, double value) {
}
#endif

#if __gfx908__ || __gfx90a__ || __gfx940__ || __gfx941__ || __gfx942__ || __gfx1200__ || __gfx1201__ || __gfx1202__
#if __gfx908__ || __gfx90a__ || __gfx940__ || __gfx941__ || __gfx942__ || __gfx1200__ || __gfx1201__ || __gfx1102__

#define ADDRSP1 __attribute__((address_space(1)))
__device__ float
Expand Down Expand Up @@ -963,7 +963,7 @@ __device__ inline int64_t GpuAtomicMin(int64_t* ptr, int64_t value) {
}
#endif

#if __gfx908__ || __gfx90a__ || __gfx940__ || __gfx941__ || __gfx942__ || __gfx1200__ || __gfx1201__ || __gfx1202__
#if __gfx908__ || __gfx90a__ || __gfx940__ || __gfx941__ || __gfx942__ || __gfx1200__ || __gfx1201__ || __gfx1102__
// Low level instructions don't return. For now, assume that return value
// is always unused.
__device__ float GpuAtomicAdd(float* dst, float val) {
Expand All @@ -978,7 +978,7 @@ __device__ inline T GpuAtomicAddShared(T* ptr, T value) {
return GpuAtomicAdd(ptr, value);
}

#if __gfx908__ || __gfx90a__ || __gfx940__ || __gfx941__ || __gfx942__ || __gfx1200__ || __gfx1201__ || __gfx1202__
#if __gfx908__ || __gfx90a__ || __gfx940__ || __gfx941__ || __gfx942__ || __gfx1200__ || __gfx1201__ || __gfx1102__
__device__ float GpuAtomicAddShared(float* dst, float val) {
atomicAdd(dst, val);
return val;
Expand Down
2 changes: 1 addition & 1 deletion tensorflow/tools/ci_build/Dockerfile.rocm
Original file line number Diff line number Diff line change
Expand Up @@ -7,7 +7,7 @@ ARG ROCM_DEB_REPO=https://repo.radeon.com/rocm/apt/6.1.2/
ARG ROCM_BUILD_NAME=ubuntu
ARG ROCM_BUILD_NUM=main
ARG ROCM_PATH=/opt/rocm/
ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1202"
ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1102"

ARG DEBIAN_FRONTEND=noninteractive
ENV TF_NEED_ROCM 1
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -8,7 +8,7 @@ COPY setup.packages.rocm.cs7.sh setup.packages.rocm.cs7.sh
COPY builder.packages.rocm.cs7.txt builder.packages.rocm.cs7.txt
RUN /setup.packages.rocm.cs7.sh /builder.packages.rocm.cs7.txt

ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1202"
ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1102"
ENV GPU_DEVICE_TARGETS=${GPU_DEVICE_TARGETS}

# Install ROCM
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -17,7 +17,7 @@ COPY setup.packages.rocm.el8.sh setup.packages.rocm.el8.sh
COPY builder.packages.rocm.el8.txt builder.packages.rocm.el8.txt
RUN /setup.packages.rocm.el8.sh /builder.packages.rocm.el8.txt

ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1202"
ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1102"
ENV GPU_DEVICE_TARGETS=${GPU_DEVICE_TARGETS}

# Install ROCM
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -2,7 +2,7 @@
FROM ubuntu:20.04
################################################################################

ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1202"
ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1102"
ENV GPU_DEVICE_TARGETS=${GPU_DEVICE_TARGETS}

# Install build dependencies
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -2,7 +2,7 @@
FROM ubuntu:22.04
################################################################################

ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1202"
ARG GPU_DEVICE_TARGETS="gfx908 gfx90a gfx940 gfx941 gfx942 gfx1030 gfx1100 gfx1200 gfx1201 gfx1102"
ENV GPU_DEVICE_TARGETS=${GPU_DEVICE_TARGETS}

# Install build dependencies
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -877,7 +877,7 @@ std::string MapGCNArchNameTokenToFeatureStr(const std::string& token,
return "+sramecc";
} else if (token == "sramecc-") {
if (gfx == "gfx90a" || gfx == "gfx940" || gfx == "gfx941" ||
gfx == "gfx942" || gfx == "gfx1200" || gfx == "gfx1201" || gfx == "gfx1202")
gfx == "gfx942" || gfx == "gfx1200" || gfx == "gfx1201" || gfx == "gfx1102")
return "";
return "-sramecc";
} else if (token == "xnack+") {
Expand Down
12 changes: 6 additions & 6 deletions third_party/xla/xla/stream_executor/device_description.h
Original file line number Diff line number Diff line change
Expand Up @@ -184,11 +184,11 @@ class RocmComputeCapability {

bool gfx10_rx69xx() const { return gfx_version() == "gfx1030"; }

bool gfx11_rx7900() const { return gfx_version() == "gfx1100"; }
bool gfx11_rx7900() const { return (gfx_version() == "gfx1100" ||
gfx_version() == "gfx1102"); }

bool gfx12_rx8900() const { return ((gfx_version() == "gfx1200") ||
(gfx_version() == "gfx1201") ||
(gfx_version() == "gfx1202")); }
bool gfx12_rx8900() const { return (gfx_version() == "gfx1200" ||
gfx_version() == "gfx1201"); }

bool has_nhwc_layout_support() const { return gfx9_mi100_or_later(); }

Expand Down Expand Up @@ -232,8 +232,8 @@ class RocmComputeCapability {
"gfx90a", // MI200
"gfx940", "gfx941", "gfx942", // MI300
"gfx1030", // RX68xx / RX69xx
"gfx1100", // RX7900
"gfx1200", "gfx1201", "gfx1202", // RX8900
"gfx1100", "gfx1102", // RX7900
"gfx1200", "gfx1201", // RX8900
};
};

Expand Down
Loading