[llvm] [mlir] [mlir] Replace MLIR_ENABLE_ROCM_CONVERSIONS with LLVM_HAS_AMDGPU_TARGET (PR #182652)
via llvm-commits
llvm-commits at lists.llvm.org
Fri Mar 13 13:53:55 PDT 2026
https://github.com/xys-syx updated https://github.com/llvm/llvm-project/pull/182652
>From b26c617e7e80b26e7cfdf684981b08335da873ee Mon Sep 17 00:00:00 2001
From: William Moses <gh at wsmoses.com>
Date: Thu, 19 Feb 2026 14:40:38 -0600
Subject: [PATCH 1/4] Change AMDGPU target initialization condition
---
mlir/lib/Target/LLVM/ROCDL/Target.cpp | 2 +-
1 file changed, 1 insertion(+), 1 deletion(-)
diff --git a/mlir/lib/Target/LLVM/ROCDL/Target.cpp b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
index e3b34c5d34c41..28a442c338d29 100644
--- a/mlir/lib/Target/LLVM/ROCDL/Target.cpp
+++ b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
@@ -112,7 +112,7 @@ void SerializeGPUModuleBase::init() {
static llvm::once_flag initializeBackendOnce;
llvm::call_once(initializeBackendOnce, []() {
// If the `AMDGPU` LLVM target was built, initialize it.
-#if MLIR_ENABLE_ROCM_CONVERSIONS
+#if LLVM_HAS_AMDGPU_TARGET
LLVMInitializeAMDGPUTarget();
LLVMInitializeAMDGPUTargetInfo();
LLVMInitializeAMDGPUTargetMC();
>From c813428fb1b5678a6e4c541f4aee82a7977e95ea Mon Sep 17 00:00:00 2001
From: Yuansui Xu <xuyuansui at outlook.com>
Date: Fri, 20 Feb 2026 22:11:36 -0600
Subject: [PATCH 2/4] add LLVM_HAS_AMDGPU_TARGET
---
mlir/CMakeLists.txt | 8 --------
mlir/include/mlir/Config/mlir-config.h.cmake | 4 ----
mlir/lib/Conversion/GPUCommon/CMakeLists.txt | 2 +-
mlir/lib/Target/LLVM/CMakeLists.txt | 4 ++--
mlir/lib/Target/LLVM/ROCDL/Target.cpp | 9 +++++----
mlir/test/CMakeLists.txt | 2 +-
mlir/test/lit.site.cfg.py.in | 2 +-
mlir/unittests/Target/LLVM/SerializeROCDLTarget.cpp | 3 ++-
utils/bazel/llvm-project-overlay/mlir/BUILD.bazel | 4 +++-
utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel | 2 +-
10 files changed, 16 insertions(+), 24 deletions(-)
diff --git a/mlir/CMakeLists.txt b/mlir/CMakeLists.txt
index 9e1e9314511e3..5b7779d035c87 100644
--- a/mlir/CMakeLists.txt
+++ b/mlir/CMakeLists.txt
@@ -132,14 +132,6 @@ option(MLIR_ENABLE_EXECUTION_ENGINE
"Enable building the MLIR Execution Engine."
${MLIR_ENABLE_EXECUTION_ENGINE_default})
-# Build the ROCm conversions and run according tests if the AMDGPU backend
-# is available.
-if ("AMDGPU" IN_LIST LLVM_TARGETS_TO_BUILD)
- set(MLIR_ENABLE_ROCM_CONVERSIONS 1)
-else()
- set(MLIR_ENABLE_ROCM_CONVERSIONS 0)
-endif()
-
# Build the XeVM conversions and run according tests if the SPIRV backend
# is available.
if ("SPIRV" IN_LIST LLVM_TARGETS_TO_BUILD)
diff --git a/mlir/include/mlir/Config/mlir-config.h.cmake b/mlir/include/mlir/Config/mlir-config.h.cmake
index dc5579439ec89..3013ec4ba5637 100644
--- a/mlir/include/mlir/Config/mlir-config.h.cmake
+++ b/mlir/include/mlir/Config/mlir-config.h.cmake
@@ -39,8 +39,4 @@
/* If set, enables features that depend on the NVIDIA's PTX compiler. */
#cmakedefine01 MLIR_ENABLE_NVPTXCOMPILER
-/* If set, enables ROCm-related features in ROCM-related transforms, pipelines,
- and targets. */
-#cmakedefine01 MLIR_ENABLE_ROCM_CONVERSIONS
-
#endif
diff --git a/mlir/lib/Conversion/GPUCommon/CMakeLists.txt b/mlir/lib/Conversion/GPUCommon/CMakeLists.txt
index ce914c0ea3dd8..9806545f713d8 100644
--- a/mlir/lib/Conversion/GPUCommon/CMakeLists.txt
+++ b/mlir/lib/Conversion/GPUCommon/CMakeLists.txt
@@ -6,7 +6,7 @@ if ("NVPTX" IN_LIST LLVM_TARGETS_TO_BUILD)
)
endif()
-if (MLIR_ENABLE_ROCM_CONVERSIONS)
+if ("AMDGPU" IN_LIST LLVM_TARGETS_TO_BUILD)
set(AMDGPU_LIBS
AMDGPUCodeGen
AMDGPUDesc
diff --git a/mlir/lib/Target/LLVM/CMakeLists.txt b/mlir/lib/Target/LLVM/CMakeLists.txt
index 2a16160267174..3b81c94dbbcd7 100644
--- a/mlir/lib/Target/LLVM/CMakeLists.txt
+++ b/mlir/lib/Target/LLVM/CMakeLists.txt
@@ -159,7 +159,7 @@ else()
endif()
-if (MLIR_ENABLE_ROCM_CONVERSIONS)
+if ("AMDGPU" IN_LIST LLVM_TARGETS_TO_BUILD)
set(AMDGPU_LIBS
AMDGPUAsmParser
AMDGPUCodeGen
@@ -188,7 +188,7 @@ add_mlir_dialect_library(MLIRROCDLTarget
MLIRROCDLToLLVMIRTranslation
)
-if(MLIR_ENABLE_ROCM_CONVERSIONS)
+if("AMDGPU" IN_LIST LLVM_TARGETS_TO_BUILD)
if (DEFINED ROCM_PATH)
set(DEFAULT_ROCM_PATH "${ROCM_PATH}" CACHE PATH "Fallback path to search for ROCm installs")
elseif(DEFINED ENV{ROCM_PATH})
diff --git a/mlir/lib/Target/LLVM/ROCDL/Target.cpp b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
index 28a442c338d29..b6656a95d6012 100644
--- a/mlir/lib/Target/LLVM/ROCDL/Target.cpp
+++ b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
@@ -19,6 +19,7 @@
#include "mlir/Target/LLVM/ROCDL/Utils.h"
#include "mlir/Target/LLVMIR/Export.h"
+#include "llvm/Config/Targets.h"
#include "llvm/IR/Constants.h"
#include "llvm/MC/MCAsmBackend.h"
#include "llvm/MC/MCAsmInfo.h"
@@ -446,7 +447,7 @@ FailureOr<SmallVector<char, 0>> SerializeGPUModuleBase::moduleToObjectImpl(
return compileToBinary(*serializedISA);
}
-#if MLIR_ENABLE_ROCM_CONVERSIONS
+#if LLVM_HAS_AMDGPU_TARGET
namespace {
class AMDGPUSerializer : public SerializeGPUModuleBase {
public:
@@ -471,7 +472,7 @@ FailureOr<SmallVector<char, 0>>
AMDGPUSerializer::moduleToObject(llvm::Module &llvmModule) {
return moduleToObjectImpl(targetOptions, llvmModule);
}
-#endif // MLIR_ENABLE_ROCM_CONVERSIONS
+#endif // LLVM_HAS_AMDGPU_TARGET
std::optional<mlir::gpu::SerializedObject>
ROCDLTargetAttrImpl::serializeToObject(
@@ -484,7 +485,7 @@ ROCDLTargetAttrImpl::serializeToObject(
module->emitError("module must be a GPU module");
return std::nullopt;
}
-#if MLIR_ENABLE_ROCM_CONVERSIONS
+#if LLVM_HAS_AMDGPU_TARGET
AMDGPUSerializer serializer(*module, cast<ROCDLTargetAttr>(attribute),
options);
serializer.init();
@@ -496,7 +497,7 @@ ROCDLTargetAttrImpl::serializeToObject(
module->emitError("the `AMDGPU` target was not built. Please enable it when "
"building LLVM");
return std::nullopt;
-#endif // MLIR_ENABLE_ROCM_CONVERSIONS
+#endif // LLVM_HAS_AMDGPU_TARGET
}
Attribute
diff --git a/mlir/test/CMakeLists.txt b/mlir/test/CMakeLists.txt
index e64935364997c..6d59122afb765 100644
--- a/mlir/test/CMakeLists.txt
+++ b/mlir/test/CMakeLists.txt
@@ -69,9 +69,9 @@ llvm_canonicalize_cmake_booleans(
LLVM_BUILD_EXAMPLES
LLVM_HAS_NVPTX_TARGET
LLVM_INCLUDE_SPIRV_TOOLS_TESTS
+ LLVM_HAS_AMDGPU_TARGET
MLIR_ENABLE_BINDINGS_PYTHON
MLIR_ENABLE_CUDA_RUNNER
- MLIR_ENABLE_ROCM_CONVERSIONS
MLIR_ENABLE_ROCM_RUNNER
MLIR_ENABLE_SPIRV_CPU_RUNNER
MLIR_ENABLE_VULKAN_RUNNER
diff --git a/mlir/test/lit.site.cfg.py.in b/mlir/test/lit.site.cfg.py.in
index 91a71aff0381c..c5bd7288dacd6 100644
--- a/mlir/test/lit.site.cfg.py.in
+++ b/mlir/test/lit.site.cfg.py.in
@@ -34,7 +34,7 @@ config.mlir_lib_dir = "@MLIR_LIB_DIR@"
config.build_examples = @LLVM_BUILD_EXAMPLES@
config.run_nvptx_tests = @LLVM_HAS_NVPTX_TARGET@
config.enable_cuda_runner = @MLIR_ENABLE_CUDA_RUNNER@
-config.run_rocm_tests = @MLIR_ENABLE_ROCM_CONVERSIONS@
+config.run_rocm_tests = @LLVM_HAS_AMDGPU_TARGET@
config.enable_rocm_runner = @MLIR_ENABLE_ROCM_RUNNER@
config.gpu_compilation_format = "@MLIR_GPU_COMPILATION_TEST_FORMAT@"
config.rocm_test_chipset = "@ROCM_TEST_CHIPSET@"
diff --git a/mlir/unittests/Target/LLVM/SerializeROCDLTarget.cpp b/mlir/unittests/Target/LLVM/SerializeROCDLTarget.cpp
index 3c71df76bb6a3..7d238494506b8 100644
--- a/mlir/unittests/Target/LLVM/SerializeROCDLTarget.cpp
+++ b/mlir/unittests/Target/LLVM/SerializeROCDLTarget.cpp
@@ -18,6 +18,7 @@
#include "mlir/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.h"
#include "mlir/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.h"
+#include "llvm/Config/Targets.h"
#include "llvm/IRReader/IRReader.h"
#include "llvm/Support/FileSystem.h"
#include "llvm/Support/MemoryBufferRef.h"
@@ -31,7 +32,7 @@
using namespace mlir;
// Skip the test if the AMDGPU target was not built.
-#if MLIR_ENABLE_ROCM_CONVERSIONS
+#if LLVM_HAS_AMDGPU_TARGET
#define SKIP_WITHOUT_AMDGPU(x) x
#else
#define SKIP_WITHOUT_AMDGPU(x) DISABLED_##x
diff --git a/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
index 3292454f49de4..bbd409666cc3c 100644
--- a/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
@@ -59,7 +59,6 @@ expand_template(
"#cmakedefine MLIR_GREEDY_REWRITE_RANDOMIZER_SEED ${MLIR_GREEDY_REWRITE_RANDOMIZER_SEED}": "/* #undef MLIR_GREEDY_REWRITE_RANDOMIZER_SEED */",
"#cmakedefine01 MLIR_ENABLE_NVPTXCOMPILER": "#define MLIR_ENABLE_NVPTXCOMPILER 0",
"#cmakedefine01 MLIR_ENABLE_PDL_IN_PATTERNMATCH": "#define MLIR_ENABLE_PDL_IN_PATTERNMATCH 1",
- "#cmakedefine01 MLIR_ENABLE_ROCM_CONVERSIONS": "#define MLIR_ENABLE_ROCM_CONVERSIONS 0",
},
template = "include/mlir/Config/mlir-config.h.cmake",
)
@@ -6592,12 +6591,15 @@ cc_library(
":Support",
":TargetLLVM",
":ToLLVMIRTranslation",
+ "//llvm:AMDGPUAsmParser",
+ "//llvm:AMDGPUCodeGen",
"//llvm:Core",
"//llvm:FrontendOffloading",
"//llvm:MC",
"//llvm:MCParser",
"//llvm:Support",
"//llvm:TargetParser",
+ "//llvm:config",
],
)
diff --git a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
index ed1500a5e1fa2..feeac1d91cfc1 100644
--- a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
@@ -41,7 +41,7 @@ expand_template(
"@LLVM_HAS_NVPTX_TARGET@": "0",
"@LLVM_INCLUDE_SPIRV_TOOLS_TESTS@": "0",
"@MLIR_ENABLE_CUDA_RUNNER@": "0",
- "@MLIR_ENABLE_ROCM_CONVERSIONS@": "0",
+ "@LLVM_HAS_AMDGPU_TARGET@": "0",
"@MLIR_ENABLE_ROCM_RUNNER@": "0",
"@MLIR_ENABLE_SYCL_RUNNER@": "0",
"@MLIR_ENABLE_LEVELZERO_RUNNER@": "0",
>From a9d5b1a9333cd9c854338e321a7432d7c51224a6 Mon Sep 17 00:00:00 2001
From: Yuansui Xu <xuyuansui at outlook.com>
Date: Thu, 12 Mar 2026 16:10:36 -0500
Subject: [PATCH 3/4] the follow up fix
---
utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel | 4 ++--
1 file changed, 2 insertions(+), 2 deletions(-)
diff --git a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
index feeac1d91cfc1..f7b871f19b193 100644
--- a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
@@ -38,10 +38,10 @@ expand_template(
"\"@MLIR_BINARY_DIR@\"": "os.environ[\"TEST_UNDECLARED_OUTPUTS_DIR\"]",
# All disabled, but required to substituted because they are not in quotes.
"@LLVM_BUILD_EXAMPLES@": "0",
- "@LLVM_HAS_NVPTX_TARGET@": "0",
+ "@LLVM_HAS_AMDGPU_TARGET@": "1" if (targets.contains("AMDGPU")) else "0",
+ "@LLVM_HAS_NVPTX_TARGET@": "1" if (targets.contains("NVPTX")) else "0",
"@LLVM_INCLUDE_SPIRV_TOOLS_TESTS@": "0",
"@MLIR_ENABLE_CUDA_RUNNER@": "0",
- "@LLVM_HAS_AMDGPU_TARGET@": "0",
"@MLIR_ENABLE_ROCM_RUNNER@": "0",
"@MLIR_ENABLE_SYCL_RUNNER@": "0",
"@MLIR_ENABLE_LEVELZERO_RUNNER@": "0",
>From e5e3095484c0d9198798273eca953c1bb7487195 Mon Sep 17 00:00:00 2001
From: Yuansui Xu <xuyuansui at outlook.com>
Date: Fri, 13 Mar 2026 15:13:17 -0500
Subject: [PATCH 4/4] rm bazel side change
---
utils/bazel/llvm-project-overlay/mlir/BUILD.bazel | 4 +---
utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel | 4 ++--
2 files changed, 3 insertions(+), 5 deletions(-)
diff --git a/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
index bbd409666cc3c..3292454f49de4 100644
--- a/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
@@ -59,6 +59,7 @@ expand_template(
"#cmakedefine MLIR_GREEDY_REWRITE_RANDOMIZER_SEED ${MLIR_GREEDY_REWRITE_RANDOMIZER_SEED}": "/* #undef MLIR_GREEDY_REWRITE_RANDOMIZER_SEED */",
"#cmakedefine01 MLIR_ENABLE_NVPTXCOMPILER": "#define MLIR_ENABLE_NVPTXCOMPILER 0",
"#cmakedefine01 MLIR_ENABLE_PDL_IN_PATTERNMATCH": "#define MLIR_ENABLE_PDL_IN_PATTERNMATCH 1",
+ "#cmakedefine01 MLIR_ENABLE_ROCM_CONVERSIONS": "#define MLIR_ENABLE_ROCM_CONVERSIONS 0",
},
template = "include/mlir/Config/mlir-config.h.cmake",
)
@@ -6591,15 +6592,12 @@ cc_library(
":Support",
":TargetLLVM",
":ToLLVMIRTranslation",
- "//llvm:AMDGPUAsmParser",
- "//llvm:AMDGPUCodeGen",
"//llvm:Core",
"//llvm:FrontendOffloading",
"//llvm:MC",
"//llvm:MCParser",
"//llvm:Support",
"//llvm:TargetParser",
- "//llvm:config",
],
)
diff --git a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
index f7b871f19b193..ed1500a5e1fa2 100644
--- a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
@@ -38,10 +38,10 @@ expand_template(
"\"@MLIR_BINARY_DIR@\"": "os.environ[\"TEST_UNDECLARED_OUTPUTS_DIR\"]",
# All disabled, but required to substituted because they are not in quotes.
"@LLVM_BUILD_EXAMPLES@": "0",
- "@LLVM_HAS_AMDGPU_TARGET@": "1" if (targets.contains("AMDGPU")) else "0",
- "@LLVM_HAS_NVPTX_TARGET@": "1" if (targets.contains("NVPTX")) else "0",
+ "@LLVM_HAS_NVPTX_TARGET@": "0",
"@LLVM_INCLUDE_SPIRV_TOOLS_TESTS@": "0",
"@MLIR_ENABLE_CUDA_RUNNER@": "0",
+ "@MLIR_ENABLE_ROCM_CONVERSIONS@": "0",
"@MLIR_ENABLE_ROCM_RUNNER@": "0",
"@MLIR_ENABLE_SYCL_RUNNER@": "0",
"@MLIR_ENABLE_LEVELZERO_RUNNER@": "0",
More information about the llvm-commits
mailing list