[mlir] Expose MLIR_CUDA_CONVERSIONS_ENABLED in mlir-config.h. #83004
Conversation
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
ingomueller-net
requested review from
ftynse,
matthias-springer,
akuegel and
jpienaar
ftynse
approved these changes
Feb 27, 2024
There was a problem hiding this comment.
As an FYI, the code relying on this does specifically: #if MLIR_CUDA_CONVERSIONS_ENABLED == 1 so you may want to define the expected value here, or update that code to use an #ifdef instead.
That macro was not defined in some cases and thus yielded warnings if compiled with `-Wundef`. In particular, they were not defined in the BUILD files, so the GPU targets were broken when built with Bazel. This commit exposes mentioned CMake variable through mlir-config.h and uses the macro that is introduced with the same name. This replaces the macro MLIR_CUDA_CONVERSIONS_ENABLED, which the CMake files previously defined manually.
ingomueller-net
force-pushed
the
missing-cuda-bazel
branch
from
a32f282
to
4c9cb9e
Compare
|
@llvm/pr-subscribers-mlir-llvm @llvm/pr-subscribers-mlir-gpu Author: Ingo Müller (ingomueller-net) ChangesThe Full diff: https://github.com/llvm/llvm-project/pull/83004.diff 9 Files Affected:
diff --git a/mlir/CMakeLists.txt b/mlir/CMakeLists.txt
index 16c898bdeb6e00..070609c94a3b38 100644
--- a/mlir/CMakeLists.txt
+++ b/mlir/CMakeLists.txt
@@ -111,8 +111,6 @@ if ("NVPTX" IN_LIST LLVM_TARGETS_TO_BUILD)
else()
set(MLIR_ENABLE_CUDA_CONVERSIONS 0)
endif()
-# TODO: we should use a config.h file like LLVM does
-add_definitions(-DMLIR_CUDA_CONVERSIONS_ENABLED=${MLIR_ENABLE_CUDA_CONVERSIONS})
# Build the ROCm conversions and run according tests if the AMDGPU backend
# is available.
diff --git a/mlir/include/mlir/Config/mlir-config.h.cmake b/mlir/include/mlir/Config/mlir-config.h.cmake
index e152a36c0ce0cf..4a7d75e2266823 100644
--- a/mlir/include/mlir/Config/mlir-config.h.cmake
+++ b/mlir/include/mlir/Config/mlir-config.h.cmake
@@ -29,4 +29,8 @@
/* If set, enables PDL usage. */
#cmakedefine01 MLIR_ENABLE_PDL_IN_PATTERNMATCH
+/* If set, enables CUDA-related features in CUDA-related transforms, pipelines,
+ and targets. */
+#cmakedefine01 MLIR_ENABLE_CUDA_CONVERSIONS
+
#endif
diff --git a/mlir/include/mlir/InitAllPasses.h b/mlir/include/mlir/InitAllPasses.h
index e28921619fe582..5d90c197a6cced 100644
--- a/mlir/include/mlir/InitAllPasses.h
+++ b/mlir/include/mlir/InitAllPasses.h
@@ -14,6 +14,7 @@
#ifndef MLIR_INITALLPASSES_H_
#define MLIR_INITALLPASSES_H_
+#include "mlir/Config/mlir-config.h"
#include "mlir/Conversion/Passes.h"
#include "mlir/Dialect/AMDGPU/Transforms/Passes.h"
#include "mlir/Dialect/Affine/Passes.h"
@@ -96,7 +97,7 @@ inline void registerAllPasses() {
bufferization::registerBufferizationPipelines();
sparse_tensor::registerSparseTensorPipelines();
tosa::registerTosaToLinalgPipelines();
-#if MLIR_CUDA_CONVERSIONS_ENABLED
+#if MLIR_ENABLE_CUDA_CONVERSIONS
gpu::registerGPUToNVVMPipeline();
#endif
}
diff --git a/mlir/lib/Dialect/GPU/Pipelines/GPUToNVVMPipeline.cpp b/mlir/lib/Dialect/GPU/Pipelines/GPUToNVVMPipeline.cpp
index 935f0deaf9c8a6..db1974ddb3773b 100644
--- a/mlir/lib/Dialect/GPU/Pipelines/GPUToNVVMPipeline.cpp
+++ b/mlir/lib/Dialect/GPU/Pipelines/GPUToNVVMPipeline.cpp
@@ -11,6 +11,7 @@
//
//===----------------------------------------------------------------------===//
+#include "mlir/Config/mlir-config.h"
#include "mlir/Conversion/AffineToStandard/AffineToStandard.h"
#include "mlir/Conversion/ArithToLLVM/ArithToLLVM.h"
#include "mlir/Conversion/FuncToLLVM/ConvertFuncToLLVMPass.h"
@@ -38,7 +39,7 @@
using namespace mlir;
-#if MLIR_CUDA_CONVERSIONS_ENABLED
+#if MLIR_ENABLE_CUDA_CONVERSIONS
namespace {
//===----------------------------------------------------------------------===//
@@ -127,4 +128,4 @@ void mlir::gpu::registerGPUToNVVMPipeline() {
buildLowerToNVVMPassPipeline);
}
-#endif // MLIR_CUDA_CONVERSIONS_ENABLED
+#endif // MLIR_ENABLE_CUDA_CONVERSIONS
diff --git a/mlir/lib/Dialect/GPU/Transforms/ModuleToBinary.cpp b/mlir/lib/Dialect/GPU/Transforms/ModuleToBinary.cpp
index 0527073da85b69..f379ea8193923d 100644
--- a/mlir/lib/Dialect/GPU/Transforms/ModuleToBinary.cpp
+++ b/mlir/lib/Dialect/GPU/Transforms/ModuleToBinary.cpp
@@ -13,6 +13,7 @@
#include "mlir/Dialect/GPU/Transforms/Passes.h"
+#include "mlir/Config/mlir-config.h"
#include "mlir/Dialect/Func/IR/FuncOps.h"
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/LLVMIR/LLVMDialect.h"
@@ -48,7 +49,7 @@ void GpuModuleToBinaryPass::getDependentDialects(
// Register all GPU related translations.
registry.insert<gpu::GPUDialect>();
registry.insert<LLVM::LLVMDialect>();
-#if MLIR_CUDA_CONVERSIONS_ENABLED == 1
+#if MLIR_ENABLE_CUDA_CONVERSIONS
registry.insert<NVVM::NVVMDialect>();
#endif
#if MLIR_ROCM_CONVERSIONS_ENABLED == 1
diff --git a/mlir/lib/Target/LLVM/NVVM/Target.cpp b/mlir/lib/Target/LLVM/NVVM/Target.cpp
index 71b15a92782ed7..d5b6645631edd6 100644
--- a/mlir/lib/Target/LLVM/NVVM/Target.cpp
+++ b/mlir/lib/Target/LLVM/NVVM/Target.cpp
@@ -13,6 +13,7 @@
#include "mlir/Target/LLVM/NVVM/Target.h"
+#include "mlir/Config/mlir-config.h"
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/LLVMIR/NVVMDialect.h"
#include "mlir/Target/LLVM/NVVM/Utils.h"
@@ -156,7 +157,7 @@ SerializeGPUModuleBase::loadBitcodeFiles(llvm::Module &module) {
return std::move(bcFiles);
}
-#if MLIR_CUDA_CONVERSIONS_ENABLED == 1
+#if MLIR_ENABLE_CUDA_CONVERSIONS
namespace {
class NVPTXSerializer : public SerializeGPUModuleBase {
public:
@@ -562,7 +563,7 @@ NVPTXSerializer::moduleToObject(llvm::Module &llvmModule) {
return compileToBinary(*serializedISA);
#endif // MLIR_NVPTXCOMPILER_ENABLED == 1
}
-#endif // MLIR_CUDA_CONVERSIONS_ENABLED == 1
+#endif // MLIR_ENABLE_CUDA_CONVERSIONS
std::optional<SmallVector<char, 0>>
NVVMTargetAttrImpl::serializeToObject(Attribute attribute, Operation *module,
@@ -574,7 +575,7 @@ NVVMTargetAttrImpl::serializeToObject(Attribute attribute, Operation *module,
module->emitError("Module must be a GPU module.");
return std::nullopt;
}
-#if MLIR_CUDA_CONVERSIONS_ENABLED == 1
+#if MLIR_ENABLE_CUDA_CONVERSIONS
NVPTXSerializer serializer(*module, cast<NVVMTargetAttr>(attribute), options);
serializer.init();
return serializer.run();
@@ -582,7 +583,7 @@ NVVMTargetAttrImpl::serializeToObject(Attribute attribute, Operation *module,
module->emitError(
"The `NVPTX` target was not built. Please enable it when building LLVM.");
return std::nullopt;
-#endif // MLIR_CUDA_CONVERSIONS_ENABLED == 1
+#endif // MLIR_ENABLE_CUDA_CONVERSIONS
}
Attribute
diff --git a/mlir/unittests/Target/LLVM/SerializeNVVMTarget.cpp b/mlir/unittests/Target/LLVM/SerializeNVVMTarget.cpp
index 26bfbd5c11e81e..06ed53152ba1fe 100644
--- a/mlir/unittests/Target/LLVM/SerializeNVVMTarget.cpp
+++ b/mlir/unittests/Target/LLVM/SerializeNVVMTarget.cpp
@@ -6,6 +6,7 @@
//
//===----------------------------------------------------------------------===//
+#include "mlir/Config/mlir-config.h"
#include "mlir/Dialect/GPU/IR/GPUDialect.h"
#include "mlir/Dialect/LLVMIR/NVVMDialect.h"
#include "mlir/IR/MLIRContext.h"
@@ -29,7 +30,7 @@
using namespace mlir;
// Skip the test if the NVPTX target was not built.
-#if MLIR_CUDA_CONVERSIONS_ENABLED == 0
+#if MLIR_ENABLE_CUDA_CONVERSIONS == 0
#define SKIP_WITHOUT_NVPTX(x) DISABLED_##x
#else
#define SKIP_WITHOUT_NVPTX(x) x
diff --git a/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
index 59ee03d9a3213f..737c8e7765f6a6 100644
--- a/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/BUILD.bazel
@@ -6,7 +6,6 @@
# The MLIR "Multi-Level Intermediate Representation" Compiler Infrastructure
load("@bazel_skylib//rules:expand_template.bzl", "expand_template")
-load("@bazel_skylib//rules:write_file.bzl", "write_file")
load(
":build_defs.bzl",
"cc_headers_only",
@@ -36,7 +35,10 @@ expand_template(
"#cmakedefine01 MLIR_ENABLE_EXPENSIVE_PATTERN_API_CHECKS": "#define MLIR_ENABLE_EXPENSIVE_PATTERN_API_CHECKS 0",
"#cmakedefine MLIR_GREEDY_REWRITE_RANDOMIZER_SEED ${MLIR_GREEDY_REWRITE_RANDOMIZER_SEED}": "/* #undef MLIR_GREEDY_REWRITE_RANDOMIZER_SEED */",
"#cmakedefine01 MLIR_ENABLE_PDL_IN_PATTERNMATCH": "#define MLIR_ENABLE_PDL_IN_PATTERNMATCH 1",
- },
+ } | if_cuda_available(
+ {"#cmakedefine01 MLIR_ENABLE_CUDA_CONVERSIONS": "#define MLIR_ENABLE_CUDA_CONVERSIONS 1"},
+ {"#cmakedefine01 MLIR_ENABLE_CUDA_CONVERSIONS": "#define MLIR_ENABLE_CUDA_CONVERSIONS 0"},
+ ),
template = "include/mlir/Config/mlir-config.h.cmake",
)
@@ -5468,7 +5470,6 @@ cc_library(
srcs = ["lib/Dialect/GPU/Pipelines/GPUToNVVMPipeline.cpp"],
hdrs = ["include/mlir/Dialect/GPU/Pipelines/Passes.h"],
includes = ["include"],
- local_defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
deps = [
":AffineToStandard",
":ArithToLLVM",
@@ -5492,6 +5493,7 @@ cc_library(
":Transforms",
":VectorToLLVM",
":VectorToSCF",
+ ":config",
],
)
@@ -5541,6 +5543,7 @@ cc_library(
":Transforms",
":VCIXToLLVMIRTranslation",
":VectorDialect",
+ ":config",
"//llvm:Core",
"//llvm:MC",
"//llvm:Support",
@@ -6176,6 +6179,7 @@ cc_library(
":NVVMToLLVMIRTranslation",
":TargetLLVM",
":ToLLVMIRTranslation",
+ ":config",
"//llvm:NVPTXCodeGen",
"//llvm:Support",
],
diff --git a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
index 68d9b23fd5643e..367c5a63bbf107 100644
--- a/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
+++ b/utils/bazel/llvm-project-overlay/mlir/test/BUILD.bazel
@@ -552,7 +552,7 @@ cc_library(
cc_library(
name = "TestTransforms",
srcs = glob(["lib/Transforms/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
":TestDialect",
@@ -579,7 +579,7 @@ cc_library(
cc_library(
name = "TestFuncToLLVM",
srcs = glob(["lib/Conversion/FuncToLLVM/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
":TestDialect",
@@ -594,7 +594,7 @@ cc_library(
cc_library(
name = "TestOneToNTypeConversion",
srcs = glob(["lib/Conversion/OneToNTypeConversion/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
":TestDialect",
@@ -653,7 +653,7 @@ cc_library(
cc_library(
name = "TestDLTI",
srcs = glob(["lib/Dialect/DLTI/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
":TestDialect",
@@ -667,7 +667,7 @@ cc_library(
cc_library(
name = "TestGPU",
srcs = glob(["lib/Dialect/GPU/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"] + if_cuda_available([
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"] + if_cuda_available([
"MLIR_GPU_TO_CUBIN_PASS_ENABLE",
]),
includes = ["lib/Dialect/Test"],
@@ -714,7 +714,7 @@ cc_library(
cc_library(
name = "TestLinalg",
srcs = glob(["lib/Dialect/Linalg/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//llvm:Support",
@@ -748,7 +748,7 @@ cc_library(
cc_library(
name = "TestLLVM",
srcs = glob(["lib/Dialect/LLVM/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:AffineToStandard",
@@ -773,7 +773,7 @@ cc_library(
cc_library(
name = "TestMath",
srcs = glob(["lib/Dialect/Math/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:ArithDialect",
@@ -790,7 +790,7 @@ cc_library(
cc_library(
name = "TestMathToVCIX",
srcs = glob(["lib/Conversion/MathToVCIX/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:ArithDialect",
@@ -807,7 +807,7 @@ cc_library(
cc_library(
name = "TestMemRef",
srcs = glob(["lib/Dialect/MemRef/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
":TestDialect",
@@ -847,7 +847,7 @@ cc_library(
cc_library(
name = "TestNVGPU",
srcs = glob(["lib/Dialect/NVGPU/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:AffineDialect",
@@ -871,7 +871,7 @@ cc_library(
cc_library(
name = "TestSCF",
srcs = glob(["lib/Dialect/SCF/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//llvm:Support",
@@ -891,7 +891,7 @@ cc_library(
cc_library(
name = "TestArith",
srcs = glob(["lib/Dialect/Arith/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:ArithDialect",
@@ -908,7 +908,7 @@ cc_library(
cc_library(
name = "TestArmSME",
srcs = glob(["lib/Dialect/ArmSME/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:ArithToArmSME",
@@ -927,7 +927,7 @@ cc_library(
cc_library(
name = "TestBufferization",
srcs = glob(["lib/Dialect/Bufferization/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:BufferizationDialect",
@@ -989,7 +989,7 @@ cc_library(
cc_library(
name = "TestFunc",
srcs = glob(["lib/Dialect/Func/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
":TestDialect",
@@ -1005,7 +1005,7 @@ cc_library(
cc_library(
name = "TestTensor",
srcs = glob(["lib/Dialect/Tensor/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:ArithDialect",
@@ -1022,7 +1022,7 @@ cc_library(
cc_library(
name = "TestVector",
srcs = glob(["lib/Dialect/Vector/*.cpp"]),
- defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"],
+ defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"],
includes = ["lib/Dialect/Test"],
deps = [
"//mlir:AffineDialect",
|
ingomueller-net
changed the title
@joker-eph raised arguments against using Note that this includes changes from I have compiled with OSS Bazel with |
|
Also, there are a few variables that have the same issue. Once we are happy with this approach, I might submit PR to apply it to them as well. Once that is done, maybe we can think about adding |
ingomueller-net
requested review from
ftynse and
joker-eph
and removed request for
akuegel
joker-eph
reviewed
Feb 27, 2024
| @@ -552,7 +552,7 @@ cc_library( | |||
| cc_library( | |||
| name = "TestTransforms", | |||
| srcs = glob(["lib/Transforms/*.cpp"]), | |||
| defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"], | |||
| defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"], | |||
There was a problem hiding this comment.
There is a discrepancy between the change in Cmake and Bazel: shouldn't remove this and rely on mlir-config.h instead?
There was a problem hiding this comment.
I need to check this tomorrow at work but my guess was that this is activated by force here for testing. Not 100% sure now if that (1) is true and (2) makes sense.
There was a problem hiding this comment.
No I think until now this matching CMake: it was defined on the command line. But now that we move it to the config.h, you removed it from the CMake definitions so I would think we should match this in bazel
There was a problem hiding this comment.
Yeah, I think you are right. The last commit makes that change.
joker-eph
approved these changes
Feb 27, 2024
This first addresses an issue raised by @joker-eph. The fixes consist of: * Adding a missing dependency to mlir-config.h in BUILD files. * Change another ´#if X == 0` test into `#if X` (with swapped branches).
ingomueller-net
commented
Feb 28, 2024
| @@ -552,7 +552,7 @@ cc_library( | |||
| cc_library( | |||
| name = "TestTransforms", | |||
| srcs = glob(["lib/Transforms/*.cpp"]), | |||
| defines = ["MLIR_CUDA_CONVERSIONS_ENABLED"], | |||
| defines = ["MLIR_ENABLE_CUDA_CONVERSIONS"], | |||
There was a problem hiding this comment.
Yeah, I think you are right. The last commit makes that change.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is a follow up of llvm#83004, which made the same change for MLIR_CUDA_CONVERSIONS_ENABLED. As the previous PR, this PR commit exposes mentioned CMake variable through mlir-config.h and uses the macro that is introduced with the same name. This replaces the macro MLIR_ROCM_CONVERSIONS_ENABLED, which the CMake files previously defined manually.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is a follow up of llvm#83004, which made the same change for MLIR_CUDA_CONVERSIONS_ENABLED. As the previous PR, this PR commit exposes mentioned CMake variable through mlir-config.h and uses the macro that is introduced with the same name. This replaces the macro MLIR_ROCM_CONVERSIONS_ENABLED, which the CMake files previously defined manually.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is another follow-up of llvm#83004. The PR replaces the macro `MLIR_GPU_TO_HSACO_PASS_ENABLE` with the more generic macro `MLIR_ENABLE_ROCM_CONVERSIONS`. Until now, the former has been defined if and only if the latter evaluated to true in CMake. However, the former was not defined when the latter evaluated to false, in which case a warning was raised if compiled with `-Wundef`. Using a single macro relies on the `#cmakedefine01` mechanism that ensures the macro is always set to either 0 or 1.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is another follow-up of llvm#83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_ENABLE_DEPRECATED_GPU_SERIALIZATION`, which the CMake files previously defined manually.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is another follow-up of llvm#83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_NVPTXCOMPILER_ENABLED`, which the CMake files previously defined manually.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is another follow-up of llvm#83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_NVPTXCOMPILER_ENABLED`, which the CMake files previously defined manually.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 5, 2024
This is another follow-up of llvm#83004, which fixed a bug due to some macros not being defined in some situations. By raising warnings on undefined macros, this kind of bug is less likely to be introduced. Similar to llvm#83004, the fix is probably adding an include to `mlir-config.h` (and potentially defining the macro there).
ingomueller-net
added a commit
that referenced
this pull request
Mar 5, 2024
This is a follow up of #83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_ROCM_CONVERSIONS_ENABLED`, which the CMake files previously defined manually.
ingomueller-net
added a commit
that referenced
this pull request
Mar 6, 2024
#84001) This is another follow-up of #83004. The PR replaces the macro `MLIR_GPU_TO_HSACO_PASS_ENABLE` with the more generic macro `MLIR_ENABLE_ROCM_CONVERSIONS`. Until now, the former has been defined if and only if the latter evaluated to true in CMake. However, the former was not defined when the latter evaluated to false, in which case a warning was raised if compiled with `-Wundef`. Using a single macro relies on the `#cmakedefine01` mechanism that ensures the macro is always set to either 0 or 1.
ingomueller-net
added a commit
that referenced
this pull request
Mar 6, 2024
This is another follow-up of #83004. `NVVM/Target.cpp` uses the macro `MLIR_NVPTXCOMPILER_ENABLED`, which is defined in `llvm-config.h` but did not include that file, yielding a warning when compiled with `-Wundef`. This PR adds the include. ~~This is another follow-up of #83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_NVPTXCOMPILER_ENABLED`, which the CMake files previously defined manually.~~
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 6, 2024
This is another follow-up of llvm#83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_ENABLE_DEPRECATED_GPU_SERIALIZATION`, which the CMake files previously defined manually.
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 6, 2024
This is another follow-up of llvm#83004, which fixed a bug due to some macros not being defined in some situations. By raising warnings on undefined macros, this kind of bug is less likely to be introduced. Similar to llvm#83004, the fix is probably adding an include to `mlir-config.h` (and potentially defining the macro there).
ingomueller-net
added a commit
to ingomueller-net/llvm-project
that referenced
this pull request
Mar 6, 2024
This is another follow-up of llvm#83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_NVPTXCOMPILER_ENABLED`, which the CMake files previously defined manually.
ingomueller-net
added a commit
that referenced
this pull request
Mar 6, 2024
#84006) This is another follow-up of #83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_ENABLE_DEPRECATED_GPU_SERIALIZATION`, which the CMake files previously defined manually.
ingomueller-net
added a commit
that referenced
this pull request
Mar 6, 2024
) This is another follow-up of #83004, which made the same change for `MLIR_CUDA_CONVERSIONS_ENABLED`. As the previous PR, this PR commit exposes mentioned CMake variable through `mlir-config.h` and uses the macro that is introduced with the same name. This replaces the macro `MLIR_NVPTXCOMPILER_ENABLED`, which the CMake files previously defined manually.
ingomueller-net
added a commit
that referenced
this pull request
Mar 6, 2024
This is another follow-up of #83004, which fixed a bug due to some macros not being defined in some situations. By raising warnings on undefined macros, this kind of bug is less likely to be introduced. Similar to #83004, the fix is probably adding an include to `mlir-config.h` (and potentially defining the macro there).
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
TheBUILDfiles of two CUDA-related build targets were missing that definition such that the corresponding MLIR target libraries did not actually work.That macro was not defined in some cases and thus yielded warnings if compiled with
-Wundef. In particular, they were not defined in the BUILD files, so the GPU targets were broken when built with Bazel. This commit exposes mentioned CMake variable through mlir-config.h and uses the macro that is introduced with the same name. This replaces the macro MLIR_CUDA_CONVERSIONS_ENABLED, which the CMake files previously defined manually.