[flang-commits] [flang] [mlir] [mlir] Support DialectRegistry extension comparison (PR #101119)
Nikhil Kalra via flang-commits
flang-commits at lists.llvm.org
Mon Aug 5 12:58:09 PDT 2024
https://github.com/nikalra updated https://github.com/llvm/llvm-project/pull/101119
>From c6532793c01b129ff7cbf6cf0db86d8ca7325069 Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Wed, 31 Jul 2024 19:24:14 -0700
Subject: [PATCH 1/8] DialectRegistry: add extension ID
---
mlir/include/mlir/IR/DialectRegistry.h | 32 ++++++++++++++++++--------
mlir/lib/IR/Dialect.cpp | 4 ++--
2 files changed, 24 insertions(+), 12 deletions(-)
diff --git a/mlir/include/mlir/IR/DialectRegistry.h b/mlir/include/mlir/IR/DialectRegistry.h
index c13a1a1858eb1..ce5a766843228 100644
--- a/mlir/include/mlir/IR/DialectRegistry.h
+++ b/mlir/include/mlir/IR/DialectRegistry.h
@@ -187,7 +187,8 @@ class DialectRegistry {
nameAndRegistrationIt.second.second);
// Merge the extensions.
for (const auto &extension : extensions)
- destination.extensions.push_back(extension->clone());
+ destination.extensions.emplace_back(extension.first,
+ extension.second->clone());
}
/// Return the names of dialects known to this registry.
@@ -206,39 +207,47 @@ class DialectRegistry {
void applyExtensions(MLIRContext *ctx) const;
/// Add the given extension to the registry.
- void addExtension(std::unique_ptr<DialectExtensionBase> extension) {
- extensions.push_back(std::move(extension));
+ void addExtension(StringRef extensionID,
+ std::unique_ptr<DialectExtensionBase> extension) {
+ extensions.emplace_back(extensionID, std::move(extension));
}
/// Add the given extensions to the registry.
template <typename... ExtensionsT>
void addExtensions() {
- (addExtension(std::make_unique<ExtensionsT>()), ...);
+ (addExtension(ExtensionsT::extensionID, std::make_unique<ExtensionsT>()),
+ ...);
}
/// Add an extension function that requires the given dialects.
/// Note: This bare functor overload is provided in addition to the
/// std::function variant to enable dialect type deduction, e.g.:
- /// registry.addExtension(+[](MLIRContext *ctx, MyDialect *dialect) { ... })
+ /// registry.addExtension("ID", +[](MLIRContext *ctx, MyDialect *dialect) {
+ /// ... })
///
/// is equivalent to:
/// registry.addExtension<MyDialect>(
+ /// "ID",
/// [](MLIRContext *ctx, MyDialect *dialect){ ... }
/// )
template <typename... DialectsT>
- void addExtension(void (*extensionFn)(MLIRContext *, DialectsT *...)) {
+ void addExtension(StringRef extensionID,
+ void (*extensionFn)(MLIRContext *, DialectsT *...)) {
addExtension<DialectsT...>(
+ extensionID,
std::function<void(MLIRContext *, DialectsT * ...)>(extensionFn));
}
template <typename... DialectsT>
void
- addExtension(std::function<void(MLIRContext *, DialectsT *...)> extensionFn) {
+ addExtension(StringRef extensionID,
+ std::function<void(MLIRContext *, DialectsT *...)> extensionFn) {
using ExtensionFnT = std::function<void(MLIRContext *, DialectsT * ...)>;
struct Extension : public DialectExtension<Extension, DialectsT...> {
Extension(const Extension &) = default;
Extension(ExtensionFnT extensionFn)
- : extensionFn(std::move(extensionFn)) {}
+ : DialectExtension<Extension, DialectsT...>(),
+ extensionFn(std::move(extensionFn)) {}
~Extension() override = default;
void apply(MLIRContext *context, DialectsT *...dialects) const final {
@@ -246,7 +255,8 @@ class DialectRegistry {
}
ExtensionFnT extensionFn;
};
- addExtension(std::make_unique<Extension>(std::move(extensionFn)));
+ addExtension(extensionID,
+ std::make_unique<Extension>(std::move(extensionFn)));
}
/// Returns true if the current registry is a subset of 'rhs', i.e. if 'rhs'
@@ -255,7 +265,9 @@ class DialectRegistry {
private:
MapTy registry;
- std::vector<std::unique_ptr<DialectExtensionBase>> extensions;
+ using KeyExtensionPair =
+ std::pair<llvm::StringRef, std::unique_ptr<DialectExtensionBase>>;
+ llvm::SmallVector<KeyExtensionPair> extensions;
};
} // namespace mlir
diff --git a/mlir/lib/IR/Dialect.cpp b/mlir/lib/IR/Dialect.cpp
index 965386681f270..0bdc34bbc5ae8 100644
--- a/mlir/lib/IR/Dialect.cpp
+++ b/mlir/lib/IR/Dialect.cpp
@@ -260,7 +260,7 @@ void DialectRegistry::applyExtensions(Dialect *dialect) const {
// Note: Additional extensions may be added while applying an extension.
for (int i = 0; i < static_cast<int>(extensions.size()); ++i)
- applyExtension(*extensions[i]);
+ applyExtension(*extensions[i].second);
}
void DialectRegistry::applyExtensions(MLIRContext *ctx) const {
@@ -287,7 +287,7 @@ void DialectRegistry::applyExtensions(MLIRContext *ctx) const {
// Note: Additional extensions may be added while applying an extension.
for (int i = 0; i < static_cast<int>(extensions.size()); ++i)
- applyExtension(*extensions[i]);
+ applyExtension(*extensions[i].second);
}
bool DialectRegistry::isSubsetOf(const DialectRegistry &rhs) const {
>From 26669bd8a53297dfe545b407d438c85d0ce32bfd Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Wed, 31 Jul 2024 19:36:25 -0700
Subject: [PATCH 2/8] attach extension IDs
---
flang/lib/Optimizer/Dialect/FIRDialect.cpp | 3 +-
.../transform/Ch2/lib/MyExtension.cpp | 3 +
.../transform/Ch3/lib/MyExtension.cpp | 3 +
.../transform/Ch4/lib/MyExtension.cpp | 3 +
.../Conversion/ArithToLLVM/ArithToLLVM.cpp | 7 +-
.../ComplexToLLVM/ComplexToLLVM.cpp | 1 +
.../ControlFlowToLLVM/ControlFlowToLLVM.cpp | 7 +-
.../ConvertToLLVM/ConvertToLLVMPass.cpp | 5 ++
mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp | 7 +-
.../Conversion/IndexToLLVM/IndexToLLVM.cpp | 7 +-
mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp | 7 +-
.../Conversion/MemRefToLLVM/MemRefToLLVM.cpp | 7 +-
mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp | 7 +-
mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp | 7 +-
.../Affine/IR/ValueBoundsOpInterfaceImpl.cpp | 11 +--
.../TransformOps/AffineTransformOps.cpp | 2 +
.../Arith/IR/ValueBoundsOpInterfaceImpl.cpp | 16 +++--
.../BufferDeallocationOpInterfaceImpl.cpp | 7 +-
.../BufferViewFlowOpInterfaceImpl.cpp | 7 +-
.../BufferizableOpInterfaceImpl.cpp | 11 +--
.../BufferizationTransformOps.cpp | 2 +
.../FuncBufferizableOpInterfaceImpl.cpp | 14 ++--
.../BufferDeallocationOpInterfaceImpl.cpp | 7 +-
.../BufferizableOpInterfaceImpl.cpp | 10 +--
.../Func/Extensions/InlinerExtension.cpp | 11 +--
.../Extensions/MeshShardingExtensions.cpp | 11 +--
.../Func/TransformOps/FuncTransformOps.cpp | 2 +
.../GPU/TransformOps/GPUTransformOps.cpp | 2 +
.../BufferDeallocationOpInterfaceImpl.cpp | 7 +-
.../Linalg/IR/ValueBoundsOpInterfaceImpl.cpp | 13 ++--
.../Linalg/TransformOps/DialectExtension.cpp | 2 +
.../BufferizableOpInterfaceImpl.cpp | 19 ++---
.../Transforms/MeshShardingInterfaceImpl.cpp | 25 +++----
.../Transforms/RuntimeOpVerification.cpp | 17 ++---
.../SubsetInsertionOpInterfaceImpl.cpp | 9 +--
.../Linalg/Transforms/TilingInterfaceImpl.cpp | 11 +--
.../BufferizableOpInterfaceImpl.cpp | 11 +--
.../Dialect/MemRef/IR/MemRefMemorySlot.cpp | 9 ++-
.../MemRef/IR/ValueBoundsOpInterfaceImpl.cpp | 25 ++++---
.../TransformOps/MemRefTransformOps.cpp | 2 +
.../Transforms/AllocationOpInterfaceImpl.cpp | 13 ++--
.../BufferViewFlowOpInterfaceImpl.cpp | 8 ++-
.../Transforms/RuntimeOpVerification.cpp | 26 +++----
.../NVGPU/TransformOps/NVGPUTransformOps.cpp | 2 +
.../SCF/IR/ValueBoundsOpInterfaceImpl.cpp | 9 +--
.../SCF/TransformOps/SCFTransformOps.cpp | 2 +
.../BufferDeallocationOpInterfaceImpl.cpp | 9 +--
.../BufferizableOpInterfaceImpl.cpp | 23 +++---
.../BufferizableOpInterfaceImpl.cpp | 10 +--
.../TransformOps/SparseTensorTransformOps.cpp | 2 +
.../BufferizableOpInterfaceImpl.cpp | 40 ++++++-----
.../IR/TensorInferTypeOpInterfaceImpl.cpp | 15 ++--
.../Tensor/IR/TensorTilingInterfaceImpl.cpp | 21 +++---
.../Tensor/IR/ValueBoundsOpInterfaceImpl.cpp | 25 ++++---
.../TransformOps/TensorTransformOps.cpp | 23 +++---
.../BufferizableOpInterfaceImpl.cpp | 46 ++++++------
.../SubsetInsertionOpInterfaceImpl.cpp | 33 ++++-----
.../Dialect/Tosa/IR/ShardingInterfaceImpl.cpp | 22 +++---
.../DebugExtension/DebugExtension.cpp | 2 +
.../Transform/IRDLExtension/IRDLExtension.cpp | 2 +
.../Transform/LoopExtension/LoopExtension.cpp | 2 +
.../Transform/PDLExtension/PDLExtension.cpp | 2 +
.../Vector/IR/ValueBoundsOpInterfaceImpl.cpp | 10 +--
.../TransformOps/VectorTransformOps.cpp | 2 +
.../BufferizableOpInterfaceImpl.cpp | 16 +++--
.../Transforms/SubsetOpInterfaceImpl.cpp | 21 +++---
mlir/lib/Interfaces/CastInterfaces.cpp | 9 +--
mlir/lib/Target/LLVM/NVVM/Target.cpp | 7 +-
mlir/lib/Target/LLVM/ROCDL/Target.cpp | 7 +-
.../Dialect/AMX/AMXToLLVMIRTranslation.cpp | 7 +-
.../ArmNeon/ArmNeonToLLVMIRTranslation.cpp | 1 +
.../ArmSME/ArmSMEToLLVMIRTranslation.cpp | 8 ++-
.../ArmSVE/ArmSVEToLLVMIRTranslation.cpp | 8 ++-
.../Builtin/BuiltinToLLVMIRTranslation.cpp | 7 +-
.../Dialect/GPU/GPUToLLVMIRTranslation.cpp | 7 +-
.../LLVMIR/Dialect/GPU/SelectObjectAttr.cpp | 7 +-
.../LLVMIR/LLVMIRToLLVMTranslation.cpp | 7 +-
.../LLVMIR/LLVMToLLVMIRTranslation.cpp | 7 +-
.../Dialect/NVVM/LLVMIRToNVVMTranslation.cpp | 7 +-
.../Dialect/NVVM/NVVMToLLVMIRTranslation.cpp | 7 +-
.../OpenACC/OpenACCToLLVMIRTranslation.cpp | 7 +-
.../OpenMP/OpenMPToLLVMIRTranslation.cpp | 7 +-
.../ROCDL/ROCDLToLLVMIRTranslation.cpp | 7 +-
.../Dialect/VCIX/VCIXToLLVMIRTranslation.cpp | 7 +-
.../X86VectorToLLVMIRTranslation.cpp | 1 +
mlir/lib/Target/SPIRV/Target.cpp | 7 +-
.../Test/TestFromLLVMIRTranslation.cpp | 1 +
.../Dialect/Test/TestToLLVMIRTranslation.cpp | 1 +
.../TestTransformDialectExtension.cpp | 2 +
.../TestTilingInterfaceTransformOps.cpp | 2 +
.../Transform/BuildOnlyExtensionTest.cpp | 2 +
mlir/unittests/IR/DialectTest.cpp | 25 +++++--
mlir/unittests/IR/InterfaceAttachmentTest.cpp | 72 +++++++++++--------
93 files changed, 567 insertions(+), 393 deletions(-)
diff --git a/flang/lib/Optimizer/Dialect/FIRDialect.cpp b/flang/lib/Optimizer/Dialect/FIRDialect.cpp
index 4b1dadaac6728..3bf8b43c3572e 100644
--- a/flang/lib/Optimizer/Dialect/FIRDialect.cpp
+++ b/flang/lib/Optimizer/Dialect/FIRDialect.cpp
@@ -74,7 +74,7 @@ void fir::FIROpsDialect::initialize() {
// Register the FIRInlinerInterface to FIROpsDialect
void fir::addFIRInlinerExtension(mlir::DialectRegistry ®istry) {
registry.addExtension(
- +[](mlir::MLIRContext *ctx, fir::FIROpsDialect *dialect) {
+ "FIR_INLINER", +[](mlir::MLIRContext *ctx, fir::FIROpsDialect *dialect) {
dialect->addInterface<FIRInlinerInterface>();
});
}
@@ -90,6 +90,7 @@ void fir::addFIRInlinerExtension(mlir::DialectRegistry ®istry) {
// when more sophisticated translation is required.
void fir::addFIRToLLVMIRExtension(mlir::DialectRegistry ®istry) {
registry.addExtension(
+ "FIR_LLVM_TRANSLATION",
+[](mlir::MLIRContext *ctx, fir::FIROpsDialect *dialect) {
dialect->addInterface<mlir::LLVMTranslationDialectInterface>();
});
diff --git a/mlir/examples/transform/Ch2/lib/MyExtension.cpp b/mlir/examples/transform/Ch2/lib/MyExtension.cpp
index 68d538a098018..a3f4f9aeb8106 100644
--- a/mlir/examples/transform/Ch2/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch2/lib/MyExtension.cpp
@@ -36,6 +36,9 @@ class MyExtension
// dialect definitions. List individual operations and dependent dialects
// here.
void init();
+
+ // Declare a unique ID for this extension.
+ static constexpr llvm::StringRef extensionID = "CH2_TRANSFORM";
};
void MyExtension::init() {
diff --git a/mlir/examples/transform/Ch3/lib/MyExtension.cpp b/mlir/examples/transform/Ch3/lib/MyExtension.cpp
index f7a99423a52ee..0eeeb3b4071d3 100644
--- a/mlir/examples/transform/Ch3/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch3/lib/MyExtension.cpp
@@ -42,6 +42,9 @@ class MyExtension
// dialect definitions. List individual operations and dependent dialects
// here.
void init();
+
+ // Declare a unique ID for this extension.
+ static constexpr llvm::StringRef extensionID = "CH3_TRANSFORM";
};
void MyExtension::init() {
diff --git a/mlir/examples/transform/Ch4/lib/MyExtension.cpp b/mlir/examples/transform/Ch4/lib/MyExtension.cpp
index 38c8ca1125a24..1d440f588e60b 100644
--- a/mlir/examples/transform/Ch4/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch4/lib/MyExtension.cpp
@@ -38,6 +38,9 @@ class MyExtension
// dialect definitions. List individual operations and dependent dialects
// here.
void init();
+
+ // Declare a unique ID for this extension.
+ static constexpr llvm::StringRef extensionID = "CH4_TRANSFORM";
};
void MyExtension::init() {
diff --git a/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp b/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp
index d882f1184f457..d2794a24cd05a 100644
--- a/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp
+++ b/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp
@@ -510,9 +510,10 @@ struct ArithToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
void mlir::arith::registerConvertArithToLLVMInterface(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, arith::ArithDialect *dialect) {
- dialect->addInterfaces<ArithToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "ARITH_TO_LLVM", +[](MLIRContext *ctx, arith::ArithDialect *dialect) {
+ dialect->addInterfaces<ArithToLLVMDialectInterface>();
+ });
}
//===----------------------------------------------------------------------===//
diff --git a/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp b/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp
index 0a3c3a330ff69..3701d62efba50 100644
--- a/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp
+++ b/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp
@@ -385,6 +385,7 @@ struct ComplexToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
void mlir::registerConvertComplexToLLVMInterface(DialectRegistry ®istry) {
registry.addExtension(
+ "COMPLEX_TO_LLVM",
+[](MLIRContext *ctx, complex::ComplexDialect *dialect) {
dialect->addInterfaces<ComplexToLLVMDialectInterface>();
});
diff --git a/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp b/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp
index b8e5aec25286d..4a154554c30bc 100644
--- a/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp
+++ b/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp
@@ -273,7 +273,8 @@ struct ControlFlowToLLVMDialectInterface
void mlir::cf::registerConvertControlFlowToLLVMInterface(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
- dialect->addInterfaces<ControlFlowToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "CF_TO_LLVM", +[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
+ dialect->addInterfaces<ControlFlowToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp b/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
index 6135117348a5b..c55607f5e7c60 100644
--- a/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
+++ b/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
@@ -15,6 +15,7 @@
#include "mlir/Pass/Pass.h"
#include "mlir/Rewrite/FrozenRewritePatternSet.h"
#include "mlir/Transforms/DialectConversion.h"
+#include "llvm/ADT/StringRef.h"
#include <memory>
#define DEBUG_TYPE "convert-to-llvm"
@@ -54,6 +55,10 @@ class LoadDependentDialectExtension : public DialectExtensionBase {
std::unique_ptr<DialectExtensionBase> clone() const final {
return std::make_unique<LoadDependentDialectExtension>(*this);
}
+
+ /// Unique ID
+ constexpr static llvm::StringRef extensionID =
+ "CONVERT_TO_LLVM_LOAD_DEPENDENT";
};
/// This is a generic pass to convert to LLVM, it uses the
diff --git a/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp b/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp
index c1f6d8bc5b361..5994a4c5fdf0f 100644
--- a/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp
+++ b/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp
@@ -786,7 +786,8 @@ struct FuncToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::registerConvertFuncToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, func::FuncDialect *dialect) {
- dialect->addInterfaces<FuncToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "FUNC_TO_LLVM", +[](MLIRContext *ctx, func::FuncDialect *dialect) {
+ dialect->addInterfaces<FuncToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp b/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp
index 9d8a5d8a0e1c0..e045605475b78 100644
--- a/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp
+++ b/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp
@@ -392,7 +392,8 @@ struct IndexToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
void mlir::index::registerConvertIndexToLLVMInterface(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, index::IndexDialect *dialect) {
- dialect->addInterfaces<IndexToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "INDEX_TO_LLVM", +[](MLIRContext *ctx, index::IndexDialect *dialect) {
+ dialect->addInterfaces<IndexToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp b/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp
index 23e957288eb95..f43ce2800e310 100644
--- a/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp
+++ b/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp
@@ -356,7 +356,8 @@ struct MathToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::registerConvertMathToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, math::MathDialect *dialect) {
- dialect->addInterfaces<MathToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "MATH_TO_LLVM", +[](MLIRContext *ctx, math::MathDialect *dialect) {
+ dialect->addInterfaces<MathToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp b/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp
index 054827d40f0f3..d0d4b159ddd57 100644
--- a/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp
+++ b/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp
@@ -1753,7 +1753,8 @@ struct MemRefToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::registerConvertMemRefToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- dialect->addInterfaces<MemRefToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "MEMREF_TO_LLVM", +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ dialect->addInterfaces<MemRefToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp b/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp
index 662ee9e483bc5..c6b856b84b10b 100644
--- a/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp
+++ b/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp
@@ -113,7 +113,8 @@ void mlir::populateNVVMToLLVMConversionPatterns(RewritePatternSet &patterns) {
}
void mlir::registerConvertNVVMToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, NVVMDialect *dialect) {
- dialect->addInterfaces<NVVMToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "NVVM_TO_LLVM", +[](MLIRContext *ctx, NVVMDialect *dialect) {
+ dialect->addInterfaces<NVVMToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp b/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp
index 0051333a35dcd..12150d2e1977b 100644
--- a/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp
+++ b/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp
@@ -119,7 +119,8 @@ struct UBToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::ub::registerConvertUBToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, ub::UBDialect *dialect) {
- dialect->addInterfaces<UBToLLVMDialectInterface>();
- });
+ registry.addExtension(
+ "UB_TO_LLVM", +[](MLIRContext *ctx, ub::UBDialect *dialect) {
+ dialect->addInterfaces<UBToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp
index 82a9fb0d49088..2fe39c757fcd3 100644
--- a/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -96,11 +96,12 @@ struct AffineMaxOpInterface
void mlir::affine::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, AffineDialect *dialect) {
- AffineApplyOp::attachInterface<AffineApplyOpInterface>(*ctx);
- AffineMaxOp::attachInterface<AffineMaxOpInterface>(*ctx);
- AffineMinOp::attachInterface<AffineMinOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "AFFINE_VALUE_BOUNDS", +[](MLIRContext *ctx, AffineDialect *dialect) {
+ AffineApplyOp::attachInterface<AffineApplyOpInterface>(*ctx);
+ AffineMaxOp::attachInterface<AffineMaxOpInterface>(*ctx);
+ AffineMinOp::attachInterface<AffineMinOpInterface>(*ctx);
+ });
}
FailureOr<int64_t>
diff --git a/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp b/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
index 6457655cfe416..c6f5cfd47b046 100644
--- a/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
+++ b/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
@@ -167,6 +167,8 @@ class AffineTransformDialectExtension
#include "mlir/Dialect/Affine/TransformOps/AffineTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "AFFINE_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp
index 7cfcc4180539c..981a52090ad66 100644
--- a/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -150,11 +150,13 @@ struct SelectOpInterface
void mlir::arith::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, arith::ArithDialect *dialect) {
- arith::AddIOp::attachInterface<arith::AddIOpInterface>(*ctx);
- arith::ConstantOp::attachInterface<arith::ConstantOpInterface>(*ctx);
- arith::SubIOp::attachInterface<arith::SubIOpInterface>(*ctx);
- arith::MulIOp::attachInterface<arith::MulIOpInterface>(*ctx);
- arith::SelectOp::attachInterface<arith::SelectOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "ARITH_VALUE_BOUNDS",
+ +[](MLIRContext *ctx, arith::ArithDialect *dialect) {
+ arith::AddIOp::attachInterface<arith::AddIOpInterface>(*ctx);
+ arith::ConstantOp::attachInterface<arith::ConstantOpInterface>(*ctx);
+ arith::SubIOp::attachInterface<arith::SubIOpInterface>(*ctx);
+ arith::MulIOp::attachInterface<arith::MulIOpInterface>(*ctx);
+ arith::SelectOp::attachInterface<arith::SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index f2e7732e8ea4a..c17f6aacd88f8 100644
--- a/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -79,7 +79,8 @@ struct SelectOpInterface
void mlir::arith::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, ArithDialect *dialect) {
- SelectOp::attachInterface<SelectOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "ARITH_BUFFER_DEALLOC", +[](MLIRContext *ctx, ArithDialect *dialect) {
+ SelectOp::attachInterface<SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp
index 9df9df86b64fb..9f03fa947085f 100644
--- a/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp
@@ -38,7 +38,8 @@ struct SelectOpInterface
void arith::registerBufferViewFlowOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, arith::ArithDialect *dialect) {
- SelectOp::attachInterface<SelectOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "ARITH_BUFFER_FLOW", +[](MLIRContext *ctx, arith::ArithDialect *dialect) {
+ SelectOp::attachInterface<SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp
index 5e69a98db8f1e..74b55c124e480 100644
--- a/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -206,9 +206,10 @@ struct SelectOpInterface
void mlir::arith::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, ArithDialect *dialect) {
- ConstantOp::attachInterface<ConstantOpInterface>(*ctx);
- IndexCastOp::attachInterface<IndexCastOpInterface>(*ctx);
- SelectOp::attachInterface<SelectOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "ARITH_BUFFERIZATION", +[](MLIRContext *ctx, ArithDialect *dialect) {
+ ConstantOp::attachInterface<ConstantOpInterface>(*ctx);
+ IndexCastOp::attachInterface<IndexCastOpInterface>(*ctx);
+ SelectOp::attachInterface<SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp b/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
index e10c7bd914e35..d0993a5b85e8d 100644
--- a/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
+++ b/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
@@ -161,6 +161,8 @@ class BufferizationTransformDialectExtension
#include "mlir/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "BUFFERIZATION_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp
index 053ea7935260a..6f2ec90bc866e 100644
--- a/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp
@@ -82,7 +82,8 @@ getBufferizedFunctionArgType(FuncOp funcOp, int64_t index,
/// Return the FuncOp called by `callOp`.
static FuncOp getCalledFunction(CallOpInterface callOp) {
- SymbolRefAttr sym = llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
+ SymbolRefAttr sym =
+ llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
if (!sym)
return nullptr;
return dyn_cast_or_null<FuncOp>(
@@ -489,9 +490,10 @@ struct FuncOpInterface
void mlir::bufferization::func_ext::
registerBufferizableOpInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, func::FuncDialect *dialect) {
- func::CallOp::attachInterface<func_ext::CallOpInterface>(*ctx);
- func::FuncOp::attachInterface<func_ext::FuncOpInterface>(*ctx);
- func::ReturnOp::attachInterface<func_ext::ReturnOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "FUNC_BUFFERIZATION", +[](MLIRContext *ctx, func::FuncDialect *dialect) {
+ func::CallOp::attachInterface<func_ext::CallOpInterface>(*ctx);
+ func::FuncOp::attachInterface<func_ext::FuncOpInterface>(*ctx);
+ func::ReturnOp::attachInterface<func_ext::ReturnOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index 89546da428fa2..f58f21b35547d 100644
--- a/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -157,7 +157,8 @@ struct CondBranchOpInterface
void mlir::cf::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, ControlFlowDialect *dialect) {
- CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "CF_BUFFER_DEALLOC", +[](MLIRContext *ctx, ControlFlowDialect *dialect) {
+ CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp
index 72f4a1a4f4c66..0db95f92dd95b 100644
--- a/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -63,8 +63,10 @@ struct CondBranchOpInterface
void mlir::cf::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
- cf::BranchOp::attachInterface<BranchOpInterface>(*ctx);
- cf::CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "CF_BUFFERIZATION",
+ +[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
+ cf::BranchOp::attachInterface<BranchOpInterface>(*ctx);
+ cf::CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp b/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp
index 719a74a29a622..43e998fbc04f8 100644
--- a/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp
+++ b/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp
@@ -80,10 +80,11 @@ struct FuncInlinerInterface : public DialectInlinerInterface {
//===----------------------------------------------------------------------===//
void mlir::func::registerInlinerExtension(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, func::FuncDialect *dialect) {
- dialect->addInterfaces<FuncInlinerInterface>();
+ registry.addExtension(
+ "FUNC_INLINER", +[](MLIRContext *ctx, func::FuncDialect *dialect) {
+ dialect->addInterfaces<FuncInlinerInterface>();
- // The inliner extension relies on the ControlFlow dialect.
- ctx->getOrLoadDialect<cf::ControlFlowDialect>();
- });
+ // The inliner extension relies on the ControlFlow dialect.
+ ctx->getOrLoadDialect<cf::ControlFlowDialect>();
+ });
}
diff --git a/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp b/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp
index da508cc95bfe1..833362041848b 100644
--- a/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp
+++ b/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp
@@ -14,11 +14,12 @@
namespace mlir::func {
void registerShardingInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, FuncDialect *dialect) {
- ReturnOp::attachInterface<
- mesh::IndependentParallelIteratorDomainShardingInterface<ReturnOp>>(
- *ctx);
- });
+ registry.addExtension(
+ "FUNC_SHARDING", +[](MLIRContext *ctx, FuncDialect *dialect) {
+ ReturnOp::attachInterface<
+ mesh::IndependentParallelIteratorDomainShardingInterface<ReturnOp>>(
+ *ctx);
+ });
}
} // namespace mlir::func
diff --git a/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp b/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
index b632b25d0cc67..711a2d9685f6d 100644
--- a/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
+++ b/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
@@ -246,6 +246,8 @@ class FuncTransformDialectExtension
#include "mlir/Dialect/Func/TransformOps/FuncTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "FUNC_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp b/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
index 3661c5dea4525..d6a044698746f 100644
--- a/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
+++ b/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
@@ -933,6 +933,8 @@ class GPUTransformDialectExtension
#include "mlir/Dialect/GPU/TransformOps/GPUTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "GPU_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index 6ccc0a26426c1..ddad79f173628 100644
--- a/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -31,7 +31,8 @@ struct GPUTerminatorOpInterface
void mlir::gpu::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, GPUDialect *dialect) {
- gpu::TerminatorOp::attachInterface<GPUTerminatorOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "GPU_BUFFER_DEALLOC", +[](MLIRContext *ctx, GPUDialect *dialect) {
+ gpu::TerminatorOp::attachInterface<GPUTerminatorOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp
index f56ef485069f8..390d6582c487e 100644
--- a/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -53,9 +53,12 @@ struct IndexOpInterface
void mlir::linalg::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- IndexOp::attachInterface<IndexOpInterface>(*ctx);
- // Note: ValueBoundsOpInterface implementation is not required for ops that
- // implement `DestinationStyleOpInterface` (for querying shaped OpResults).
- });
+ registry.addExtension(
+ "LINALG_VALUE_BOUNDS",
+ +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ IndexOp::attachInterface<IndexOpInterface>(*ctx);
+ // Note: ValueBoundsOpInterface implementation is not required for ops
+ // that implement `DestinationStyleOpInterface` (for querying shaped
+ // OpResults).
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp b/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
index f4244ca962232..01c6814bc2082 100644
--- a/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
+++ b/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
@@ -52,6 +52,8 @@ class LinalgTransformDialectExtension
#include "mlir/Dialect/Linalg/TransformOps/LinalgMatchOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "LINALG_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp
index be158af09d398..dc3bea56da8de 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -195,15 +195,18 @@ struct SoftmaxOpInterface
void mlir::linalg::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- // Register all Linalg structured ops. `LinalgOp` is an interface and it is
- // not possible to attach an external interface to an existing interface.
- // Therefore, attach the `BufferizableOpInterface` to all ops one-by-one.
- LinalgOpInterfaceHelper<
+ registry.addExtension(
+ "LINALG_BUFFERIZATION",
+ +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ // Register all Linalg structured ops. `LinalgOp` is an interface and it
+ // is not possible to attach an external interface to an existing
+ // interface. Therefore, attach the `BufferizableOpInterface` to all ops
+ // one-by-one.
+ LinalgOpInterfaceHelper<
#define GET_OP_LIST
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >::registerOpInterface(ctx);
+ >::registerOpInterface(ctx);
- SoftmaxOp::attachInterface<SoftmaxOpInterface>(*ctx);
- });
+ SoftmaxOp::attachInterface<SoftmaxOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp
index 36b6088b83cc2..afb5ffcf31613 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp
@@ -347,20 +347,21 @@ static void registerAll(MLIRContext *ctx) {
}
void registerMeshShardingInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, LinalgDialect *dialect) {
- DialectRegistry registry;
- registry.insert<affine::AffineDialect, arith::ArithDialect, scf::SCFDialect,
- tensor::TensorDialect>();
- ctx->appendDialectRegistry(registry);
- for (StringRef name : registry.getDialectNames())
- ctx->getOrLoadDialect(name);
-
- registerOne<linalg::GenericOp>(ctx);
- registerAll<
+ registry.addExtension(
+ "LINALG_SHARDING", +[](MLIRContext *ctx, LinalgDialect *dialect) {
+ DialectRegistry registry;
+ registry.insert<affine::AffineDialect, arith::ArithDialect,
+ scf::SCFDialect, tensor::TensorDialect>();
+ ctx->appendDialectRegistry(registry);
+ for (StringRef name : registry.getDialectNames())
+ ctx->getOrLoadDialect(name);
+
+ registerOne<linalg::GenericOp>(ctx);
+ registerAll<
#define GET_OP_LIST
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >(ctx);
- });
+ >(ctx);
+ });
}
} // namespace mlir::linalg
diff --git a/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp b/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp
index b30182dc84079..6b3ab1a42bd99 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp
@@ -121,15 +121,16 @@ void attachInterface(MLIRContext *ctx) {
void mlir::linalg::registerRuntimeVerifiableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, LinalgDialect *) {
- attachInterface<
+ registry.addExtension(
+ "LINALG_RUNTIME_VERIFICATION", +[](MLIRContext *ctx, LinalgDialect *) {
+ attachInterface<
#define GET_OP_LIST
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >(ctx);
+ >(ctx);
- // Load additional dialects of which ops may get created.
- ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
- cf::ControlFlowDialect, index::IndexDialect,
- tensor::TensorDialect>();
- });
+ // Load additional dialects of which ops may get created.
+ ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
+ cf::ControlFlowDialect, index::IndexDialect,
+ tensor::TensorDialect>();
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp
index 6fcfa05468eea..04e4b86f8a855 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp
@@ -74,8 +74,9 @@ struct LinalgCopyOpInterface
void mlir::linalg::registerSubsetOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- linalg::CopyOp::attachInterface<LinalgCopyOpSubsetOpInterface>(*ctx);
- linalg::CopyOp::attachInterface<LinalgCopyOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "LINALG_SUBSET", +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ linalg::CopyOp::attachInterface<LinalgCopyOpSubsetOpInterface>(*ctx);
+ linalg::CopyOp::attachInterface<LinalgCopyOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp
index 2133458efe74c..9cd4b4c6adbd9 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp
@@ -506,10 +506,11 @@ static void registerAll(MLIRContext *ctx) {
void mlir::linalg::registerTilingInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- registerOne<linalg::GenericOp>(ctx);
- registerAll<
+ registry.addExtension(
+ "LINALG_TILING", +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ registerOne<linalg::GenericOp>(ctx);
+ registerAll<
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >(ctx);
- });
+ >(ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp
index 926d580ac7852..a07a412f11e85 100644
--- a/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -149,11 +149,12 @@ struct GlobalStoreOpInterface
} // namespace
void registerBufferizableOpInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, MLProgramDialect *) {
- GlobalOp::attachInterface<GlobalOpInterface>(*ctx);
- GlobalLoadOp::attachInterface<GlobalLoadOpInterface>(*ctx);
- GlobalStoreOp::attachInterface<GlobalStoreOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "MLPROGRAM_BUFFERIZATION", +[](MLIRContext *ctx, MLProgramDialect *) {
+ GlobalOp::attachInterface<GlobalOpInterface>(*ctx);
+ GlobalLoadOp::attachInterface<GlobalLoadOpInterface>(*ctx);
+ GlobalStoreOp::attachInterface<GlobalStoreOpInterface>(*ctx);
+ });
}
} // namespace ml_program
} // namespace mlir
diff --git a/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp b/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp
index f630c48cdcaa1..7c21858adc5fb 100644
--- a/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp
+++ b/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp
@@ -347,7 +347,10 @@ struct MemRefDestructurableTypeExternalModel
//===----------------------------------------------------------------------===//
void mlir::memref::registerMemorySlotExternalModels(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- MemRefType::attachInterface<MemRefDestructurableTypeExternalModel>(*ctx);
- });
+ registry.addExtension(
+ "MEMREF_MEMORY_SLOT_MODELS",
+ +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ MemRefType::attachInterface<MemRefDestructurableTypeExternalModel>(
+ *ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp
index daec22cf6ebdc..fde96046b3ca2 100644
--- a/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -115,15 +115,18 @@ struct SubViewOpInterface
void mlir::memref::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- memref::AllocOp::attachInterface<memref::AllocOpInterface<memref::AllocOp>>(
- *ctx);
- memref::AllocaOp::attachInterface<
- memref::AllocOpInterface<memref::AllocaOp>>(*ctx);
- memref::CastOp::attachInterface<memref::CastOpInterface>(*ctx);
- memref::DimOp::attachInterface<memref::DimOpInterface>(*ctx);
- memref::GetGlobalOp::attachInterface<memref::GetGlobalOpInterface>(*ctx);
- memref::RankOp::attachInterface<memref::RankOpInterface>(*ctx);
- memref::SubViewOp::attachInterface<memref::SubViewOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "MEMREF_VALUE_BOUNDS",
+ +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ memref::AllocOp::attachInterface<
+ memref::AllocOpInterface<memref::AllocOp>>(*ctx);
+ memref::AllocaOp::attachInterface<
+ memref::AllocOpInterface<memref::AllocaOp>>(*ctx);
+ memref::CastOp::attachInterface<memref::CastOpInterface>(*ctx);
+ memref::DimOp::attachInterface<memref::DimOpInterface>(*ctx);
+ memref::GetGlobalOp::attachInterface<memref::GetGlobalOpInterface>(
+ *ctx);
+ memref::RankOp::attachInterface<memref::RankOpInterface>(*ctx);
+ memref::SubViewOp::attachInterface<memref::SubViewOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp b/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
index 8469e84c668cb..20f96a018d330 100644
--- a/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
+++ b/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
@@ -323,6 +323,8 @@ class MemRefTransformDialectExtension
#include "mlir/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "MEMREF_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp
index c433415944323..2cde261180e9e 100644
--- a/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp
@@ -60,10 +60,11 @@ struct DefaultReallocationInterface
void mlir::memref::registerAllocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- memref::AllocOp::attachInterface<DefaultAllocationInterface>(*ctx);
- memref::AllocaOp::attachInterface<
- DefaultAutomaticAllocationHoistingInterface>(*ctx);
- memref::ReallocOp::attachInterface<DefaultReallocationInterface>(*ctx);
- });
+ registry.addExtension(
+ "MEMREF_ALLOC", +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ memref::AllocOp::attachInterface<DefaultAllocationInterface>(*ctx);
+ memref::AllocaOp::attachInterface<
+ DefaultAutomaticAllocationHoistingInterface>(*ctx);
+ memref::ReallocOp::attachInterface<DefaultReallocationInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp b/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp
index bbb269bd00161..5076c0be916c9 100644
--- a/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp
@@ -42,7 +42,9 @@ struct ReallocOpInterface
void memref::registerBufferViewFlowOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- ReallocOp::attachInterface<ReallocOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "MEMREF_BUFFER_FLOW",
+ +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ ReallocOp::attachInterface<ReallocOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp b/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp
index 450bfa0cec0c7..80fc16d058b41 100644
--- a/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp
+++ b/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp
@@ -333,16 +333,18 @@ struct ExpandShapeOpInterface
void mlir::memref::registerRuntimeVerifiableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- CastOp::attachInterface<CastOpInterface>(*ctx);
- ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
- LoadOp::attachInterface<LoadStoreOpInterface<LoadOp>>(*ctx);
- ReinterpretCastOp::attachInterface<ReinterpretCastOpInterface>(*ctx);
- StoreOp::attachInterface<LoadStoreOpInterface<StoreOp>>(*ctx);
- SubViewOp::attachInterface<SubViewOpInterface>(*ctx);
-
- // Load additional dialects of which ops may get created.
- ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
- cf::ControlFlowDialect>();
- });
+ registry.addExtension(
+ "MEMREF_RUNTIME_VERIFICATION",
+ +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ CastOp::attachInterface<CastOpInterface>(*ctx);
+ ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
+ LoadOp::attachInterface<LoadStoreOpInterface<LoadOp>>(*ctx);
+ ReinterpretCastOp::attachInterface<ReinterpretCastOpInterface>(*ctx);
+ StoreOp::attachInterface<LoadStoreOpInterface<StoreOp>>(*ctx);
+ SubViewOp::attachInterface<SubViewOpInterface>(*ctx);
+
+ // Load additional dialects of which ops may get created.
+ ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
+ cf::ControlFlowDialect>();
+ });
}
diff --git a/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp b/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
index 7d3d868b326c6..8315ad2f5892b 100644
--- a/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
+++ b/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
@@ -1146,6 +1146,8 @@ class NVGPUTransformDialectExtension
#include "mlir/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "NVGPU_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp
index 17a1c016ea16d..c6ede24e605df 100644
--- a/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -159,8 +159,9 @@ struct IfOpInterface
void mlir::scf::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, scf::SCFDialect *dialect) {
- scf::ForOp::attachInterface<scf::ForOpInterface>(*ctx);
- scf::IfOp::attachInterface<scf::IfOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "SCF_VALUE_BOUNDS", +[](MLIRContext *ctx, scf::SCFDialect *dialect) {
+ scf::ForOp::attachInterface<scf::ForOpInterface>(*ctx);
+ scf::IfOp::attachInterface<scf::IfOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp b/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
index c4a55c302d0a3..c56d1e46bc0fd 100644
--- a/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
+++ b/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
@@ -624,6 +624,8 @@ class SCFTransformDialectExtension
#include "mlir/Dialect/SCF/TransformOps/SCFTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "SCF_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index 24fbc1dca8361..bea0728e16987 100644
--- a/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -75,8 +75,9 @@ struct ReduceReturnOpInterface
void mlir::scf::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, SCFDialect *dialect) {
- InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
- ReduceReturnOp::attachInterface<ReduceReturnOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "SCF_BUFFER_DEALLOC", +[](MLIRContext *ctx, SCFDialect *dialect) {
+ InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
+ ReduceReturnOp::attachInterface<ReduceReturnOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp
index cf40443ff3839..0aa9fd4385bb5 100644
--- a/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -1352,15 +1352,16 @@ struct InParallelOpInterface
void mlir::scf::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, scf::SCFDialect *dialect) {
- ConditionOp::attachInterface<ConditionOpInterface>(*ctx);
- ExecuteRegionOp::attachInterface<ExecuteRegionOpInterface>(*ctx);
- ForOp::attachInterface<ForOpInterface>(*ctx);
- IfOp::attachInterface<IfOpInterface>(*ctx);
- IndexSwitchOp::attachInterface<IndexSwitchOpInterface>(*ctx);
- ForallOp::attachInterface<ForallOpInterface>(*ctx);
- InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
- WhileOp::attachInterface<WhileOpInterface>(*ctx);
- YieldOp::attachInterface<YieldOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "SCF_BUFFERIZATION", +[](MLIRContext *ctx, scf::SCFDialect *dialect) {
+ ConditionOp::attachInterface<ConditionOpInterface>(*ctx);
+ ExecuteRegionOp::attachInterface<ExecuteRegionOpInterface>(*ctx);
+ ForOp::attachInterface<ForOpInterface>(*ctx);
+ IfOp::attachInterface<IfOpInterface>(*ctx);
+ IndexSwitchOp::attachInterface<IndexSwitchOpInterface>(*ctx);
+ ForallOp::attachInterface<ForallOpInterface>(*ctx);
+ InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
+ WhileOp::attachInterface<WhileOpInterface>(*ctx);
+ YieldOp::attachInterface<YieldOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp
index 66a2e45001781..b82c6624d0c32 100644
--- a/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -137,8 +137,10 @@ struct AssumingYieldOpInterface
void mlir::shape::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, shape::ShapeDialect *dialect) {
- shape::AssumingOp::attachInterface<AssumingOpInterface>(*ctx);
- shape::AssumingYieldOp::attachInterface<AssumingYieldOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "SHAPE_BUFFERIZATION",
+ +[](MLIRContext *ctx, shape::ShapeDialect *dialect) {
+ shape::AssumingOp::attachInterface<AssumingOpInterface>(*ctx);
+ shape::AssumingYieldOp::attachInterface<AssumingYieldOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp b/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
index ca19259ebffa6..e2b0c77f5f5a0 100644
--- a/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
+++ b/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
@@ -45,6 +45,8 @@ class SparseTensorTransformDialectExtension
#include "mlir/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "SPARSE_TENSOR_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp
index 7734d1d258453..aca1a13a8b913 100644
--- a/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -326,22 +326,26 @@ struct ToValuesOpInterface
void mlir::sparse_tensor::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx,
- sparse_tensor::SparseTensorDialect *dialect) {
- sparse_tensor::ConcatenateOp::attachInterface<ConcatenateOpInterface>(*ctx);
- sparse_tensor::ConvertOp::attachInterface<ConvertOpInterface>(*ctx);
- sparse_tensor::LoadOp::attachInterface<LoadOpInterface>(*ctx);
- sparse_tensor::NewOp::attachInterface<NewOpInterface>(*ctx);
- sparse_tensor::NumberOfEntriesOp::attachInterface<
- NumberOfEntriesOpInterface>(*ctx);
- sparse_tensor::AssembleOp::attachInterface<AssembleOpInterface>(*ctx);
- sparse_tensor::DisassembleOp::attachInterface<DisassembleOpInterface>(*ctx);
- sparse_tensor::ForeachOp::attachInterface<ForeachOpInterface>(*ctx);
- sparse_tensor::ToCoordinatesBufferOp::attachInterface<
- ToCoordinatesBufferOpInterface>(*ctx);
- sparse_tensor::ToCoordinatesOp::attachInterface<ToCoordinatesOpInterface>(
- *ctx);
- sparse_tensor::ToPositionsOp::attachInterface<ToPositionsOpInterface>(*ctx);
- sparse_tensor::ToValuesOp::attachInterface<ToValuesOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "SPARSE_TENSOR_BUFFERIZATION",
+ +[](MLIRContext *ctx, sparse_tensor::SparseTensorDialect *dialect) {
+ sparse_tensor::ConcatenateOp::attachInterface<ConcatenateOpInterface>(
+ *ctx);
+ sparse_tensor::ConvertOp::attachInterface<ConvertOpInterface>(*ctx);
+ sparse_tensor::LoadOp::attachInterface<LoadOpInterface>(*ctx);
+ sparse_tensor::NewOp::attachInterface<NewOpInterface>(*ctx);
+ sparse_tensor::NumberOfEntriesOp::attachInterface<
+ NumberOfEntriesOpInterface>(*ctx);
+ sparse_tensor::AssembleOp::attachInterface<AssembleOpInterface>(*ctx);
+ sparse_tensor::DisassembleOp::attachInterface<DisassembleOpInterface>(
+ *ctx);
+ sparse_tensor::ForeachOp::attachInterface<ForeachOpInterface>(*ctx);
+ sparse_tensor::ToCoordinatesBufferOp::attachInterface<
+ ToCoordinatesBufferOpInterface>(*ctx);
+ sparse_tensor::ToCoordinatesOp::attachInterface<
+ ToCoordinatesOpInterface>(*ctx);
+ sparse_tensor::ToPositionsOp::attachInterface<ToPositionsOpInterface>(
+ *ctx);
+ sparse_tensor::ToValuesOp::attachInterface<ToValuesOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp
index 7ff435a033985..48b1dc2bdaa6c 100644
--- a/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp
@@ -201,11 +201,12 @@ struct ReifyPadOp
void mlir::tensor::registerInferTypeOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, TensorDialect *dialect) {
- ExpandShapeOp::attachInterface<
- ReifyExpandOrCollapseShapeOp<tensor::ExpandShapeOp>>(*ctx);
- CollapseShapeOp::attachInterface<
- ReifyExpandOrCollapseShapeOp<tensor::CollapseShapeOp>>(*ctx);
- PadOp::attachInterface<ReifyPadOp>(*ctx);
- });
+ registry.addExtension(
+ "TENSOR_INFER_TYPE", +[](MLIRContext *ctx, TensorDialect *dialect) {
+ ExpandShapeOp::attachInterface<
+ ReifyExpandOrCollapseShapeOp<tensor::ExpandShapeOp>>(*ctx);
+ CollapseShapeOp::attachInterface<
+ ReifyExpandOrCollapseShapeOp<tensor::CollapseShapeOp>>(*ctx);
+ PadOp::attachInterface<ReifyPadOp>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp
index 9b2a97eb2b006..88801be553148 100644
--- a/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp
@@ -785,17 +785,20 @@ FailureOr<TilingResult> tensor::bubbleUpPadSlice(OpBuilder &b,
void mlir::tensor::registerTilingInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, TensorDialect *dialect) {
- tensor::PadOp::attachInterface<PadOpTiling>(*ctx);
- tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
- tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
- });
+ registry.addExtension(
+ "TENSOR_TILING", +[](MLIRContext *ctx, TensorDialect *dialect) {
+ tensor::PadOp::attachInterface<PadOpTiling>(*ctx);
+ tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
+ tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
+ });
}
void mlir::tensor::registerTilingInterfaceExternalModelsForPackUnPackOps(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, TensorDialect *dialect) {
- tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
- tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
- });
+ registry.addExtension(
+ "TENOR_TILING_PACK_UNPACK",
+ +[](MLIRContext *ctx, TensorDialect *dialect) {
+ tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
+ tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp
index 06f2c16406d3c..3b68822a4813d 100644
--- a/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -114,15 +114,18 @@ struct RankOpInterface
void mlir::tensor::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- tensor::CastOp::attachInterface<tensor::CastOpInterface>(*ctx);
- tensor::DimOp::attachInterface<tensor::DimOpInterface>(*ctx);
- tensor::EmptyOp::attachInterface<tensor::EmptyOpInterface>(*ctx);
- tensor::ExtractSliceOp::attachInterface<tensor::ExtractSliceOpInterface>(
- *ctx);
- tensor::PadOp::attachInterface<tensor::PadOpInterface>(*ctx);
- tensor::RankOp::attachInterface<tensor::RankOpInterface>(*ctx);
- // Note: ValueBoundsOpInterface implementation is not required for ops that
- // implement `DestinationStyleOpInterface` (for querying shaped OpResults).
- });
+ registry.addExtension(
+ "TENSOR_VALUE_BOUNDS",
+ +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ tensor::CastOp::attachInterface<tensor::CastOpInterface>(*ctx);
+ tensor::DimOp::attachInterface<tensor::DimOpInterface>(*ctx);
+ tensor::EmptyOp::attachInterface<tensor::EmptyOpInterface>(*ctx);
+ tensor::ExtractSliceOp::attachInterface<
+ tensor::ExtractSliceOpInterface>(*ctx);
+ tensor::PadOp::attachInterface<tensor::PadOpInterface>(*ctx);
+ tensor::RankOp::attachInterface<tensor::RankOpInterface>(*ctx);
+ // Note: ValueBoundsOpInterface implementation is not required for ops
+ // that implement `DestinationStyleOpInterface` (for querying shaped
+ // OpResults).
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp b/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
index 33016f84056e9..74f795bf02ba4 100644
--- a/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
+++ b/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
@@ -70,15 +70,18 @@ struct ReassociativeReshapeOpReplacementInterface
void tensor::registerFindPayloadReplacementOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- CollapseShapeOp::attachInterface<
- ReassociativeReshapeOpReplacementInterface<CollapseShapeOp>>(*ctx);
- ExpandShapeOp::attachInterface<
- ReassociativeReshapeOpReplacementInterface<ExpandShapeOp>>(*ctx);
- ExtractSliceOp::attachInterface<ExtractSliceOpReplacementInterface>(*ctx);
- InsertSliceOp::attachInterface<InsertSliceOpReplacementInterface>(*ctx);
- ReshapeOp::attachInterface<ReshapeOpReplacementInterface>(*ctx);
- });
+ registry.addExtension(
+ "TENSOR_PAYLOAD_REPLACEMENT",
+ +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ CollapseShapeOp::attachInterface<
+ ReassociativeReshapeOpReplacementInterface<CollapseShapeOp>>(*ctx);
+ ExpandShapeOp::attachInterface<
+ ReassociativeReshapeOpReplacementInterface<ExpandShapeOp>>(*ctx);
+ ExtractSliceOp::attachInterface<ExtractSliceOpReplacementInterface>(
+ *ctx);
+ InsertSliceOp::attachInterface<InsertSliceOpReplacementInterface>(*ctx);
+ ReshapeOp::attachInterface<ReshapeOpReplacementInterface>(*ctx);
+ });
}
//===----------------------------------------------------------------------===//
@@ -247,6 +250,8 @@ class TensorTransformDialectExtension
#include "mlir/Dialect/Tensor/TransformOps/TensorTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "TENSOR_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp
index 87464ccb71720..b7f5fe676a7d2 100644
--- a/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -1055,28 +1055,30 @@ struct SplatOpInterface
void mlir::tensor::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- CastOp::attachInterface<CastOpInterface>(*ctx);
- CollapseShapeOp::attachInterface<CollapseShapeOpInterface>(*ctx);
- DimOp::attachInterface<DimOpInterface>(*ctx);
- EmptyOp::attachInterface<EmptyOpInterface>(*ctx);
- ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
- ExtractSliceOp::attachInterface<ExtractSliceOpInterface>(*ctx);
- ExtractOp::attachInterface<ExtractOpInterface>(*ctx);
- FromElementsOp::attachInterface<FromElementsOpInterface>(*ctx);
- GenerateOp::attachInterface<GenerateOpInterface>(*ctx);
- InsertOp::attachInterface<InsertOpInterface>(*ctx);
- InsertSliceOp::attachInterface<InsertSliceOpInterface>(*ctx);
- PadOp::attachInterface<PadOpInterface>(*ctx);
- ParallelInsertSliceOp::attachInterface<ParallelInsertSliceOpInterface>(
- *ctx);
- RankOp::attachInterface<RankOpInterface>(*ctx);
- ReshapeOp::attachInterface<ReshapeOpInterface>(*ctx);
- SplatOp::attachInterface<SplatOpInterface>(*ctx);
-
- // Load additional dialects of which ops may get created.
- ctx->loadDialect<arith::ArithDialect, linalg::LinalgDialect>();
- });
+ registry.addExtension(
+ "TENSOR_BUFFERIZATION",
+ +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ CastOp::attachInterface<CastOpInterface>(*ctx);
+ CollapseShapeOp::attachInterface<CollapseShapeOpInterface>(*ctx);
+ DimOp::attachInterface<DimOpInterface>(*ctx);
+ EmptyOp::attachInterface<EmptyOpInterface>(*ctx);
+ ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
+ ExtractSliceOp::attachInterface<ExtractSliceOpInterface>(*ctx);
+ ExtractOp::attachInterface<ExtractOpInterface>(*ctx);
+ FromElementsOp::attachInterface<FromElementsOpInterface>(*ctx);
+ GenerateOp::attachInterface<GenerateOpInterface>(*ctx);
+ InsertOp::attachInterface<InsertOpInterface>(*ctx);
+ InsertSliceOp::attachInterface<InsertSliceOpInterface>(*ctx);
+ PadOp::attachInterface<PadOpInterface>(*ctx);
+ ParallelInsertSliceOp::attachInterface<ParallelInsertSliceOpInterface>(
+ *ctx);
+ RankOp::attachInterface<RankOpInterface>(*ctx);
+ ReshapeOp::attachInterface<ReshapeOpInterface>(*ctx);
+ SplatOp::attachInterface<SplatOpInterface>(*ctx);
+
+ // Load additional dialects of which ops may get created.
+ ctx->loadDialect<arith::ArithDialect, linalg::LinalgDialect>();
+ });
// Bufferization requires SubsetInsertionOpInterface models. Make sure that
// they are registered.
diff --git a/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp
index d50d7c62b789c..2b052cbb762e1 100644
--- a/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp
@@ -86,20 +86,21 @@ struct InsertSliceLikeOpSubsetInsertionOpInterface
void mlir::tensor::registerSubsetOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- // Note: `SubsetExtractionOpInterface` and `SubsetInsertionOpInterface`
- // require `SubsetOpInterface`.
- ExtractSliceOp::attachInterface<ExtractSliceOpSubsetOpInterface>(*ctx);
- ExtractSliceOp::attachInterface<ExtractSliceOpSubsetExtractionOpInterface>(
- *ctx);
- InsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetOpInterface<InsertSliceOp>>(*ctx);
- InsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetInsertionOpInterface<InsertSliceOp>>(*ctx);
- ParallelInsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetOpInterface<ParallelInsertSliceOp>>(*ctx);
- ParallelInsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetInsertionOpInterface<ParallelInsertSliceOp>>(
- *ctx);
- });
+ registry.addExtension(
+ "TENSOR_SUBSET", +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ // Note: `SubsetExtractionOpInterface` and `SubsetInsertionOpInterface`
+ // require `SubsetOpInterface`.
+ ExtractSliceOp::attachInterface<ExtractSliceOpSubsetOpInterface>(*ctx);
+ ExtractSliceOp::attachInterface<
+ ExtractSliceOpSubsetExtractionOpInterface>(*ctx);
+ InsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetOpInterface<InsertSliceOp>>(*ctx);
+ InsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetInsertionOpInterface<InsertSliceOp>>(*ctx);
+ ParallelInsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetOpInterface<ParallelInsertSliceOp>>(*ctx);
+ ParallelInsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetInsertionOpInterface<ParallelInsertSliceOp>>(
+ *ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp b/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp
index ffbb707344b8c..6161d5ae7e3ee 100644
--- a/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp
@@ -76,15 +76,17 @@ static void registerElemwiseAll(MLIRContext *ctx) {
void mlir::tosa::registerShardingInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, TosaDialect *dialect) {
- registerElemwiseAll<
- ClampOp, SigmoidOp, TanhOp, AddOp, ArithmeticRightShiftOp, BitwiseAndOp,
- BitwiseOrOp, BitwiseXorOp, IntDivOp, LogicalAndOp, LogicalLeftShiftOp,
- LogicalRightShiftOp, LogicalOrOp, LogicalXorOp, MaximumOp, MinimumOp,
- MulOp, PowOp, SubOp, AbsOp, BitwiseNotOp, CeilOp, ClzOp, ExpOp, FloorOp,
- LogOp, LogicalNotOp, NegateOp, ReciprocalOp, RsqrtOp, SelectOp, EqualOp,
- GreaterOp, GreaterEqualOp>(ctx);
+ registry.addExtension(
+ "TOSA_SHARDING", +[](MLIRContext *ctx, TosaDialect *dialect) {
+ registerElemwiseAll<
+ ClampOp, SigmoidOp, TanhOp, AddOp, ArithmeticRightShiftOp,
+ BitwiseAndOp, BitwiseOrOp, BitwiseXorOp, IntDivOp, LogicalAndOp,
+ LogicalLeftShiftOp, LogicalRightShiftOp, LogicalOrOp, LogicalXorOp,
+ MaximumOp, MinimumOp, MulOp, PowOp, SubOp, AbsOp, BitwiseNotOp,
+ CeilOp, ClzOp, ExpOp, FloorOp, LogOp, LogicalNotOp, NegateOp,
+ ReciprocalOp, RsqrtOp, SelectOp, EqualOp, GreaterOp,
+ GreaterEqualOp>(ctx);
- MatMulOp::attachInterface<MatMulOpSharding>(*ctx);
- });
+ MatMulOp::attachInterface<MatMulOpSharding>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp b/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
index e369daddb00cb..1e3c426162462 100644
--- a/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
+++ b/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
@@ -26,6 +26,8 @@ class DebugExtension
#include "mlir/Dialect/Transform/DebugExtension/DebugExtensionOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "TRANSFORM_DEBUG";
};
} // namespace
diff --git a/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp b/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
index 94004365b8a1a..8cae64ff8de0a 100644
--- a/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
+++ b/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
@@ -26,6 +26,8 @@ class IRDLExtension
declareDependentDialect<irdl::IRDLDialect>();
}
+
+ static constexpr llvm::StringRef extensionID = "IRDL_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp b/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
index b33288fd7b991..032cd65e1065f 100644
--- a/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
+++ b/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
@@ -26,6 +26,8 @@ class LoopExtension
#include "mlir/Dialect/Transform/LoopExtension/LoopExtensionOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "TRANSFORM_LOOP";
};
} // namespace
diff --git a/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp b/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
index 2c770abd56d52..c7f1e63e3d921 100644
--- a/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
+++ b/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
@@ -61,6 +61,8 @@ class PDLExtension : public transform::TransformDialectExtension<PDLExtension> {
PDLOperationTypeTransformHandleTypeInterfaceImpl>(*context);
});
}
+
+ static constexpr llvm::StringRef extensionID = "TRANSFORM_PDL";
};
} // namespace
diff --git a/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp
index ca95072d9bb0f..91bf4ef633a6d 100644
--- a/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -44,8 +44,10 @@ struct VectorScaleOpInterface
void mlir::vector::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, vector::VectorDialect *dialect) {
- vector::VectorScaleOp::attachInterface<vector::VectorScaleOpInterface>(
- *ctx);
- });
+ registry.addExtension(
+ "VECTOR_VALUE_BOUNDS",
+ +[](MLIRContext *ctx, vector::VectorDialect *dialect) {
+ vector::VectorScaleOp::attachInterface<vector::VectorScaleOpInterface>(
+ *ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp b/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
index 2e9aa88011825..0b42040a4d63b 100644
--- a/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
+++ b/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
@@ -220,6 +220,8 @@ class VectorTransformDialectExtension
#include "mlir/Dialect/Vector/TransformOps/VectorTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "VECTOR_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp
index 1caec5bb8644f..d3e01e33b3f1d 100644
--- a/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -317,11 +317,13 @@ struct YieldOpInterface
void mlir::vector::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, vector::VectorDialect *dialect) {
- TransferReadOp::attachInterface<TransferReadOpInterface>(*ctx);
- TransferWriteOp::attachInterface<TransferWriteOpInterface>(*ctx);
- GatherOp::attachInterface<GatherOpInterface>(*ctx);
- MaskOp::attachInterface<MaskOpInterface>(*ctx);
- YieldOp::attachInterface<YieldOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "VECTOR_BUFFERIZATION",
+ +[](MLIRContext *ctx, vector::VectorDialect *dialect) {
+ TransferReadOp::attachInterface<TransferReadOpInterface>(*ctx);
+ TransferWriteOp::attachInterface<TransferWriteOpInterface>(*ctx);
+ GatherOp::attachInterface<GatherOpInterface>(*ctx);
+ MaskOp::attachInterface<MaskOpInterface>(*ctx);
+ YieldOp::attachInterface<YieldOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp b/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp
index b450d5b78a466..91a014a7ab24b 100644
--- a/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp
@@ -69,14 +69,15 @@ struct TransferWriteOpSubsetInsertionOpInterface
void mlir::vector::registerSubsetOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, vector::VectorDialect *dialect) {
- TransferReadOp::attachInterface<XferOpSubsetOpInterface<TransferReadOp>>(
- *ctx);
- TransferReadOp::attachInterface<TransferReadOpSubsetExtractionOpInterface>(
- *ctx);
- TransferWriteOp::attachInterface<XferOpSubsetOpInterface<TransferWriteOp>>(
- *ctx);
- TransferWriteOp::attachInterface<TransferWriteOpSubsetInsertionOpInterface>(
- *ctx);
- });
+ registry.addExtension(
+ "VECTOR_SUBSET", +[](MLIRContext *ctx, vector::VectorDialect *dialect) {
+ TransferReadOp::attachInterface<
+ XferOpSubsetOpInterface<TransferReadOp>>(*ctx);
+ TransferReadOp::attachInterface<
+ TransferReadOpSubsetExtractionOpInterface>(*ctx);
+ TransferWriteOp::attachInterface<
+ XferOpSubsetOpInterface<TransferWriteOp>>(*ctx);
+ TransferWriteOp::attachInterface<
+ TransferWriteOpSubsetInsertionOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Interfaces/CastInterfaces.cpp b/mlir/lib/Interfaces/CastInterfaces.cpp
index 05c872daf5dab..11e581a5fefa8 100644
--- a/mlir/lib/Interfaces/CastInterfaces.cpp
+++ b/mlir/lib/Interfaces/CastInterfaces.cpp
@@ -79,10 +79,11 @@ struct UnrealizedConversionCastOpInterface
void mlir::builtin::registerCastOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- UnrealizedConversionCastOp::attachInterface<
- UnrealizedConversionCastOpInterface>(*ctx);
- });
+ registry.addExtension(
+ "BUILTIN_CAST", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ UnrealizedConversionCastOp::attachInterface<
+ UnrealizedConversionCastOpInterface>(*ctx);
+ });
}
//===----------------------------------------------------------------------===//
diff --git a/mlir/lib/Target/LLVM/NVVM/Target.cpp b/mlir/lib/Target/LLVM/NVVM/Target.cpp
index e608d26e8d2ec..f5d1babd01180 100644
--- a/mlir/lib/Target/LLVM/NVVM/Target.cpp
+++ b/mlir/lib/Target/LLVM/NVVM/Target.cpp
@@ -58,9 +58,10 @@ class NVVMTargetAttrImpl
// Register the NVVM dialect, the NVVM translation & the target interface.
void mlir::NVVM::registerNVVMTargetInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
- NVVMTargetAttr::attachInterface<NVVMTargetAttrImpl>(*ctx);
- });
+ registry.addExtension(
+ "NVVM_TARGET", +[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
+ NVVMTargetAttr::attachInterface<NVVMTargetAttrImpl>(*ctx);
+ });
}
void mlir::NVVM::registerNVVMTargetInterfaceExternalModels(
diff --git a/mlir/lib/Target/LLVM/ROCDL/Target.cpp b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
index 4d23f987eb05e..3a56688117fda 100644
--- a/mlir/lib/Target/LLVM/ROCDL/Target.cpp
+++ b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
@@ -68,9 +68,10 @@ class ROCDLTargetAttrImpl
// Register the ROCDL dialect, the ROCDL translation and the target interface.
void mlir::ROCDL::registerROCDLTargetInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
- ROCDLTargetAttr::attachInterface<ROCDLTargetAttrImpl>(*ctx);
- });
+ registry.addExtension(
+ "ROCDL_TARGET", +[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
+ ROCDLTargetAttr::attachInterface<ROCDLTargetAttrImpl>(*ctx);
+ });
}
void mlir::ROCDL::registerROCDLTargetInterfaceExternalModels(
diff --git a/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp
index 044462d33cfd1..4ebcfa5941d0e 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp
@@ -44,9 +44,10 @@ class AMXDialectLLVMIRTranslationInterface
void mlir::registerAMXDialectTranslation(DialectRegistry ®istry) {
registry.insert<amx::AMXDialect>();
- registry.addExtension(+[](MLIRContext *ctx, amx::AMXDialect *dialect) {
- dialect->addInterfaces<AMXDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "AMX_TO_LLVMIR", +[](MLIRContext *ctx, amx::AMXDialect *dialect) {
+ dialect->addInterfaces<AMXDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerAMXDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp
index 7098592d506e0..d317f68731eb8 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp
@@ -46,6 +46,7 @@ class ArmNeonDialectLLVMIRTranslationInterface
void mlir::registerArmNeonDialectTranslation(DialectRegistry ®istry) {
registry.insert<arm_neon::ArmNeonDialect>();
registry.addExtension(
+ "ARM_NEON_TO_LLVMIR",
+[](MLIRContext *ctx, arm_neon::ArmNeonDialect *dialect) {
dialect->addInterfaces<ArmNeonDialectLLVMIRTranslationInterface>();
});
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp
index e6ee41188d594..a1f63b9f07bdd 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp
@@ -45,9 +45,11 @@ class ArmSMEDialectLLVMIRTranslationInterface
void mlir::registerArmSMEDialectTranslation(DialectRegistry ®istry) {
registry.insert<arm_sme::ArmSMEDialect>();
- registry.addExtension(+[](MLIRContext *ctx, arm_sme::ArmSMEDialect *dialect) {
- dialect->addInterfaces<ArmSMEDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "ARM_SME_TO_LLVMIR",
+ +[](MLIRContext *ctx, arm_sme::ArmSMEDialect *dialect) {
+ dialect->addInterfaces<ArmSMEDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerArmSMEDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp
index cd10811b68f02..3411f5d147a34 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp
@@ -44,9 +44,11 @@ class ArmSVEDialectLLVMIRTranslationInterface
void mlir::registerArmSVEDialectTranslation(DialectRegistry ®istry) {
registry.insert<arm_sve::ArmSVEDialect>();
- registry.addExtension(+[](MLIRContext *ctx, arm_sve::ArmSVEDialect *dialect) {
- dialect->addInterfaces<ArmSVEDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "ARM_SVE_TO_LLVMIR",
+ +[](MLIRContext *ctx, arm_sve::ArmSVEDialect *dialect) {
+ dialect->addInterfaces<ArmSVEDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerArmSVEDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp
index 51c304cfbb8e5..c120eb243b783 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp
@@ -34,9 +34,10 @@ class BuiltinDialectLLVMIRTranslationInterface
} // namespace
void mlir::registerBuiltinDialectTranslation(DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- dialect->addInterfaces<BuiltinDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "BUILTIN_TO_LLVMIR", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ dialect->addInterfaces<BuiltinDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerBuiltinDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp
index eecc8f1001ca4..c4fff8cfee7d1 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp
@@ -65,9 +65,10 @@ class GPUDialectLLVMIRTranslationInterface
void mlir::registerGPUDialectTranslation(DialectRegistry ®istry) {
registry.insert<gpu::GPUDialect>();
- registry.addExtension(+[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
- dialect->addInterfaces<GPUDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "GPU_TO_LLVMIR", +[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
+ dialect->addInterfaces<GPUDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerGPUDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp b/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp
index b023c4c126da3..c982e5f67900f 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp
@@ -56,9 +56,10 @@ std::string getBinaryIdentifier(StringRef binaryName) {
void mlir::gpu::registerOffloadingLLVMTranslationInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
- SelectObjectAttr::attachInterface<SelectObjectAttrImpl>(*ctx);
- });
+ registry.addExtension(
+ "GPU_OFFLOADING_LLVM", +[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
+ SelectObjectAttr::attachInterface<SelectObjectAttrImpl>(*ctx);
+ });
}
gpu::ObjectAttr
diff --git a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp
index 06673965245c0..14485c54ffdde 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp
@@ -281,9 +281,10 @@ class LLVMDialectLLVMIRImportInterface : public LLVMImportDialectInterface {
void mlir::registerLLVMDialectImport(DialectRegistry ®istry) {
registry.insert<LLVM::LLVMDialect>();
- registry.addExtension(+[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
- dialect->addInterfaces<LLVMDialectLLVMIRImportInterface>();
- });
+ registry.addExtension(
+ "LLVMIR_TO_LLVM", +[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
+ dialect->addInterfaces<LLVMDialectLLVMIRImportInterface>();
+ });
}
void mlir::registerLLVMDialectImport(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp
index bdb15a290209b..d5bc600fd7e13 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp
@@ -439,9 +439,10 @@ class LLVMDialectLLVMIRTranslationInterface
void mlir::registerLLVMDialectTranslation(DialectRegistry ®istry) {
registry.insert<LLVM::LLVMDialect>();
- registry.addExtension(+[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
- dialect->addInterfaces<LLVMDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "LLVM_TO_LLVMIR", +[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
+ dialect->addInterfaces<LLVMDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerLLVMDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp
index 855abc12a909e..9978513e59ec0 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp
@@ -81,9 +81,10 @@ class NVVMDialectLLVMIRImportInterface : public LLVMImportDialectInterface {
void mlir::registerNVVMDialectImport(DialectRegistry ®istry) {
registry.insert<NVVM::NVVMDialect>();
- registry.addExtension(+[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
- dialect->addInterfaces<NVVMDialectLLVMIRImportInterface>();
- });
+ registry.addExtension(
+ "LLVMIR_TO_NVVM", +[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
+ dialect->addInterfaces<NVVMDialectLLVMIRImportInterface>();
+ });
}
void mlir::registerNVVMDialectImport(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp
index a09c24dda82af..e12364b1122f0 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp
@@ -261,9 +261,10 @@ class NVVMDialectLLVMIRTranslationInterface
void mlir::registerNVVMDialectTranslation(DialectRegistry ®istry) {
registry.insert<NVVM::NVVMDialect>();
- registry.addExtension(+[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
- dialect->addInterfaces<NVVMDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "NVVM_TO_LLVMIR", +[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
+ dialect->addInterfaces<NVVMDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerNVVMDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp
index d9cf85e4aecab..67f0f8fc4f374 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp
@@ -532,9 +532,10 @@ LogicalResult OpenACCDialectLLVMIRTranslationInterface::convertOperation(
void mlir::registerOpenACCDialectTranslation(DialectRegistry ®istry) {
registry.insert<acc::OpenACCDialect>();
- registry.addExtension(+[](MLIRContext *ctx, acc::OpenACCDialect *dialect) {
- dialect->addInterfaces<OpenACCDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "OPENACC_TO_LLVMIR", +[](MLIRContext *ctx, acc::OpenACCDialect *dialect) {
+ dialect->addInterfaces<OpenACCDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerOpenACCDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp
index ddee117838697..a43761460fa31 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp
@@ -3741,9 +3741,10 @@ LogicalResult OpenMPDialectLLVMIRTranslationInterface::convertOperation(
void mlir::registerOpenMPDialectTranslation(DialectRegistry ®istry) {
registry.insert<omp::OpenMPDialect>();
- registry.addExtension(+[](MLIRContext *ctx, omp::OpenMPDialect *dialect) {
- dialect->addInterfaces<OpenMPDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "OPENMV_TO_LLVMIR", +[](MLIRContext *ctx, omp::OpenMPDialect *dialect) {
+ dialect->addInterfaces<OpenMPDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerOpenMPDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp
index 2a146f5efed30..7fe58e5d161a8 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp
@@ -190,9 +190,10 @@ class ROCDLDialectLLVMIRTranslationInterface
void mlir::registerROCDLDialectTranslation(DialectRegistry ®istry) {
registry.insert<ROCDL::ROCDLDialect>();
- registry.addExtension(+[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
- dialect->addInterfaces<ROCDLDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "ROCDL_TO_LLVMIR", +[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
+ dialect->addInterfaces<ROCDLDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerROCDLDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp
index b78b002d32292..17d63863d9ddc 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp
@@ -76,9 +76,10 @@ class VCIXDialectLLVMIRTranslationInterface
void mlir::registerVCIXDialectTranslation(DialectRegistry ®istry) {
registry.insert<vcix::VCIXDialect>();
- registry.addExtension(+[](MLIRContext *ctx, vcix::VCIXDialect *dialect) {
- dialect->addInterfaces<VCIXDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(
+ "VCIX_TO_LLVMIR", +[](MLIRContext *ctx, vcix::VCIXDialect *dialect) {
+ dialect->addInterfaces<VCIXDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerVCIXDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp
index fa5f61420ee8a..5a6104162123b 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp
@@ -46,6 +46,7 @@ class X86VectorDialectLLVMIRTranslationInterface
void mlir::registerX86VectorDialectTranslation(DialectRegistry ®istry) {
registry.insert<x86vector::X86VectorDialect>();
registry.addExtension(
+ "X86_TO_LLVMIR",
+[](MLIRContext *ctx, x86vector::X86VectorDialect *dialect) {
dialect->addInterfaces<X86VectorDialectLLVMIRTranslationInterface>();
});
diff --git a/mlir/lib/Target/SPIRV/Target.cpp b/mlir/lib/Target/SPIRV/Target.cpp
index 4c416abe71cac..15832a43b6620 100644
--- a/mlir/lib/Target/SPIRV/Target.cpp
+++ b/mlir/lib/Target/SPIRV/Target.cpp
@@ -43,9 +43,10 @@ class SPIRVTargetAttrImpl
// Register the SPIR-V dialect, the SPIR-V translation & the target interface.
void mlir::spirv::registerSPIRVTargetInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(+[](MLIRContext *ctx, spirv::SPIRVDialect *dialect) {
- spirv::TargetEnvAttr::attachInterface<SPIRVTargetAttrImpl>(*ctx);
- });
+ registry.addExtension(
+ "SPIRV_TARGET", +[](MLIRContext *ctx, spirv::SPIRVDialect *dialect) {
+ spirv::TargetEnvAttr::attachInterface<SPIRVTargetAttrImpl>(*ctx);
+ });
}
void mlir::spirv::registerSPIRVTargetInterfaceExternalModels(
diff --git a/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp b/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp
index dc6413b25707e..0422705663b81 100644
--- a/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp
+++ b/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp
@@ -104,6 +104,7 @@ void registerTestFromLLVMIR() {
registry.insert<test::TestDialect>();
registerLLVMDialectImport(registry);
registry.addExtension(
+ "TEST_FROM_LLVMIR",
+[](MLIRContext *ctx, test::TestDialect *dialect) {
dialect->addInterfaces<TestDialectLLVMImportDialectInterface>();
});
diff --git a/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp b/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp
index 157c6265be834..3d4b8dc6b1495 100644
--- a/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp
+++ b/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp
@@ -140,6 +140,7 @@ void registerTestToLLVMIR() {
registerBuiltinDialectTranslation(registry);
registerLLVMDialectTranslation(registry);
registry.addExtension(
+ "TEST_TO_LLVMIR",
+[](MLIRContext *ctx, test::TestDialect *dialect) {
dialect->addInterfaces<TestDialectLLVMIRTranslationInterface>();
});
diff --git a/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp b/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
index b8a4b9470d736..cc8f0e8988ea6 100644
--- a/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
+++ b/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
@@ -905,6 +905,8 @@ class TestTransformDialectExtension
hooks.mergeInPDLMatchHooks(std::move(constraints));
});
}
+
+ static constexpr llvm::StringRef extensionID = "TEST_TRANSFORM";
};
} // namespace
diff --git a/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp b/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
index 8f206d9077272..172001123d4da 100644
--- a/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
+++ b/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
@@ -399,6 +399,8 @@ class TestTilingInterfaceDialectExtension
#include "TestTilingInterfaceTransformOps.cpp.inc"
>();
}
+
+ static constexpr llvm::StringRef extensionID = "TEST_TILING_TRANSFORM";
};
} // namespace
diff --git a/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp b/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
index 40fb752ffd6eb..2d8fbf078618f 100644
--- a/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
+++ b/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
@@ -20,6 +20,8 @@ class Extension : public TransformDialectExtension<Extension> {
public:
using Base::Base;
void init() { declareGeneratedDialect<func::FuncDialect>(); }
+
+ static constexpr llvm::StringRef extensionID = "TRANSFORM_BUILD_ONLY";
};
} // end namespace
diff --git a/mlir/unittests/IR/DialectTest.cpp b/mlir/unittests/IR/DialectTest.cpp
index e99d46e6d2643..b79d7941a12a0 100644
--- a/mlir/unittests/IR/DialectTest.cpp
+++ b/mlir/unittests/IR/DialectTest.cpp
@@ -75,7 +75,8 @@ TEST(Dialect, DelayedInterfaceRegistration) {
registry.insert<TestDialect, SecondTestDialect>();
// Delayed registration of an interface for TestDialect.
- registry.addExtension(+[](MLIRContext *ctx, TestDialect *dialect) {
+ registry.addExtension(
+ "TEST_DIALECT_DELAYED", +[](MLIRContext *ctx, TestDialect *dialect) {
dialect->addInterfaces<TestDialectInterface>();
});
@@ -100,7 +101,7 @@ TEST(Dialect, DelayedInterfaceRegistration) {
DialectRegistry secondRegistry;
secondRegistry.insert<SecondTestDialect>();
secondRegistry.addExtension(
- +[](MLIRContext *ctx, SecondTestDialect *dialect) {
+ "SECOND_TEST", +[](MLIRContext *ctx, SecondTestDialect *dialect) {
dialect->addInterfaces<SecondTestDialectInterface>();
});
context.appendDialectRegistry(secondRegistry);
@@ -113,7 +114,8 @@ TEST(Dialect, RepeatedDelayedRegistration) {
// Set up the delayed registration.
DialectRegistry registry;
registry.insert<TestDialect>();
- registry.addExtension(+[](MLIRContext *ctx, TestDialect *dialect) {
+ registry.addExtension(
+ "TEST_DIALECT", +[](MLIRContext *ctx, TestDialect *dialect) {
dialect->addInterfaces<TestDialectInterface>();
});
MLIRContext context(registry);
@@ -128,7 +130,8 @@ TEST(Dialect, RepeatedDelayedRegistration) {
// on repeated interface registration.
DialectRegistry secondRegistry;
secondRegistry.insert<TestDialect>();
- secondRegistry.addExtension(+[](MLIRContext *ctx, TestDialect *dialect) {
+ secondRegistry.addExtension(
+ "TEST_DIALECT", +[](MLIRContext *ctx, TestDialect *dialect) {
dialect->addInterfaces<TestDialectInterface>();
});
context.appendDialectRegistry(secondRegistry);
@@ -143,12 +146,18 @@ struct DummyExtension : DialectExtension<DummyExtension, TestDialect> {
DummyExtension(int *counter, int numRecursive)
: DialectExtension(), counter(counter), numRecursive(numRecursive) {}
+ inline static std::vector<std::string> extensionIDs;
+
void apply(MLIRContext *ctx, TestDialect *dialect) const final {
++(*counter);
DialectRegistry nestedRegistry;
- for (int i = 0; i < numRecursive; ++i)
+ extensionIDs.reserve(extensionIDs.size() + numRecursive);
+ for (int i = 0; i < numRecursive; ++i) {
+ extensionIDs.push_back("DUMMY_" + std::to_string(i));
nestedRegistry.addExtension(
+ extensionIDs.back(),
std::make_unique<DummyExtension>(counter, /*numRecursive=*/0));
+ }
// Adding additional extensions may trigger a reallocation of the
// `extensions` vector in the dialect registry.
ctx->appendDialectRegistry(nestedRegistry);
@@ -166,10 +175,12 @@ TEST(Dialect, NestedDialectExtension) {
// Add an extension that adds 100 more extensions.
int counter1 = 0;
- registry.addExtension(std::make_unique<DummyExtension>(&counter1, 100));
+ registry.addExtension("DUMMY",
+ std::make_unique<DummyExtension>(&counter1, 100));
// Add one more extension. This should not crash.
int counter2 = 0;
- registry.addExtension(std::make_unique<DummyExtension>(&counter2, 0));
+ registry.addExtension("DUMMY2",
+ std::make_unique<DummyExtension>(&counter2, 0));
// Load dialect and apply extensions.
MLIRContext context(registry);
diff --git a/mlir/unittests/IR/InterfaceAttachmentTest.cpp b/mlir/unittests/IR/InterfaceAttachmentTest.cpp
index b6066dd5685dc..a5b835153b636 100644
--- a/mlir/unittests/IR/InterfaceAttachmentTest.cpp
+++ b/mlir/unittests/IR/InterfaceAttachmentTest.cpp
@@ -103,9 +103,11 @@ TEST(InterfaceAttachment, TypeDelayedContextConstruct) {
// Put the interface in the registry.
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::TestType::attachInterface<TestTypeModel>(*ctx);
- });
+ registry.addExtension(
+ "TYPE_DELAYED_CONSTRUCT",
+ +[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::TestType::attachInterface<TestTypeModel>(*ctx);
+ });
// Check that when a context is constructed with the given registry, the type
// interface gets registered.
@@ -122,9 +124,10 @@ TEST(InterfaceAttachment, TypeDelayedContextAppend) {
// Put the interface in the registry.
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::TestType::attachInterface<TestTypeModel>(*ctx);
- });
+ registry.addExtension(
+ "TYPE_DELAYED_APPEND", +[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::TestType::attachInterface<TestTypeModel>(*ctx);
+ });
// Check that when the registry gets appended to the context, the interface
// becomes available for objects in loaded dialects.
@@ -138,9 +141,10 @@ TEST(InterfaceAttachment, TypeDelayedContextAppend) {
TEST(InterfaceAttachment, RepeatedRegistration) {
DialectRegistry registry;
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- IntegerType::attachInterface<Model>(*ctx);
- });
+ registry.addExtension(
+ "REPEATED", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ IntegerType::attachInterface<Model>(*ctx);
+ });
MLIRContext context(registry);
// Should't fail on repeated registration through the dialect registry.
@@ -151,9 +155,10 @@ TEST(InterfaceAttachment, TypeBuiltinDelayed) {
// Builtin dialect needs to registration or loading, but delayed interface
// registration must still work.
DialectRegistry registry;
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- IntegerType::attachInterface<Model>(*ctx);
- });
+ registry.addExtension(
+ "BUILTIN_DELAYED", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ IntegerType::attachInterface<Model>(*ctx);
+ });
MLIRContext context(registry);
IntegerType i16 = IntegerType::get(&context, 16);
@@ -246,9 +251,10 @@ TEST(InterfaceAttachmentTest, AttributeDelayed) {
// that the delayed registration work for attributes.
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::SimpleAAttr::attachInterface<TestExternalSimpleAAttrModel>(*ctx);
- });
+ registry.addExtension(
+ "ATTRIBUTE_DELAYED", +[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::SimpleAAttr::attachInterface<TestExternalSimpleAAttrModel>(*ctx);
+ });
MLIRContext context(registry);
context.loadDialect<test::TestDialect>();
@@ -352,13 +358,17 @@ struct TestExternalTestOpModel
TEST(InterfaceAttachment, OperationDelayedContextConstruct) {
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
- });
- registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
- test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
- });
+ registry.addExtension(
+ "OPERATION_DELAYED_BUILTIN",
+ +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
+ });
+ registry.addExtension(
+ "OPERATION_DELAYED_TEST",
+ +[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
+ test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
+ });
// Construct the context directly from a registry. The interfaces are
// expected to be readily available on operations.
@@ -383,13 +393,17 @@ TEST(InterfaceAttachment, OperationDelayedContextConstruct) {
TEST(InterfaceAttachment, OperationDelayedContextAppend) {
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
- ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
- });
- registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
- test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
- });
+ registry.addExtension(
+ "OPERATION_DELAYED_BUILTIN",
+ +[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
+ });
+ registry.addExtension(
+ "OPERATION_DELAYED_TEST",
+ +[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
+ test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
+ });
// Construct the context, create ops, and only then append the registry. The
// interfaces are expected to be available after appending the registry.
>From 0cf68e846b20027c7903551c6a0d2321f4c06f88 Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Wed, 31 Jul 2024 19:36:45 -0700
Subject: [PATCH 3/8] add dialect registry extension comparison by key
---
mlir/lib/IR/Dialect.cpp | 14 ++++++++++--
mlir/unittests/IR/DialectTest.cpp | 37 ++++++++++++++++++++++++++-----
2 files changed, 43 insertions(+), 8 deletions(-)
diff --git a/mlir/lib/IR/Dialect.cpp b/mlir/lib/IR/Dialect.cpp
index 0bdc34bbc5ae8..c8b55bf751091 100644
--- a/mlir/lib/IR/Dialect.cpp
+++ b/mlir/lib/IR/Dialect.cpp
@@ -291,9 +291,19 @@ void DialectRegistry::applyExtensions(MLIRContext *ctx) const {
}
bool DialectRegistry::isSubsetOf(const DialectRegistry &rhs) const {
- // Treat any extensions conservatively.
- if (!extensions.empty())
+ // Check that all extension keys are present in 'rhs'.
+ llvm::DenseSet<llvm::StringRef> rhsExtensionKeys;
+ {
+ auto rhsKeys = llvm::map_range(rhs.extensions,
+ [](const auto &item) { return item.first; });
+ rhsExtensionKeys.insert(rhsKeys.begin(), rhsKeys.end());
+ }
+
+ if (!llvm::all_of(extensions, [&rhsExtensionKeys](const auto &extension) {
+ return rhsExtensionKeys.contains(extension.first);
+ }))
return false;
+
// Check that the current dialects fully overlap with the dialects in 'rhs'.
return llvm::all_of(
registry, [&](const auto &it) { return rhs.registry.count(it.first); });
diff --git a/mlir/unittests/IR/DialectTest.cpp b/mlir/unittests/IR/DialectTest.cpp
index b79d7941a12a0..2623978639674 100644
--- a/mlir/unittests/IR/DialectTest.cpp
+++ b/mlir/unittests/IR/DialectTest.cpp
@@ -77,8 +77,8 @@ TEST(Dialect, DelayedInterfaceRegistration) {
// Delayed registration of an interface for TestDialect.
registry.addExtension(
"TEST_DIALECT_DELAYED", +[](MLIRContext *ctx, TestDialect *dialect) {
- dialect->addInterfaces<TestDialectInterface>();
- });
+ dialect->addInterfaces<TestDialectInterface>();
+ });
MLIRContext context(registry);
@@ -116,8 +116,8 @@ TEST(Dialect, RepeatedDelayedRegistration) {
registry.insert<TestDialect>();
registry.addExtension(
"TEST_DIALECT", +[](MLIRContext *ctx, TestDialect *dialect) {
- dialect->addInterfaces<TestDialectInterface>();
- });
+ dialect->addInterfaces<TestDialectInterface>();
+ });
MLIRContext context(registry);
// Load the TestDialect and check that the interface got registered for it.
@@ -132,8 +132,8 @@ TEST(Dialect, RepeatedDelayedRegistration) {
secondRegistry.insert<TestDialect>();
secondRegistry.addExtension(
"TEST_DIALECT", +[](MLIRContext *ctx, TestDialect *dialect) {
- dialect->addInterfaces<TestDialectInterface>();
- });
+ dialect->addInterfaces<TestDialectInterface>();
+ });
context.appendDialectRegistry(secondRegistry);
testDialectInterface = dyn_cast<TestDialectInterfaceBase>(testDialect);
EXPECT_TRUE(testDialectInterface != nullptr);
@@ -193,4 +193,29 @@ TEST(Dialect, NestedDialectExtension) {
EXPECT_GE(counter2, 1);
}
+TEST(Dialect, SubsetWithExtensions) {
+ DialectRegistry registry1, registry2;
+ registry1.insert<TestDialect>();
+ registry2.insert<TestDialect>();
+
+ // Validate that the registries are equivalent.
+ ASSERT_TRUE(registry1.isSubsetOf(registry2));
+ ASSERT_TRUE(registry2.isSubsetOf(registry1));
+
+ // Add extensions to registry2.
+ int counter;
+ registry2.addExtension("EXT", std::make_unique<DummyExtension>(&counter, 0));
+
+ // Expect that (1) is a subset of (2) but not the other way around.
+ ASSERT_TRUE(registry1.isSubsetOf(registry2));
+ ASSERT_FALSE(registry2.isSubsetOf(registry1));
+
+ // Add extensions to registry1.
+ registry1.addExtension("EXT", std::make_unique<DummyExtension>(&counter, 0));
+
+ // Expect that (1) and (2) are equivalent.
+ ASSERT_TRUE(registry1.isSubsetOf(registry2));
+ ASSERT_TRUE(registry2.isSubsetOf(registry1));
+}
+
} // namespace
>From 07de1d3fa32f7cfd671f2942badf079b4ffafdc6 Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Thu, 1 Aug 2024 08:39:41 -0700
Subject: [PATCH 4/8] Revert "attach extension IDs"
This reverts commit 26669bd8a53297dfe545b407d438c85d0ce32bfd.
---
flang/lib/Optimizer/Dialect/FIRDialect.cpp | 3 +-
.../transform/Ch2/lib/MyExtension.cpp | 3 -
.../transform/Ch3/lib/MyExtension.cpp | 3 -
.../transform/Ch4/lib/MyExtension.cpp | 3 -
.../Conversion/ArithToLLVM/ArithToLLVM.cpp | 7 +-
.../ComplexToLLVM/ComplexToLLVM.cpp | 1 -
.../ControlFlowToLLVM/ControlFlowToLLVM.cpp | 7 +-
.../ConvertToLLVM/ConvertToLLVMPass.cpp | 5 --
mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp | 7 +-
.../Conversion/IndexToLLVM/IndexToLLVM.cpp | 7 +-
mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp | 7 +-
.../Conversion/MemRefToLLVM/MemRefToLLVM.cpp | 7 +-
mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp | 7 +-
mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp | 7 +-
.../Affine/IR/ValueBoundsOpInterfaceImpl.cpp | 11 ++-
.../TransformOps/AffineTransformOps.cpp | 2 -
.../Arith/IR/ValueBoundsOpInterfaceImpl.cpp | 16 ++---
.../BufferDeallocationOpInterfaceImpl.cpp | 7 +-
.../BufferViewFlowOpInterfaceImpl.cpp | 7 +-
.../BufferizableOpInterfaceImpl.cpp | 11 ++-
.../BufferizationTransformOps.cpp | 2 -
.../FuncBufferizableOpInterfaceImpl.cpp | 14 ++--
.../BufferDeallocationOpInterfaceImpl.cpp | 7 +-
.../BufferizableOpInterfaceImpl.cpp | 10 ++-
.../Func/Extensions/InlinerExtension.cpp | 11 ++-
.../Extensions/MeshShardingExtensions.cpp | 11 ++-
.../Func/TransformOps/FuncTransformOps.cpp | 2 -
.../GPU/TransformOps/GPUTransformOps.cpp | 2 -
.../BufferDeallocationOpInterfaceImpl.cpp | 7 +-
.../Linalg/IR/ValueBoundsOpInterfaceImpl.cpp | 13 ++--
.../Linalg/TransformOps/DialectExtension.cpp | 2 -
.../BufferizableOpInterfaceImpl.cpp | 19 +++--
.../Transforms/MeshShardingInterfaceImpl.cpp | 25 ++++---
.../Transforms/RuntimeOpVerification.cpp | 17 +++--
.../SubsetInsertionOpInterfaceImpl.cpp | 9 ++-
.../Linalg/Transforms/TilingInterfaceImpl.cpp | 11 ++-
.../BufferizableOpInterfaceImpl.cpp | 11 ++-
.../Dialect/MemRef/IR/MemRefMemorySlot.cpp | 9 +--
.../MemRef/IR/ValueBoundsOpInterfaceImpl.cpp | 25 +++----
.../TransformOps/MemRefTransformOps.cpp | 2 -
.../Transforms/AllocationOpInterfaceImpl.cpp | 13 ++--
.../BufferViewFlowOpInterfaceImpl.cpp | 8 +--
.../Transforms/RuntimeOpVerification.cpp | 26 ++++---
.../NVGPU/TransformOps/NVGPUTransformOps.cpp | 2 -
.../SCF/IR/ValueBoundsOpInterfaceImpl.cpp | 9 ++-
.../SCF/TransformOps/SCFTransformOps.cpp | 2 -
.../BufferDeallocationOpInterfaceImpl.cpp | 9 ++-
.../BufferizableOpInterfaceImpl.cpp | 23 +++---
.../BufferizableOpInterfaceImpl.cpp | 10 ++-
.../TransformOps/SparseTensorTransformOps.cpp | 2 -
.../BufferizableOpInterfaceImpl.cpp | 40 +++++------
.../IR/TensorInferTypeOpInterfaceImpl.cpp | 15 ++--
.../Tensor/IR/TensorTilingInterfaceImpl.cpp | 21 +++---
.../Tensor/IR/ValueBoundsOpInterfaceImpl.cpp | 25 +++----
.../TransformOps/TensorTransformOps.cpp | 23 +++---
.../BufferizableOpInterfaceImpl.cpp | 46 ++++++------
.../SubsetInsertionOpInterfaceImpl.cpp | 33 +++++----
.../Dialect/Tosa/IR/ShardingInterfaceImpl.cpp | 22 +++---
.../DebugExtension/DebugExtension.cpp | 2 -
.../Transform/IRDLExtension/IRDLExtension.cpp | 2 -
.../Transform/LoopExtension/LoopExtension.cpp | 2 -
.../Transform/PDLExtension/PDLExtension.cpp | 2 -
.../Vector/IR/ValueBoundsOpInterfaceImpl.cpp | 10 ++-
.../TransformOps/VectorTransformOps.cpp | 2 -
.../BufferizableOpInterfaceImpl.cpp | 16 ++---
.../Transforms/SubsetOpInterfaceImpl.cpp | 21 +++---
mlir/lib/Interfaces/CastInterfaces.cpp | 9 ++-
mlir/lib/Target/LLVM/NVVM/Target.cpp | 7 +-
mlir/lib/Target/LLVM/ROCDL/Target.cpp | 7 +-
.../Dialect/AMX/AMXToLLVMIRTranslation.cpp | 7 +-
.../ArmNeon/ArmNeonToLLVMIRTranslation.cpp | 1 -
.../ArmSME/ArmSMEToLLVMIRTranslation.cpp | 8 +--
.../ArmSVE/ArmSVEToLLVMIRTranslation.cpp | 8 +--
.../Builtin/BuiltinToLLVMIRTranslation.cpp | 7 +-
.../Dialect/GPU/GPUToLLVMIRTranslation.cpp | 7 +-
.../LLVMIR/Dialect/GPU/SelectObjectAttr.cpp | 7 +-
.../LLVMIR/LLVMIRToLLVMTranslation.cpp | 7 +-
.../LLVMIR/LLVMToLLVMIRTranslation.cpp | 7 +-
.../Dialect/NVVM/LLVMIRToNVVMTranslation.cpp | 7 +-
.../Dialect/NVVM/NVVMToLLVMIRTranslation.cpp | 7 +-
.../OpenACC/OpenACCToLLVMIRTranslation.cpp | 7 +-
.../OpenMP/OpenMPToLLVMIRTranslation.cpp | 7 +-
.../ROCDL/ROCDLToLLVMIRTranslation.cpp | 7 +-
.../Dialect/VCIX/VCIXToLLVMIRTranslation.cpp | 7 +-
.../X86VectorToLLVMIRTranslation.cpp | 1 -
mlir/lib/Target/SPIRV/Target.cpp | 7 +-
.../Test/TestFromLLVMIRTranslation.cpp | 1 -
.../Dialect/Test/TestToLLVMIRTranslation.cpp | 1 -
.../TestTransformDialectExtension.cpp | 2 -
.../TestTilingInterfaceTransformOps.cpp | 2 -
.../Transform/BuildOnlyExtensionTest.cpp | 2 -
mlir/unittests/IR/DialectTest.cpp | 60 +++++++++-------
mlir/unittests/IR/InterfaceAttachmentTest.cpp | 72 ++++++++-----------
93 files changed, 420 insertions(+), 575 deletions(-)
diff --git a/flang/lib/Optimizer/Dialect/FIRDialect.cpp b/flang/lib/Optimizer/Dialect/FIRDialect.cpp
index 3bf8b43c3572e..4b1dadaac6728 100644
--- a/flang/lib/Optimizer/Dialect/FIRDialect.cpp
+++ b/flang/lib/Optimizer/Dialect/FIRDialect.cpp
@@ -74,7 +74,7 @@ void fir::FIROpsDialect::initialize() {
// Register the FIRInlinerInterface to FIROpsDialect
void fir::addFIRInlinerExtension(mlir::DialectRegistry ®istry) {
registry.addExtension(
- "FIR_INLINER", +[](mlir::MLIRContext *ctx, fir::FIROpsDialect *dialect) {
+ +[](mlir::MLIRContext *ctx, fir::FIROpsDialect *dialect) {
dialect->addInterface<FIRInlinerInterface>();
});
}
@@ -90,7 +90,6 @@ void fir::addFIRInlinerExtension(mlir::DialectRegistry ®istry) {
// when more sophisticated translation is required.
void fir::addFIRToLLVMIRExtension(mlir::DialectRegistry ®istry) {
registry.addExtension(
- "FIR_LLVM_TRANSLATION",
+[](mlir::MLIRContext *ctx, fir::FIROpsDialect *dialect) {
dialect->addInterface<mlir::LLVMTranslationDialectInterface>();
});
diff --git a/mlir/examples/transform/Ch2/lib/MyExtension.cpp b/mlir/examples/transform/Ch2/lib/MyExtension.cpp
index a3f4f9aeb8106..68d538a098018 100644
--- a/mlir/examples/transform/Ch2/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch2/lib/MyExtension.cpp
@@ -36,9 +36,6 @@ class MyExtension
// dialect definitions. List individual operations and dependent dialects
// here.
void init();
-
- // Declare a unique ID for this extension.
- static constexpr llvm::StringRef extensionID = "CH2_TRANSFORM";
};
void MyExtension::init() {
diff --git a/mlir/examples/transform/Ch3/lib/MyExtension.cpp b/mlir/examples/transform/Ch3/lib/MyExtension.cpp
index 0eeeb3b4071d3..f7a99423a52ee 100644
--- a/mlir/examples/transform/Ch3/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch3/lib/MyExtension.cpp
@@ -42,9 +42,6 @@ class MyExtension
// dialect definitions. List individual operations and dependent dialects
// here.
void init();
-
- // Declare a unique ID for this extension.
- static constexpr llvm::StringRef extensionID = "CH3_TRANSFORM";
};
void MyExtension::init() {
diff --git a/mlir/examples/transform/Ch4/lib/MyExtension.cpp b/mlir/examples/transform/Ch4/lib/MyExtension.cpp
index 1d440f588e60b..38c8ca1125a24 100644
--- a/mlir/examples/transform/Ch4/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch4/lib/MyExtension.cpp
@@ -38,9 +38,6 @@ class MyExtension
// dialect definitions. List individual operations and dependent dialects
// here.
void init();
-
- // Declare a unique ID for this extension.
- static constexpr llvm::StringRef extensionID = "CH4_TRANSFORM";
};
void MyExtension::init() {
diff --git a/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp b/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp
index d2794a24cd05a..d882f1184f457 100644
--- a/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp
+++ b/mlir/lib/Conversion/ArithToLLVM/ArithToLLVM.cpp
@@ -510,10 +510,9 @@ struct ArithToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
void mlir::arith::registerConvertArithToLLVMInterface(
DialectRegistry ®istry) {
- registry.addExtension(
- "ARITH_TO_LLVM", +[](MLIRContext *ctx, arith::ArithDialect *dialect) {
- dialect->addInterfaces<ArithToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, arith::ArithDialect *dialect) {
+ dialect->addInterfaces<ArithToLLVMDialectInterface>();
+ });
}
//===----------------------------------------------------------------------===//
diff --git a/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp b/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp
index 3701d62efba50..0a3c3a330ff69 100644
--- a/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp
+++ b/mlir/lib/Conversion/ComplexToLLVM/ComplexToLLVM.cpp
@@ -385,7 +385,6 @@ struct ComplexToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
void mlir::registerConvertComplexToLLVMInterface(DialectRegistry ®istry) {
registry.addExtension(
- "COMPLEX_TO_LLVM",
+[](MLIRContext *ctx, complex::ComplexDialect *dialect) {
dialect->addInterfaces<ComplexToLLVMDialectInterface>();
});
diff --git a/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp b/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp
index 4a154554c30bc..b8e5aec25286d 100644
--- a/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp
+++ b/mlir/lib/Conversion/ControlFlowToLLVM/ControlFlowToLLVM.cpp
@@ -273,8 +273,7 @@ struct ControlFlowToLLVMDialectInterface
void mlir::cf::registerConvertControlFlowToLLVMInterface(
DialectRegistry ®istry) {
- registry.addExtension(
- "CF_TO_LLVM", +[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
- dialect->addInterfaces<ControlFlowToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
+ dialect->addInterfaces<ControlFlowToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp b/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
index c55607f5e7c60..6135117348a5b 100644
--- a/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
+++ b/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
@@ -15,7 +15,6 @@
#include "mlir/Pass/Pass.h"
#include "mlir/Rewrite/FrozenRewritePatternSet.h"
#include "mlir/Transforms/DialectConversion.h"
-#include "llvm/ADT/StringRef.h"
#include <memory>
#define DEBUG_TYPE "convert-to-llvm"
@@ -55,10 +54,6 @@ class LoadDependentDialectExtension : public DialectExtensionBase {
std::unique_ptr<DialectExtensionBase> clone() const final {
return std::make_unique<LoadDependentDialectExtension>(*this);
}
-
- /// Unique ID
- constexpr static llvm::StringRef extensionID =
- "CONVERT_TO_LLVM_LOAD_DEPENDENT";
};
/// This is a generic pass to convert to LLVM, it uses the
diff --git a/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp b/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp
index 5994a4c5fdf0f..c1f6d8bc5b361 100644
--- a/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp
+++ b/mlir/lib/Conversion/FuncToLLVM/FuncToLLVM.cpp
@@ -786,8 +786,7 @@ struct FuncToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::registerConvertFuncToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(
- "FUNC_TO_LLVM", +[](MLIRContext *ctx, func::FuncDialect *dialect) {
- dialect->addInterfaces<FuncToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, func::FuncDialect *dialect) {
+ dialect->addInterfaces<FuncToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp b/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp
index e045605475b78..9d8a5d8a0e1c0 100644
--- a/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp
+++ b/mlir/lib/Conversion/IndexToLLVM/IndexToLLVM.cpp
@@ -392,8 +392,7 @@ struct IndexToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
void mlir::index::registerConvertIndexToLLVMInterface(
DialectRegistry ®istry) {
- registry.addExtension(
- "INDEX_TO_LLVM", +[](MLIRContext *ctx, index::IndexDialect *dialect) {
- dialect->addInterfaces<IndexToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, index::IndexDialect *dialect) {
+ dialect->addInterfaces<IndexToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp b/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp
index f43ce2800e310..23e957288eb95 100644
--- a/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp
+++ b/mlir/lib/Conversion/MathToLLVM/MathToLLVM.cpp
@@ -356,8 +356,7 @@ struct MathToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::registerConvertMathToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(
- "MATH_TO_LLVM", +[](MLIRContext *ctx, math::MathDialect *dialect) {
- dialect->addInterfaces<MathToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, math::MathDialect *dialect) {
+ dialect->addInterfaces<MathToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp b/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp
index d0d4b159ddd57..054827d40f0f3 100644
--- a/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp
+++ b/mlir/lib/Conversion/MemRefToLLVM/MemRefToLLVM.cpp
@@ -1753,8 +1753,7 @@ struct MemRefToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::registerConvertMemRefToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(
- "MEMREF_TO_LLVM", +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- dialect->addInterfaces<MemRefToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ dialect->addInterfaces<MemRefToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp b/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp
index c6b856b84b10b..662ee9e483bc5 100644
--- a/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp
+++ b/mlir/lib/Conversion/NVVMToLLVM/NVVMToLLVM.cpp
@@ -113,8 +113,7 @@ void mlir::populateNVVMToLLVMConversionPatterns(RewritePatternSet &patterns) {
}
void mlir::registerConvertNVVMToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(
- "NVVM_TO_LLVM", +[](MLIRContext *ctx, NVVMDialect *dialect) {
- dialect->addInterfaces<NVVMToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, NVVMDialect *dialect) {
+ dialect->addInterfaces<NVVMToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp b/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp
index 12150d2e1977b..0051333a35dcd 100644
--- a/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp
+++ b/mlir/lib/Conversion/UBToLLVM/UBToLLVM.cpp
@@ -119,8 +119,7 @@ struct UBToLLVMDialectInterface : public ConvertToLLVMPatternInterface {
} // namespace
void mlir::ub::registerConvertUBToLLVMInterface(DialectRegistry ®istry) {
- registry.addExtension(
- "UB_TO_LLVM", +[](MLIRContext *ctx, ub::UBDialect *dialect) {
- dialect->addInterfaces<UBToLLVMDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, ub::UBDialect *dialect) {
+ dialect->addInterfaces<UBToLLVMDialectInterface>();
+ });
}
diff --git a/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp
index 2fe39c757fcd3..82a9fb0d49088 100644
--- a/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Affine/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -96,12 +96,11 @@ struct AffineMaxOpInterface
void mlir::affine::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "AFFINE_VALUE_BOUNDS", +[](MLIRContext *ctx, AffineDialect *dialect) {
- AffineApplyOp::attachInterface<AffineApplyOpInterface>(*ctx);
- AffineMaxOp::attachInterface<AffineMaxOpInterface>(*ctx);
- AffineMinOp::attachInterface<AffineMinOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, AffineDialect *dialect) {
+ AffineApplyOp::attachInterface<AffineApplyOpInterface>(*ctx);
+ AffineMaxOp::attachInterface<AffineMaxOpInterface>(*ctx);
+ AffineMinOp::attachInterface<AffineMinOpInterface>(*ctx);
+ });
}
FailureOr<int64_t>
diff --git a/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp b/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
index c6f5cfd47b046..6457655cfe416 100644
--- a/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
+++ b/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
@@ -167,8 +167,6 @@ class AffineTransformDialectExtension
#include "mlir/Dialect/Affine/TransformOps/AffineTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "AFFINE_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp
index 981a52090ad66..7cfcc4180539c 100644
--- a/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -150,13 +150,11 @@ struct SelectOpInterface
void mlir::arith::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "ARITH_VALUE_BOUNDS",
- +[](MLIRContext *ctx, arith::ArithDialect *dialect) {
- arith::AddIOp::attachInterface<arith::AddIOpInterface>(*ctx);
- arith::ConstantOp::attachInterface<arith::ConstantOpInterface>(*ctx);
- arith::SubIOp::attachInterface<arith::SubIOpInterface>(*ctx);
- arith::MulIOp::attachInterface<arith::MulIOpInterface>(*ctx);
- arith::SelectOp::attachInterface<arith::SelectOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, arith::ArithDialect *dialect) {
+ arith::AddIOp::attachInterface<arith::AddIOpInterface>(*ctx);
+ arith::ConstantOp::attachInterface<arith::ConstantOpInterface>(*ctx);
+ arith::SubIOp::attachInterface<arith::SubIOpInterface>(*ctx);
+ arith::MulIOp::attachInterface<arith::MulIOpInterface>(*ctx);
+ arith::SelectOp::attachInterface<arith::SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index c17f6aacd88f8..f2e7732e8ea4a 100644
--- a/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -79,8 +79,7 @@ struct SelectOpInterface
void mlir::arith::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "ARITH_BUFFER_DEALLOC", +[](MLIRContext *ctx, ArithDialect *dialect) {
- SelectOp::attachInterface<SelectOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, ArithDialect *dialect) {
+ SelectOp::attachInterface<SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp
index 9f03fa947085f..9df9df86b64fb 100644
--- a/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/Transforms/BufferViewFlowOpInterfaceImpl.cpp
@@ -38,8 +38,7 @@ struct SelectOpInterface
void arith::registerBufferViewFlowOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "ARITH_BUFFER_FLOW", +[](MLIRContext *ctx, arith::ArithDialect *dialect) {
- SelectOp::attachInterface<SelectOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, arith::ArithDialect *dialect) {
+ SelectOp::attachInterface<SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp
index 74b55c124e480..5e69a98db8f1e 100644
--- a/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Arith/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -206,10 +206,9 @@ struct SelectOpInterface
void mlir::arith::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "ARITH_BUFFERIZATION", +[](MLIRContext *ctx, ArithDialect *dialect) {
- ConstantOp::attachInterface<ConstantOpInterface>(*ctx);
- IndexCastOp::attachInterface<IndexCastOpInterface>(*ctx);
- SelectOp::attachInterface<SelectOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, ArithDialect *dialect) {
+ ConstantOp::attachInterface<ConstantOpInterface>(*ctx);
+ IndexCastOp::attachInterface<IndexCastOpInterface>(*ctx);
+ SelectOp::attachInterface<SelectOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp b/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
index d0993a5b85e8d..e10c7bd914e35 100644
--- a/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
+++ b/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
@@ -161,8 +161,6 @@ class BufferizationTransformDialectExtension
#include "mlir/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "BUFFERIZATION_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp
index 6f2ec90bc866e..053ea7935260a 100644
--- a/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Bufferization/Transforms/FuncBufferizableOpInterfaceImpl.cpp
@@ -82,8 +82,7 @@ getBufferizedFunctionArgType(FuncOp funcOp, int64_t index,
/// Return the FuncOp called by `callOp`.
static FuncOp getCalledFunction(CallOpInterface callOp) {
- SymbolRefAttr sym =
- llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
+ SymbolRefAttr sym = llvm::dyn_cast_if_present<SymbolRefAttr>(callOp.getCallableForCallee());
if (!sym)
return nullptr;
return dyn_cast_or_null<FuncOp>(
@@ -490,10 +489,9 @@ struct FuncOpInterface
void mlir::bufferization::func_ext::
registerBufferizableOpInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(
- "FUNC_BUFFERIZATION", +[](MLIRContext *ctx, func::FuncDialect *dialect) {
- func::CallOp::attachInterface<func_ext::CallOpInterface>(*ctx);
- func::FuncOp::attachInterface<func_ext::FuncOpInterface>(*ctx);
- func::ReturnOp::attachInterface<func_ext::ReturnOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, func::FuncDialect *dialect) {
+ func::CallOp::attachInterface<func_ext::CallOpInterface>(*ctx);
+ func::FuncOp::attachInterface<func_ext::FuncOpInterface>(*ctx);
+ func::ReturnOp::attachInterface<func_ext::ReturnOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index f58f21b35547d..89546da428fa2 100644
--- a/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/ControlFlow/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -157,8 +157,7 @@ struct CondBranchOpInterface
void mlir::cf::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "CF_BUFFER_DEALLOC", +[](MLIRContext *ctx, ControlFlowDialect *dialect) {
- CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, ControlFlowDialect *dialect) {
+ CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp
index 0db95f92dd95b..72f4a1a4f4c66 100644
--- a/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/ControlFlow/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -63,10 +63,8 @@ struct CondBranchOpInterface
void mlir::cf::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "CF_BUFFERIZATION",
- +[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
- cf::BranchOp::attachInterface<BranchOpInterface>(*ctx);
- cf::CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, cf::ControlFlowDialect *dialect) {
+ cf::BranchOp::attachInterface<BranchOpInterface>(*ctx);
+ cf::CondBranchOp::attachInterface<CondBranchOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp b/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp
index 43e998fbc04f8..719a74a29a622 100644
--- a/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp
+++ b/mlir/lib/Dialect/Func/Extensions/InlinerExtension.cpp
@@ -80,11 +80,10 @@ struct FuncInlinerInterface : public DialectInlinerInterface {
//===----------------------------------------------------------------------===//
void mlir::func::registerInlinerExtension(DialectRegistry ®istry) {
- registry.addExtension(
- "FUNC_INLINER", +[](MLIRContext *ctx, func::FuncDialect *dialect) {
- dialect->addInterfaces<FuncInlinerInterface>();
+ registry.addExtension(+[](MLIRContext *ctx, func::FuncDialect *dialect) {
+ dialect->addInterfaces<FuncInlinerInterface>();
- // The inliner extension relies on the ControlFlow dialect.
- ctx->getOrLoadDialect<cf::ControlFlowDialect>();
- });
+ // The inliner extension relies on the ControlFlow dialect.
+ ctx->getOrLoadDialect<cf::ControlFlowDialect>();
+ });
}
diff --git a/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp b/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp
index 833362041848b..da508cc95bfe1 100644
--- a/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp
+++ b/mlir/lib/Dialect/Func/Extensions/MeshShardingExtensions.cpp
@@ -14,12 +14,11 @@
namespace mlir::func {
void registerShardingInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(
- "FUNC_SHARDING", +[](MLIRContext *ctx, FuncDialect *dialect) {
- ReturnOp::attachInterface<
- mesh::IndependentParallelIteratorDomainShardingInterface<ReturnOp>>(
- *ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, FuncDialect *dialect) {
+ ReturnOp::attachInterface<
+ mesh::IndependentParallelIteratorDomainShardingInterface<ReturnOp>>(
+ *ctx);
+ });
}
} // namespace mlir::func
diff --git a/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp b/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
index 711a2d9685f6d..b632b25d0cc67 100644
--- a/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
+++ b/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
@@ -246,8 +246,6 @@ class FuncTransformDialectExtension
#include "mlir/Dialect/Func/TransformOps/FuncTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "FUNC_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp b/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
index d6a044698746f..3661c5dea4525 100644
--- a/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
+++ b/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
@@ -933,8 +933,6 @@ class GPUTransformDialectExtension
#include "mlir/Dialect/GPU/TransformOps/GPUTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "GPU_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index ddad79f173628..6ccc0a26426c1 100644
--- a/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/GPU/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -31,8 +31,7 @@ struct GPUTerminatorOpInterface
void mlir::gpu::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "GPU_BUFFER_DEALLOC", +[](MLIRContext *ctx, GPUDialect *dialect) {
- gpu::TerminatorOp::attachInterface<GPUTerminatorOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, GPUDialect *dialect) {
+ gpu::TerminatorOp::attachInterface<GPUTerminatorOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp
index 390d6582c487e..f56ef485069f8 100644
--- a/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -53,12 +53,9 @@ struct IndexOpInterface
void mlir::linalg::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "LINALG_VALUE_BOUNDS",
- +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- IndexOp::attachInterface<IndexOpInterface>(*ctx);
- // Note: ValueBoundsOpInterface implementation is not required for ops
- // that implement `DestinationStyleOpInterface` (for querying shaped
- // OpResults).
- });
+ registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ IndexOp::attachInterface<IndexOpInterface>(*ctx);
+ // Note: ValueBoundsOpInterface implementation is not required for ops that
+ // implement `DestinationStyleOpInterface` (for querying shaped OpResults).
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp b/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
index 01c6814bc2082..f4244ca962232 100644
--- a/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
+++ b/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
@@ -52,8 +52,6 @@ class LinalgTransformDialectExtension
#include "mlir/Dialect/Linalg/TransformOps/LinalgMatchOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "LINALG_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp
index dc3bea56da8de..be158af09d398 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -195,18 +195,15 @@ struct SoftmaxOpInterface
void mlir::linalg::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "LINALG_BUFFERIZATION",
- +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- // Register all Linalg structured ops. `LinalgOp` is an interface and it
- // is not possible to attach an external interface to an existing
- // interface. Therefore, attach the `BufferizableOpInterface` to all ops
- // one-by-one.
- LinalgOpInterfaceHelper<
+ registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ // Register all Linalg structured ops. `LinalgOp` is an interface and it is
+ // not possible to attach an external interface to an existing interface.
+ // Therefore, attach the `BufferizableOpInterface` to all ops one-by-one.
+ LinalgOpInterfaceHelper<
#define GET_OP_LIST
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >::registerOpInterface(ctx);
+ >::registerOpInterface(ctx);
- SoftmaxOp::attachInterface<SoftmaxOpInterface>(*ctx);
- });
+ SoftmaxOp::attachInterface<SoftmaxOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp
index afb5ffcf31613..36b6088b83cc2 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/MeshShardingInterfaceImpl.cpp
@@ -347,21 +347,20 @@ static void registerAll(MLIRContext *ctx) {
}
void registerMeshShardingInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(
- "LINALG_SHARDING", +[](MLIRContext *ctx, LinalgDialect *dialect) {
- DialectRegistry registry;
- registry.insert<affine::AffineDialect, arith::ArithDialect,
- scf::SCFDialect, tensor::TensorDialect>();
- ctx->appendDialectRegistry(registry);
- for (StringRef name : registry.getDialectNames())
- ctx->getOrLoadDialect(name);
-
- registerOne<linalg::GenericOp>(ctx);
- registerAll<
+ registry.addExtension(+[](MLIRContext *ctx, LinalgDialect *dialect) {
+ DialectRegistry registry;
+ registry.insert<affine::AffineDialect, arith::ArithDialect, scf::SCFDialect,
+ tensor::TensorDialect>();
+ ctx->appendDialectRegistry(registry);
+ for (StringRef name : registry.getDialectNames())
+ ctx->getOrLoadDialect(name);
+
+ registerOne<linalg::GenericOp>(ctx);
+ registerAll<
#define GET_OP_LIST
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >(ctx);
- });
+ >(ctx);
+ });
}
} // namespace mlir::linalg
diff --git a/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp b/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp
index 6b3ab1a42bd99..b30182dc84079 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/RuntimeOpVerification.cpp
@@ -121,16 +121,15 @@ void attachInterface(MLIRContext *ctx) {
void mlir::linalg::registerRuntimeVerifiableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "LINALG_RUNTIME_VERIFICATION", +[](MLIRContext *ctx, LinalgDialect *) {
- attachInterface<
+ registry.addExtension(+[](MLIRContext *ctx, LinalgDialect *) {
+ attachInterface<
#define GET_OP_LIST
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >(ctx);
+ >(ctx);
- // Load additional dialects of which ops may get created.
- ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
- cf::ControlFlowDialect, index::IndexDialect,
- tensor::TensorDialect>();
- });
+ // Load additional dialects of which ops may get created.
+ ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
+ cf::ControlFlowDialect, index::IndexDialect,
+ tensor::TensorDialect>();
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp
index 04e4b86f8a855..6fcfa05468eea 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/SubsetInsertionOpInterfaceImpl.cpp
@@ -74,9 +74,8 @@ struct LinalgCopyOpInterface
void mlir::linalg::registerSubsetOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "LINALG_SUBSET", +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- linalg::CopyOp::attachInterface<LinalgCopyOpSubsetOpInterface>(*ctx);
- linalg::CopyOp::attachInterface<LinalgCopyOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ linalg::CopyOp::attachInterface<LinalgCopyOpSubsetOpInterface>(*ctx);
+ linalg::CopyOp::attachInterface<LinalgCopyOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp b/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp
index 9cd4b4c6adbd9..2133458efe74c 100644
--- a/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Linalg/Transforms/TilingInterfaceImpl.cpp
@@ -506,11 +506,10 @@ static void registerAll(MLIRContext *ctx) {
void mlir::linalg::registerTilingInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "LINALG_TILING", +[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
- registerOne<linalg::GenericOp>(ctx);
- registerAll<
+ registry.addExtension(+[](MLIRContext *ctx, linalg::LinalgDialect *dialect) {
+ registerOne<linalg::GenericOp>(ctx);
+ registerAll<
#include "mlir/Dialect/Linalg/IR/LinalgStructuredOps.cpp.inc"
- >(ctx);
- });
+ >(ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp
index a07a412f11e85..926d580ac7852 100644
--- a/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MLProgram/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -149,12 +149,11 @@ struct GlobalStoreOpInterface
} // namespace
void registerBufferizableOpInterfaceExternalModels(DialectRegistry ®istry) {
- registry.addExtension(
- "MLPROGRAM_BUFFERIZATION", +[](MLIRContext *ctx, MLProgramDialect *) {
- GlobalOp::attachInterface<GlobalOpInterface>(*ctx);
- GlobalLoadOp::attachInterface<GlobalLoadOpInterface>(*ctx);
- GlobalStoreOp::attachInterface<GlobalStoreOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, MLProgramDialect *) {
+ GlobalOp::attachInterface<GlobalOpInterface>(*ctx);
+ GlobalLoadOp::attachInterface<GlobalLoadOpInterface>(*ctx);
+ GlobalStoreOp::attachInterface<GlobalStoreOpInterface>(*ctx);
+ });
}
} // namespace ml_program
} // namespace mlir
diff --git a/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp b/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp
index 7c21858adc5fb..f630c48cdcaa1 100644
--- a/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp
+++ b/mlir/lib/Dialect/MemRef/IR/MemRefMemorySlot.cpp
@@ -347,10 +347,7 @@ struct MemRefDestructurableTypeExternalModel
//===----------------------------------------------------------------------===//
void mlir::memref::registerMemorySlotExternalModels(DialectRegistry ®istry) {
- registry.addExtension(
- "MEMREF_MEMORY_SLOT_MODELS",
- +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- MemRefType::attachInterface<MemRefDestructurableTypeExternalModel>(
- *ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ MemRefType::attachInterface<MemRefDestructurableTypeExternalModel>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp
index fde96046b3ca2..daec22cf6ebdc 100644
--- a/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MemRef/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -115,18 +115,15 @@ struct SubViewOpInterface
void mlir::memref::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "MEMREF_VALUE_BOUNDS",
- +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- memref::AllocOp::attachInterface<
- memref::AllocOpInterface<memref::AllocOp>>(*ctx);
- memref::AllocaOp::attachInterface<
- memref::AllocOpInterface<memref::AllocaOp>>(*ctx);
- memref::CastOp::attachInterface<memref::CastOpInterface>(*ctx);
- memref::DimOp::attachInterface<memref::DimOpInterface>(*ctx);
- memref::GetGlobalOp::attachInterface<memref::GetGlobalOpInterface>(
- *ctx);
- memref::RankOp::attachInterface<memref::RankOpInterface>(*ctx);
- memref::SubViewOp::attachInterface<memref::SubViewOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ memref::AllocOp::attachInterface<memref::AllocOpInterface<memref::AllocOp>>(
+ *ctx);
+ memref::AllocaOp::attachInterface<
+ memref::AllocOpInterface<memref::AllocaOp>>(*ctx);
+ memref::CastOp::attachInterface<memref::CastOpInterface>(*ctx);
+ memref::DimOp::attachInterface<memref::DimOpInterface>(*ctx);
+ memref::GetGlobalOp::attachInterface<memref::GetGlobalOpInterface>(*ctx);
+ memref::RankOp::attachInterface<memref::RankOpInterface>(*ctx);
+ memref::SubViewOp::attachInterface<memref::SubViewOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp b/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
index 20f96a018d330..8469e84c668cb 100644
--- a/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
+++ b/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
@@ -323,8 +323,6 @@ class MemRefTransformDialectExtension
#include "mlir/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "MEMREF_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp
index 2cde261180e9e..c433415944323 100644
--- a/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MemRef/Transforms/AllocationOpInterfaceImpl.cpp
@@ -60,11 +60,10 @@ struct DefaultReallocationInterface
void mlir::memref::registerAllocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "MEMREF_ALLOC", +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- memref::AllocOp::attachInterface<DefaultAllocationInterface>(*ctx);
- memref::AllocaOp::attachInterface<
- DefaultAutomaticAllocationHoistingInterface>(*ctx);
- memref::ReallocOp::attachInterface<DefaultReallocationInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ memref::AllocOp::attachInterface<DefaultAllocationInterface>(*ctx);
+ memref::AllocaOp::attachInterface<
+ DefaultAutomaticAllocationHoistingInterface>(*ctx);
+ memref::ReallocOp::attachInterface<DefaultReallocationInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp b/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp
index 5076c0be916c9..bbb269bd00161 100644
--- a/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/MemRef/Transforms/BufferViewFlowOpInterfaceImpl.cpp
@@ -42,9 +42,7 @@ struct ReallocOpInterface
void memref::registerBufferViewFlowOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "MEMREF_BUFFER_FLOW",
- +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- ReallocOp::attachInterface<ReallocOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ ReallocOp::attachInterface<ReallocOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp b/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp
index 80fc16d058b41..450bfa0cec0c7 100644
--- a/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp
+++ b/mlir/lib/Dialect/MemRef/Transforms/RuntimeOpVerification.cpp
@@ -333,18 +333,16 @@ struct ExpandShapeOpInterface
void mlir::memref::registerRuntimeVerifiableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "MEMREF_RUNTIME_VERIFICATION",
- +[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
- CastOp::attachInterface<CastOpInterface>(*ctx);
- ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
- LoadOp::attachInterface<LoadStoreOpInterface<LoadOp>>(*ctx);
- ReinterpretCastOp::attachInterface<ReinterpretCastOpInterface>(*ctx);
- StoreOp::attachInterface<LoadStoreOpInterface<StoreOp>>(*ctx);
- SubViewOp::attachInterface<SubViewOpInterface>(*ctx);
-
- // Load additional dialects of which ops may get created.
- ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
- cf::ControlFlowDialect>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, memref::MemRefDialect *dialect) {
+ CastOp::attachInterface<CastOpInterface>(*ctx);
+ ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
+ LoadOp::attachInterface<LoadStoreOpInterface<LoadOp>>(*ctx);
+ ReinterpretCastOp::attachInterface<ReinterpretCastOpInterface>(*ctx);
+ StoreOp::attachInterface<LoadStoreOpInterface<StoreOp>>(*ctx);
+ SubViewOp::attachInterface<SubViewOpInterface>(*ctx);
+
+ // Load additional dialects of which ops may get created.
+ ctx->loadDialect<affine::AffineDialect, arith::ArithDialect,
+ cf::ControlFlowDialect>();
+ });
}
diff --git a/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp b/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
index 8315ad2f5892b..7d3d868b326c6 100644
--- a/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
+++ b/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
@@ -1146,8 +1146,6 @@ class NVGPUTransformDialectExtension
#include "mlir/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "NVGPU_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp
index c6ede24e605df..17a1c016ea16d 100644
--- a/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SCF/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -159,9 +159,8 @@ struct IfOpInterface
void mlir::scf::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "SCF_VALUE_BOUNDS", +[](MLIRContext *ctx, scf::SCFDialect *dialect) {
- scf::ForOp::attachInterface<scf::ForOpInterface>(*ctx);
- scf::IfOp::attachInterface<scf::IfOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, scf::SCFDialect *dialect) {
+ scf::ForOp::attachInterface<scf::ForOpInterface>(*ctx);
+ scf::IfOp::attachInterface<scf::IfOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp b/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
index c56d1e46bc0fd..c4a55c302d0a3 100644
--- a/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
+++ b/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
@@ -624,8 +624,6 @@ class SCFTransformDialectExtension
#include "mlir/Dialect/SCF/TransformOps/SCFTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "SCF_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp b/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp
index bea0728e16987..24fbc1dca8361 100644
--- a/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SCF/Transforms/BufferDeallocationOpInterfaceImpl.cpp
@@ -75,9 +75,8 @@ struct ReduceReturnOpInterface
void mlir::scf::registerBufferDeallocationOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "SCF_BUFFER_DEALLOC", +[](MLIRContext *ctx, SCFDialect *dialect) {
- InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
- ReduceReturnOp::attachInterface<ReduceReturnOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, SCFDialect *dialect) {
+ InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
+ ReduceReturnOp::attachInterface<ReduceReturnOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp
index 0aa9fd4385bb5..cf40443ff3839 100644
--- a/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SCF/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -1352,16 +1352,15 @@ struct InParallelOpInterface
void mlir::scf::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "SCF_BUFFERIZATION", +[](MLIRContext *ctx, scf::SCFDialect *dialect) {
- ConditionOp::attachInterface<ConditionOpInterface>(*ctx);
- ExecuteRegionOp::attachInterface<ExecuteRegionOpInterface>(*ctx);
- ForOp::attachInterface<ForOpInterface>(*ctx);
- IfOp::attachInterface<IfOpInterface>(*ctx);
- IndexSwitchOp::attachInterface<IndexSwitchOpInterface>(*ctx);
- ForallOp::attachInterface<ForallOpInterface>(*ctx);
- InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
- WhileOp::attachInterface<WhileOpInterface>(*ctx);
- YieldOp::attachInterface<YieldOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, scf::SCFDialect *dialect) {
+ ConditionOp::attachInterface<ConditionOpInterface>(*ctx);
+ ExecuteRegionOp::attachInterface<ExecuteRegionOpInterface>(*ctx);
+ ForOp::attachInterface<ForOpInterface>(*ctx);
+ IfOp::attachInterface<IfOpInterface>(*ctx);
+ IndexSwitchOp::attachInterface<IndexSwitchOpInterface>(*ctx);
+ ForallOp::attachInterface<ForallOpInterface>(*ctx);
+ InParallelOp::attachInterface<InParallelOpInterface>(*ctx);
+ WhileOp::attachInterface<WhileOpInterface>(*ctx);
+ YieldOp::attachInterface<YieldOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp
index b82c6624d0c32..66a2e45001781 100644
--- a/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Shape/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -137,10 +137,8 @@ struct AssumingYieldOpInterface
void mlir::shape::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "SHAPE_BUFFERIZATION",
- +[](MLIRContext *ctx, shape::ShapeDialect *dialect) {
- shape::AssumingOp::attachInterface<AssumingOpInterface>(*ctx);
- shape::AssumingYieldOp::attachInterface<AssumingYieldOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, shape::ShapeDialect *dialect) {
+ shape::AssumingOp::attachInterface<AssumingOpInterface>(*ctx);
+ shape::AssumingYieldOp::attachInterface<AssumingYieldOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp b/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
index e2b0c77f5f5a0..ca19259ebffa6 100644
--- a/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
+++ b/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
@@ -45,8 +45,6 @@ class SparseTensorTransformDialectExtension
#include "mlir/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "SPARSE_TENSOR_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp
index aca1a13a8b913..7734d1d258453 100644
--- a/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/SparseTensor/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -326,26 +326,22 @@ struct ToValuesOpInterface
void mlir::sparse_tensor::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "SPARSE_TENSOR_BUFFERIZATION",
- +[](MLIRContext *ctx, sparse_tensor::SparseTensorDialect *dialect) {
- sparse_tensor::ConcatenateOp::attachInterface<ConcatenateOpInterface>(
- *ctx);
- sparse_tensor::ConvertOp::attachInterface<ConvertOpInterface>(*ctx);
- sparse_tensor::LoadOp::attachInterface<LoadOpInterface>(*ctx);
- sparse_tensor::NewOp::attachInterface<NewOpInterface>(*ctx);
- sparse_tensor::NumberOfEntriesOp::attachInterface<
- NumberOfEntriesOpInterface>(*ctx);
- sparse_tensor::AssembleOp::attachInterface<AssembleOpInterface>(*ctx);
- sparse_tensor::DisassembleOp::attachInterface<DisassembleOpInterface>(
- *ctx);
- sparse_tensor::ForeachOp::attachInterface<ForeachOpInterface>(*ctx);
- sparse_tensor::ToCoordinatesBufferOp::attachInterface<
- ToCoordinatesBufferOpInterface>(*ctx);
- sparse_tensor::ToCoordinatesOp::attachInterface<
- ToCoordinatesOpInterface>(*ctx);
- sparse_tensor::ToPositionsOp::attachInterface<ToPositionsOpInterface>(
- *ctx);
- sparse_tensor::ToValuesOp::attachInterface<ToValuesOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx,
+ sparse_tensor::SparseTensorDialect *dialect) {
+ sparse_tensor::ConcatenateOp::attachInterface<ConcatenateOpInterface>(*ctx);
+ sparse_tensor::ConvertOp::attachInterface<ConvertOpInterface>(*ctx);
+ sparse_tensor::LoadOp::attachInterface<LoadOpInterface>(*ctx);
+ sparse_tensor::NewOp::attachInterface<NewOpInterface>(*ctx);
+ sparse_tensor::NumberOfEntriesOp::attachInterface<
+ NumberOfEntriesOpInterface>(*ctx);
+ sparse_tensor::AssembleOp::attachInterface<AssembleOpInterface>(*ctx);
+ sparse_tensor::DisassembleOp::attachInterface<DisassembleOpInterface>(*ctx);
+ sparse_tensor::ForeachOp::attachInterface<ForeachOpInterface>(*ctx);
+ sparse_tensor::ToCoordinatesBufferOp::attachInterface<
+ ToCoordinatesBufferOpInterface>(*ctx);
+ sparse_tensor::ToCoordinatesOp::attachInterface<ToCoordinatesOpInterface>(
+ *ctx);
+ sparse_tensor::ToPositionsOp::attachInterface<ToPositionsOpInterface>(*ctx);
+ sparse_tensor::ToValuesOp::attachInterface<ToValuesOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp
index 48b1dc2bdaa6c..7ff435a033985 100644
--- a/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/IR/TensorInferTypeOpInterfaceImpl.cpp
@@ -201,12 +201,11 @@ struct ReifyPadOp
void mlir::tensor::registerInferTypeOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENSOR_INFER_TYPE", +[](MLIRContext *ctx, TensorDialect *dialect) {
- ExpandShapeOp::attachInterface<
- ReifyExpandOrCollapseShapeOp<tensor::ExpandShapeOp>>(*ctx);
- CollapseShapeOp::attachInterface<
- ReifyExpandOrCollapseShapeOp<tensor::CollapseShapeOp>>(*ctx);
- PadOp::attachInterface<ReifyPadOp>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, TensorDialect *dialect) {
+ ExpandShapeOp::attachInterface<
+ ReifyExpandOrCollapseShapeOp<tensor::ExpandShapeOp>>(*ctx);
+ CollapseShapeOp::attachInterface<
+ ReifyExpandOrCollapseShapeOp<tensor::CollapseShapeOp>>(*ctx);
+ PadOp::attachInterface<ReifyPadOp>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp
index 88801be553148..9b2a97eb2b006 100644
--- a/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/IR/TensorTilingInterfaceImpl.cpp
@@ -785,20 +785,17 @@ FailureOr<TilingResult> tensor::bubbleUpPadSlice(OpBuilder &b,
void mlir::tensor::registerTilingInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENSOR_TILING", +[](MLIRContext *ctx, TensorDialect *dialect) {
- tensor::PadOp::attachInterface<PadOpTiling>(*ctx);
- tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
- tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, TensorDialect *dialect) {
+ tensor::PadOp::attachInterface<PadOpTiling>(*ctx);
+ tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
+ tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
+ });
}
void mlir::tensor::registerTilingInterfaceExternalModelsForPackUnPackOps(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENOR_TILING_PACK_UNPACK",
- +[](MLIRContext *ctx, TensorDialect *dialect) {
- tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
- tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, TensorDialect *dialect) {
+ tensor::PackOp::attachInterface<PackOpTiling>(*ctx);
+ tensor::UnPackOp::attachInterface<UnPackOpTiling>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp
index 3b68822a4813d..06f2c16406d3c 100644
--- a/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -114,18 +114,15 @@ struct RankOpInterface
void mlir::tensor::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENSOR_VALUE_BOUNDS",
- +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- tensor::CastOp::attachInterface<tensor::CastOpInterface>(*ctx);
- tensor::DimOp::attachInterface<tensor::DimOpInterface>(*ctx);
- tensor::EmptyOp::attachInterface<tensor::EmptyOpInterface>(*ctx);
- tensor::ExtractSliceOp::attachInterface<
- tensor::ExtractSliceOpInterface>(*ctx);
- tensor::PadOp::attachInterface<tensor::PadOpInterface>(*ctx);
- tensor::RankOp::attachInterface<tensor::RankOpInterface>(*ctx);
- // Note: ValueBoundsOpInterface implementation is not required for ops
- // that implement `DestinationStyleOpInterface` (for querying shaped
- // OpResults).
- });
+ registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ tensor::CastOp::attachInterface<tensor::CastOpInterface>(*ctx);
+ tensor::DimOp::attachInterface<tensor::DimOpInterface>(*ctx);
+ tensor::EmptyOp::attachInterface<tensor::EmptyOpInterface>(*ctx);
+ tensor::ExtractSliceOp::attachInterface<tensor::ExtractSliceOpInterface>(
+ *ctx);
+ tensor::PadOp::attachInterface<tensor::PadOpInterface>(*ctx);
+ tensor::RankOp::attachInterface<tensor::RankOpInterface>(*ctx);
+ // Note: ValueBoundsOpInterface implementation is not required for ops that
+ // implement `DestinationStyleOpInterface` (for querying shaped OpResults).
+ });
}
diff --git a/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp b/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
index 74f795bf02ba4..33016f84056e9 100644
--- a/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
+++ b/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
@@ -70,18 +70,15 @@ struct ReassociativeReshapeOpReplacementInterface
void tensor::registerFindPayloadReplacementOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENSOR_PAYLOAD_REPLACEMENT",
- +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- CollapseShapeOp::attachInterface<
- ReassociativeReshapeOpReplacementInterface<CollapseShapeOp>>(*ctx);
- ExpandShapeOp::attachInterface<
- ReassociativeReshapeOpReplacementInterface<ExpandShapeOp>>(*ctx);
- ExtractSliceOp::attachInterface<ExtractSliceOpReplacementInterface>(
- *ctx);
- InsertSliceOp::attachInterface<InsertSliceOpReplacementInterface>(*ctx);
- ReshapeOp::attachInterface<ReshapeOpReplacementInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ CollapseShapeOp::attachInterface<
+ ReassociativeReshapeOpReplacementInterface<CollapseShapeOp>>(*ctx);
+ ExpandShapeOp::attachInterface<
+ ReassociativeReshapeOpReplacementInterface<ExpandShapeOp>>(*ctx);
+ ExtractSliceOp::attachInterface<ExtractSliceOpReplacementInterface>(*ctx);
+ InsertSliceOp::attachInterface<InsertSliceOpReplacementInterface>(*ctx);
+ ReshapeOp::attachInterface<ReshapeOpReplacementInterface>(*ctx);
+ });
}
//===----------------------------------------------------------------------===//
@@ -250,8 +247,6 @@ class TensorTransformDialectExtension
#include "mlir/Dialect/Tensor/TransformOps/TensorTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "TENSOR_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp
index b7f5fe676a7d2..87464ccb71720 100644
--- a/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -1055,30 +1055,28 @@ struct SplatOpInterface
void mlir::tensor::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENSOR_BUFFERIZATION",
- +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- CastOp::attachInterface<CastOpInterface>(*ctx);
- CollapseShapeOp::attachInterface<CollapseShapeOpInterface>(*ctx);
- DimOp::attachInterface<DimOpInterface>(*ctx);
- EmptyOp::attachInterface<EmptyOpInterface>(*ctx);
- ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
- ExtractSliceOp::attachInterface<ExtractSliceOpInterface>(*ctx);
- ExtractOp::attachInterface<ExtractOpInterface>(*ctx);
- FromElementsOp::attachInterface<FromElementsOpInterface>(*ctx);
- GenerateOp::attachInterface<GenerateOpInterface>(*ctx);
- InsertOp::attachInterface<InsertOpInterface>(*ctx);
- InsertSliceOp::attachInterface<InsertSliceOpInterface>(*ctx);
- PadOp::attachInterface<PadOpInterface>(*ctx);
- ParallelInsertSliceOp::attachInterface<ParallelInsertSliceOpInterface>(
- *ctx);
- RankOp::attachInterface<RankOpInterface>(*ctx);
- ReshapeOp::attachInterface<ReshapeOpInterface>(*ctx);
- SplatOp::attachInterface<SplatOpInterface>(*ctx);
-
- // Load additional dialects of which ops may get created.
- ctx->loadDialect<arith::ArithDialect, linalg::LinalgDialect>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ CastOp::attachInterface<CastOpInterface>(*ctx);
+ CollapseShapeOp::attachInterface<CollapseShapeOpInterface>(*ctx);
+ DimOp::attachInterface<DimOpInterface>(*ctx);
+ EmptyOp::attachInterface<EmptyOpInterface>(*ctx);
+ ExpandShapeOp::attachInterface<ExpandShapeOpInterface>(*ctx);
+ ExtractSliceOp::attachInterface<ExtractSliceOpInterface>(*ctx);
+ ExtractOp::attachInterface<ExtractOpInterface>(*ctx);
+ FromElementsOp::attachInterface<FromElementsOpInterface>(*ctx);
+ GenerateOp::attachInterface<GenerateOpInterface>(*ctx);
+ InsertOp::attachInterface<InsertOpInterface>(*ctx);
+ InsertSliceOp::attachInterface<InsertSliceOpInterface>(*ctx);
+ PadOp::attachInterface<PadOpInterface>(*ctx);
+ ParallelInsertSliceOp::attachInterface<ParallelInsertSliceOpInterface>(
+ *ctx);
+ RankOp::attachInterface<RankOpInterface>(*ctx);
+ ReshapeOp::attachInterface<ReshapeOpInterface>(*ctx);
+ SplatOp::attachInterface<SplatOpInterface>(*ctx);
+
+ // Load additional dialects of which ops may get created.
+ ctx->loadDialect<arith::ArithDialect, linalg::LinalgDialect>();
+ });
// Bufferization requires SubsetInsertionOpInterface models. Make sure that
// they are registered.
diff --git a/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp b/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp
index 2b052cbb762e1..d50d7c62b789c 100644
--- a/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tensor/Transforms/SubsetInsertionOpInterfaceImpl.cpp
@@ -86,21 +86,20 @@ struct InsertSliceLikeOpSubsetInsertionOpInterface
void mlir::tensor::registerSubsetOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TENSOR_SUBSET", +[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
- // Note: `SubsetExtractionOpInterface` and `SubsetInsertionOpInterface`
- // require `SubsetOpInterface`.
- ExtractSliceOp::attachInterface<ExtractSliceOpSubsetOpInterface>(*ctx);
- ExtractSliceOp::attachInterface<
- ExtractSliceOpSubsetExtractionOpInterface>(*ctx);
- InsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetOpInterface<InsertSliceOp>>(*ctx);
- InsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetInsertionOpInterface<InsertSliceOp>>(*ctx);
- ParallelInsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetOpInterface<ParallelInsertSliceOp>>(*ctx);
- ParallelInsertSliceOp::attachInterface<
- InsertSliceLikeOpSubsetInsertionOpInterface<ParallelInsertSliceOp>>(
- *ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, tensor::TensorDialect *dialect) {
+ // Note: `SubsetExtractionOpInterface` and `SubsetInsertionOpInterface`
+ // require `SubsetOpInterface`.
+ ExtractSliceOp::attachInterface<ExtractSliceOpSubsetOpInterface>(*ctx);
+ ExtractSliceOp::attachInterface<ExtractSliceOpSubsetExtractionOpInterface>(
+ *ctx);
+ InsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetOpInterface<InsertSliceOp>>(*ctx);
+ InsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetInsertionOpInterface<InsertSliceOp>>(*ctx);
+ ParallelInsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetOpInterface<ParallelInsertSliceOp>>(*ctx);
+ ParallelInsertSliceOp::attachInterface<
+ InsertSliceLikeOpSubsetInsertionOpInterface<ParallelInsertSliceOp>>(
+ *ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp b/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp
index 6161d5ae7e3ee..ffbb707344b8c 100644
--- a/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Tosa/IR/ShardingInterfaceImpl.cpp
@@ -76,17 +76,15 @@ static void registerElemwiseAll(MLIRContext *ctx) {
void mlir::tosa::registerShardingInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "TOSA_SHARDING", +[](MLIRContext *ctx, TosaDialect *dialect) {
- registerElemwiseAll<
- ClampOp, SigmoidOp, TanhOp, AddOp, ArithmeticRightShiftOp,
- BitwiseAndOp, BitwiseOrOp, BitwiseXorOp, IntDivOp, LogicalAndOp,
- LogicalLeftShiftOp, LogicalRightShiftOp, LogicalOrOp, LogicalXorOp,
- MaximumOp, MinimumOp, MulOp, PowOp, SubOp, AbsOp, BitwiseNotOp,
- CeilOp, ClzOp, ExpOp, FloorOp, LogOp, LogicalNotOp, NegateOp,
- ReciprocalOp, RsqrtOp, SelectOp, EqualOp, GreaterOp,
- GreaterEqualOp>(ctx);
+ registry.addExtension(+[](MLIRContext *ctx, TosaDialect *dialect) {
+ registerElemwiseAll<
+ ClampOp, SigmoidOp, TanhOp, AddOp, ArithmeticRightShiftOp, BitwiseAndOp,
+ BitwiseOrOp, BitwiseXorOp, IntDivOp, LogicalAndOp, LogicalLeftShiftOp,
+ LogicalRightShiftOp, LogicalOrOp, LogicalXorOp, MaximumOp, MinimumOp,
+ MulOp, PowOp, SubOp, AbsOp, BitwiseNotOp, CeilOp, ClzOp, ExpOp, FloorOp,
+ LogOp, LogicalNotOp, NegateOp, ReciprocalOp, RsqrtOp, SelectOp, EqualOp,
+ GreaterOp, GreaterEqualOp>(ctx);
- MatMulOp::attachInterface<MatMulOpSharding>(*ctx);
- });
+ MatMulOp::attachInterface<MatMulOpSharding>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp b/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
index 1e3c426162462..e369daddb00cb 100644
--- a/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
+++ b/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
@@ -26,8 +26,6 @@ class DebugExtension
#include "mlir/Dialect/Transform/DebugExtension/DebugExtensionOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "TRANSFORM_DEBUG";
};
} // namespace
diff --git a/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp b/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
index 8cae64ff8de0a..94004365b8a1a 100644
--- a/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
+++ b/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
@@ -26,8 +26,6 @@ class IRDLExtension
declareDependentDialect<irdl::IRDLDialect>();
}
-
- static constexpr llvm::StringRef extensionID = "IRDL_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp b/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
index 032cd65e1065f..b33288fd7b991 100644
--- a/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
+++ b/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
@@ -26,8 +26,6 @@ class LoopExtension
#include "mlir/Dialect/Transform/LoopExtension/LoopExtensionOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "TRANSFORM_LOOP";
};
} // namespace
diff --git a/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp b/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
index c7f1e63e3d921..2c770abd56d52 100644
--- a/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
+++ b/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
@@ -61,8 +61,6 @@ class PDLExtension : public transform::TransformDialectExtension<PDLExtension> {
PDLOperationTypeTransformHandleTypeInterfaceImpl>(*context);
});
}
-
- static constexpr llvm::StringRef extensionID = "TRANSFORM_PDL";
};
} // namespace
diff --git a/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp b/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp
index 91bf4ef633a6d..ca95072d9bb0f 100644
--- a/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Vector/IR/ValueBoundsOpInterfaceImpl.cpp
@@ -44,10 +44,8 @@ struct VectorScaleOpInterface
void mlir::vector::registerValueBoundsOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "VECTOR_VALUE_BOUNDS",
- +[](MLIRContext *ctx, vector::VectorDialect *dialect) {
- vector::VectorScaleOp::attachInterface<vector::VectorScaleOpInterface>(
- *ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, vector::VectorDialect *dialect) {
+ vector::VectorScaleOp::attachInterface<vector::VectorScaleOpInterface>(
+ *ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp b/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
index 0b42040a4d63b..2e9aa88011825 100644
--- a/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
+++ b/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
@@ -220,8 +220,6 @@ class VectorTransformDialectExtension
#include "mlir/Dialect/Vector/TransformOps/VectorTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "VECTOR_TRANSFORM";
};
} // namespace
diff --git a/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp b/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp
index d3e01e33b3f1d..1caec5bb8644f 100644
--- a/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Vector/Transforms/BufferizableOpInterfaceImpl.cpp
@@ -317,13 +317,11 @@ struct YieldOpInterface
void mlir::vector::registerBufferizableOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "VECTOR_BUFFERIZATION",
- +[](MLIRContext *ctx, vector::VectorDialect *dialect) {
- TransferReadOp::attachInterface<TransferReadOpInterface>(*ctx);
- TransferWriteOp::attachInterface<TransferWriteOpInterface>(*ctx);
- GatherOp::attachInterface<GatherOpInterface>(*ctx);
- MaskOp::attachInterface<MaskOpInterface>(*ctx);
- YieldOp::attachInterface<YieldOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, vector::VectorDialect *dialect) {
+ TransferReadOp::attachInterface<TransferReadOpInterface>(*ctx);
+ TransferWriteOp::attachInterface<TransferWriteOpInterface>(*ctx);
+ GatherOp::attachInterface<GatherOpInterface>(*ctx);
+ MaskOp::attachInterface<MaskOpInterface>(*ctx);
+ YieldOp::attachInterface<YieldOpInterface>(*ctx);
+ });
}
diff --git a/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp b/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp
index 91a014a7ab24b..b450d5b78a466 100644
--- a/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp
+++ b/mlir/lib/Dialect/Vector/Transforms/SubsetOpInterfaceImpl.cpp
@@ -69,15 +69,14 @@ struct TransferWriteOpSubsetInsertionOpInterface
void mlir::vector::registerSubsetOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "VECTOR_SUBSET", +[](MLIRContext *ctx, vector::VectorDialect *dialect) {
- TransferReadOp::attachInterface<
- XferOpSubsetOpInterface<TransferReadOp>>(*ctx);
- TransferReadOp::attachInterface<
- TransferReadOpSubsetExtractionOpInterface>(*ctx);
- TransferWriteOp::attachInterface<
- XferOpSubsetOpInterface<TransferWriteOp>>(*ctx);
- TransferWriteOp::attachInterface<
- TransferWriteOpSubsetInsertionOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, vector::VectorDialect *dialect) {
+ TransferReadOp::attachInterface<XferOpSubsetOpInterface<TransferReadOp>>(
+ *ctx);
+ TransferReadOp::attachInterface<TransferReadOpSubsetExtractionOpInterface>(
+ *ctx);
+ TransferWriteOp::attachInterface<XferOpSubsetOpInterface<TransferWriteOp>>(
+ *ctx);
+ TransferWriteOp::attachInterface<TransferWriteOpSubsetInsertionOpInterface>(
+ *ctx);
+ });
}
diff --git a/mlir/lib/Interfaces/CastInterfaces.cpp b/mlir/lib/Interfaces/CastInterfaces.cpp
index 11e581a5fefa8..05c872daf5dab 100644
--- a/mlir/lib/Interfaces/CastInterfaces.cpp
+++ b/mlir/lib/Interfaces/CastInterfaces.cpp
@@ -79,11 +79,10 @@ struct UnrealizedConversionCastOpInterface
void mlir::builtin::registerCastOpInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "BUILTIN_CAST", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- UnrealizedConversionCastOp::attachInterface<
- UnrealizedConversionCastOpInterface>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ UnrealizedConversionCastOp::attachInterface<
+ UnrealizedConversionCastOpInterface>(*ctx);
+ });
}
//===----------------------------------------------------------------------===//
diff --git a/mlir/lib/Target/LLVM/NVVM/Target.cpp b/mlir/lib/Target/LLVM/NVVM/Target.cpp
index f5d1babd01180..e608d26e8d2ec 100644
--- a/mlir/lib/Target/LLVM/NVVM/Target.cpp
+++ b/mlir/lib/Target/LLVM/NVVM/Target.cpp
@@ -58,10 +58,9 @@ class NVVMTargetAttrImpl
// Register the NVVM dialect, the NVVM translation & the target interface.
void mlir::NVVM::registerNVVMTargetInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "NVVM_TARGET", +[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
- NVVMTargetAttr::attachInterface<NVVMTargetAttrImpl>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
+ NVVMTargetAttr::attachInterface<NVVMTargetAttrImpl>(*ctx);
+ });
}
void mlir::NVVM::registerNVVMTargetInterfaceExternalModels(
diff --git a/mlir/lib/Target/LLVM/ROCDL/Target.cpp b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
index 3a56688117fda..4d23f987eb05e 100644
--- a/mlir/lib/Target/LLVM/ROCDL/Target.cpp
+++ b/mlir/lib/Target/LLVM/ROCDL/Target.cpp
@@ -68,10 +68,9 @@ class ROCDLTargetAttrImpl
// Register the ROCDL dialect, the ROCDL translation and the target interface.
void mlir::ROCDL::registerROCDLTargetInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "ROCDL_TARGET", +[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
- ROCDLTargetAttr::attachInterface<ROCDLTargetAttrImpl>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
+ ROCDLTargetAttr::attachInterface<ROCDLTargetAttrImpl>(*ctx);
+ });
}
void mlir::ROCDL::registerROCDLTargetInterfaceExternalModels(
diff --git a/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp
index 4ebcfa5941d0e..044462d33cfd1 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/AMX/AMXToLLVMIRTranslation.cpp
@@ -44,10 +44,9 @@ class AMXDialectLLVMIRTranslationInterface
void mlir::registerAMXDialectTranslation(DialectRegistry ®istry) {
registry.insert<amx::AMXDialect>();
- registry.addExtension(
- "AMX_TO_LLVMIR", +[](MLIRContext *ctx, amx::AMXDialect *dialect) {
- dialect->addInterfaces<AMXDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, amx::AMXDialect *dialect) {
+ dialect->addInterfaces<AMXDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerAMXDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp
index d317f68731eb8..7098592d506e0 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ArmNeon/ArmNeonToLLVMIRTranslation.cpp
@@ -46,7 +46,6 @@ class ArmNeonDialectLLVMIRTranslationInterface
void mlir::registerArmNeonDialectTranslation(DialectRegistry ®istry) {
registry.insert<arm_neon::ArmNeonDialect>();
registry.addExtension(
- "ARM_NEON_TO_LLVMIR",
+[](MLIRContext *ctx, arm_neon::ArmNeonDialect *dialect) {
dialect->addInterfaces<ArmNeonDialectLLVMIRTranslationInterface>();
});
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp
index a1f63b9f07bdd..e6ee41188d594 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ArmSME/ArmSMEToLLVMIRTranslation.cpp
@@ -45,11 +45,9 @@ class ArmSMEDialectLLVMIRTranslationInterface
void mlir::registerArmSMEDialectTranslation(DialectRegistry ®istry) {
registry.insert<arm_sme::ArmSMEDialect>();
- registry.addExtension(
- "ARM_SME_TO_LLVMIR",
- +[](MLIRContext *ctx, arm_sme::ArmSMEDialect *dialect) {
- dialect->addInterfaces<ArmSMEDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, arm_sme::ArmSMEDialect *dialect) {
+ dialect->addInterfaces<ArmSMEDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerArmSMEDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp
index 3411f5d147a34..cd10811b68f02 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ArmSVE/ArmSVEToLLVMIRTranslation.cpp
@@ -44,11 +44,9 @@ class ArmSVEDialectLLVMIRTranslationInterface
void mlir::registerArmSVEDialectTranslation(DialectRegistry ®istry) {
registry.insert<arm_sve::ArmSVEDialect>();
- registry.addExtension(
- "ARM_SVE_TO_LLVMIR",
- +[](MLIRContext *ctx, arm_sve::ArmSVEDialect *dialect) {
- dialect->addInterfaces<ArmSVEDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, arm_sve::ArmSVEDialect *dialect) {
+ dialect->addInterfaces<ArmSVEDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerArmSVEDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp
index c120eb243b783..51c304cfbb8e5 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/Builtin/BuiltinToLLVMIRTranslation.cpp
@@ -34,10 +34,9 @@ class BuiltinDialectLLVMIRTranslationInterface
} // namespace
void mlir::registerBuiltinDialectTranslation(DialectRegistry ®istry) {
- registry.addExtension(
- "BUILTIN_TO_LLVMIR", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- dialect->addInterfaces<BuiltinDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ dialect->addInterfaces<BuiltinDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerBuiltinDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp
index c4fff8cfee7d1..eecc8f1001ca4 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/GPU/GPUToLLVMIRTranslation.cpp
@@ -65,10 +65,9 @@ class GPUDialectLLVMIRTranslationInterface
void mlir::registerGPUDialectTranslation(DialectRegistry ®istry) {
registry.insert<gpu::GPUDialect>();
- registry.addExtension(
- "GPU_TO_LLVMIR", +[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
- dialect->addInterfaces<GPUDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
+ dialect->addInterfaces<GPUDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerGPUDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp b/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp
index c982e5f67900f..b023c4c126da3 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/GPU/SelectObjectAttr.cpp
@@ -56,10 +56,9 @@ std::string getBinaryIdentifier(StringRef binaryName) {
void mlir::gpu::registerOffloadingLLVMTranslationInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "GPU_OFFLOADING_LLVM", +[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
- SelectObjectAttr::attachInterface<SelectObjectAttrImpl>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, gpu::GPUDialect *dialect) {
+ SelectObjectAttr::attachInterface<SelectObjectAttrImpl>(*ctx);
+ });
}
gpu::ObjectAttr
diff --git a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp
index 14485c54ffdde..06673965245c0 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMIRToLLVMTranslation.cpp
@@ -281,10 +281,9 @@ class LLVMDialectLLVMIRImportInterface : public LLVMImportDialectInterface {
void mlir::registerLLVMDialectImport(DialectRegistry ®istry) {
registry.insert<LLVM::LLVMDialect>();
- registry.addExtension(
- "LLVMIR_TO_LLVM", +[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
- dialect->addInterfaces<LLVMDialectLLVMIRImportInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
+ dialect->addInterfaces<LLVMDialectLLVMIRImportInterface>();
+ });
}
void mlir::registerLLVMDialectImport(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp
index d5bc600fd7e13..bdb15a290209b 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/LLVMIR/LLVMToLLVMIRTranslation.cpp
@@ -439,10 +439,9 @@ class LLVMDialectLLVMIRTranslationInterface
void mlir::registerLLVMDialectTranslation(DialectRegistry ®istry) {
registry.insert<LLVM::LLVMDialect>();
- registry.addExtension(
- "LLVM_TO_LLVMIR", +[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
- dialect->addInterfaces<LLVMDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, LLVM::LLVMDialect *dialect) {
+ dialect->addInterfaces<LLVMDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerLLVMDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp
index 9978513e59ec0..855abc12a909e 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/NVVM/LLVMIRToNVVMTranslation.cpp
@@ -81,10 +81,9 @@ class NVVMDialectLLVMIRImportInterface : public LLVMImportDialectInterface {
void mlir::registerNVVMDialectImport(DialectRegistry ®istry) {
registry.insert<NVVM::NVVMDialect>();
- registry.addExtension(
- "LLVMIR_TO_NVVM", +[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
- dialect->addInterfaces<NVVMDialectLLVMIRImportInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
+ dialect->addInterfaces<NVVMDialectLLVMIRImportInterface>();
+ });
}
void mlir::registerNVVMDialectImport(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp
index e12364b1122f0..a09c24dda82af 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/NVVM/NVVMToLLVMIRTranslation.cpp
@@ -261,10 +261,9 @@ class NVVMDialectLLVMIRTranslationInterface
void mlir::registerNVVMDialectTranslation(DialectRegistry ®istry) {
registry.insert<NVVM::NVVMDialect>();
- registry.addExtension(
- "NVVM_TO_LLVMIR", +[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
- dialect->addInterfaces<NVVMDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, NVVM::NVVMDialect *dialect) {
+ dialect->addInterfaces<NVVMDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerNVVMDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp
index 67f0f8fc4f374..d9cf85e4aecab 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/OpenACC/OpenACCToLLVMIRTranslation.cpp
@@ -532,10 +532,9 @@ LogicalResult OpenACCDialectLLVMIRTranslationInterface::convertOperation(
void mlir::registerOpenACCDialectTranslation(DialectRegistry ®istry) {
registry.insert<acc::OpenACCDialect>();
- registry.addExtension(
- "OPENACC_TO_LLVMIR", +[](MLIRContext *ctx, acc::OpenACCDialect *dialect) {
- dialect->addInterfaces<OpenACCDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, acc::OpenACCDialect *dialect) {
+ dialect->addInterfaces<OpenACCDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerOpenACCDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp
index a43761460fa31..ddee117838697 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/OpenMP/OpenMPToLLVMIRTranslation.cpp
@@ -3741,10 +3741,9 @@ LogicalResult OpenMPDialectLLVMIRTranslationInterface::convertOperation(
void mlir::registerOpenMPDialectTranslation(DialectRegistry ®istry) {
registry.insert<omp::OpenMPDialect>();
- registry.addExtension(
- "OPENMV_TO_LLVMIR", +[](MLIRContext *ctx, omp::OpenMPDialect *dialect) {
- dialect->addInterfaces<OpenMPDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, omp::OpenMPDialect *dialect) {
+ dialect->addInterfaces<OpenMPDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerOpenMPDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp
index 7fe58e5d161a8..2a146f5efed30 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/ROCDL/ROCDLToLLVMIRTranslation.cpp
@@ -190,10 +190,9 @@ class ROCDLDialectLLVMIRTranslationInterface
void mlir::registerROCDLDialectTranslation(DialectRegistry ®istry) {
registry.insert<ROCDL::ROCDLDialect>();
- registry.addExtension(
- "ROCDL_TO_LLVMIR", +[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
- dialect->addInterfaces<ROCDLDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, ROCDL::ROCDLDialect *dialect) {
+ dialect->addInterfaces<ROCDLDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerROCDLDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp
index 17d63863d9ddc..b78b002d32292 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/VCIX/VCIXToLLVMIRTranslation.cpp
@@ -76,10 +76,9 @@ class VCIXDialectLLVMIRTranslationInterface
void mlir::registerVCIXDialectTranslation(DialectRegistry ®istry) {
registry.insert<vcix::VCIXDialect>();
- registry.addExtension(
- "VCIX_TO_LLVMIR", +[](MLIRContext *ctx, vcix::VCIXDialect *dialect) {
- dialect->addInterfaces<VCIXDialectLLVMIRTranslationInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, vcix::VCIXDialect *dialect) {
+ dialect->addInterfaces<VCIXDialectLLVMIRTranslationInterface>();
+ });
}
void mlir::registerVCIXDialectTranslation(MLIRContext &context) {
diff --git a/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp b/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp
index 5a6104162123b..fa5f61420ee8a 100644
--- a/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp
+++ b/mlir/lib/Target/LLVMIR/Dialect/X86Vector/X86VectorToLLVMIRTranslation.cpp
@@ -46,7 +46,6 @@ class X86VectorDialectLLVMIRTranslationInterface
void mlir::registerX86VectorDialectTranslation(DialectRegistry ®istry) {
registry.insert<x86vector::X86VectorDialect>();
registry.addExtension(
- "X86_TO_LLVMIR",
+[](MLIRContext *ctx, x86vector::X86VectorDialect *dialect) {
dialect->addInterfaces<X86VectorDialectLLVMIRTranslationInterface>();
});
diff --git a/mlir/lib/Target/SPIRV/Target.cpp b/mlir/lib/Target/SPIRV/Target.cpp
index 15832a43b6620..4c416abe71cac 100644
--- a/mlir/lib/Target/SPIRV/Target.cpp
+++ b/mlir/lib/Target/SPIRV/Target.cpp
@@ -43,10 +43,9 @@ class SPIRVTargetAttrImpl
// Register the SPIR-V dialect, the SPIR-V translation & the target interface.
void mlir::spirv::registerSPIRVTargetInterfaceExternalModels(
DialectRegistry ®istry) {
- registry.addExtension(
- "SPIRV_TARGET", +[](MLIRContext *ctx, spirv::SPIRVDialect *dialect) {
- spirv::TargetEnvAttr::attachInterface<SPIRVTargetAttrImpl>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, spirv::SPIRVDialect *dialect) {
+ spirv::TargetEnvAttr::attachInterface<SPIRVTargetAttrImpl>(*ctx);
+ });
}
void mlir::spirv::registerSPIRVTargetInterfaceExternalModels(
diff --git a/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp b/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp
index 0422705663b81..dc6413b25707e 100644
--- a/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp
+++ b/mlir/test/lib/Dialect/Test/TestFromLLVMIRTranslation.cpp
@@ -104,7 +104,6 @@ void registerTestFromLLVMIR() {
registry.insert<test::TestDialect>();
registerLLVMDialectImport(registry);
registry.addExtension(
- "TEST_FROM_LLVMIR",
+[](MLIRContext *ctx, test::TestDialect *dialect) {
dialect->addInterfaces<TestDialectLLVMImportDialectInterface>();
});
diff --git a/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp b/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp
index 3d4b8dc6b1495..157c6265be834 100644
--- a/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp
+++ b/mlir/test/lib/Dialect/Test/TestToLLVMIRTranslation.cpp
@@ -140,7 +140,6 @@ void registerTestToLLVMIR() {
registerBuiltinDialectTranslation(registry);
registerLLVMDialectTranslation(registry);
registry.addExtension(
- "TEST_TO_LLVMIR",
+[](MLIRContext *ctx, test::TestDialect *dialect) {
dialect->addInterfaces<TestDialectLLVMIRTranslationInterface>();
});
diff --git a/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp b/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
index cc8f0e8988ea6..b8a4b9470d736 100644
--- a/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
+++ b/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
@@ -905,8 +905,6 @@ class TestTransformDialectExtension
hooks.mergeInPDLMatchHooks(std::move(constraints));
});
}
-
- static constexpr llvm::StringRef extensionID = "TEST_TRANSFORM";
};
} // namespace
diff --git a/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp b/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
index 172001123d4da..8f206d9077272 100644
--- a/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
+++ b/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
@@ -399,8 +399,6 @@ class TestTilingInterfaceDialectExtension
#include "TestTilingInterfaceTransformOps.cpp.inc"
>();
}
-
- static constexpr llvm::StringRef extensionID = "TEST_TILING_TRANSFORM";
};
} // namespace
diff --git a/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp b/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
index 2d8fbf078618f..40fb752ffd6eb 100644
--- a/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
+++ b/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
@@ -20,8 +20,6 @@ class Extension : public TransformDialectExtension<Extension> {
public:
using Base::Base;
void init() { declareGeneratedDialect<func::FuncDialect>(); }
-
- static constexpr llvm::StringRef extensionID = "TRANSFORM_BUILD_ONLY";
};
} // end namespace
diff --git a/mlir/unittests/IR/DialectTest.cpp b/mlir/unittests/IR/DialectTest.cpp
index 2623978639674..4be392faa31d2 100644
--- a/mlir/unittests/IR/DialectTest.cpp
+++ b/mlir/unittests/IR/DialectTest.cpp
@@ -8,6 +8,7 @@
#include "mlir/IR/Dialect.h"
#include "mlir/IR/DialectInterface.h"
+#include "mlir/Support/TypeID.h"
#include "gtest/gtest.h"
using namespace mlir;
@@ -75,10 +76,9 @@ TEST(Dialect, DelayedInterfaceRegistration) {
registry.insert<TestDialect, SecondTestDialect>();
// Delayed registration of an interface for TestDialect.
- registry.addExtension(
- "TEST_DIALECT_DELAYED", +[](MLIRContext *ctx, TestDialect *dialect) {
- dialect->addInterfaces<TestDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, TestDialect *dialect) {
+ dialect->addInterfaces<TestDialectInterface>();
+ });
MLIRContext context(registry);
@@ -101,7 +101,7 @@ TEST(Dialect, DelayedInterfaceRegistration) {
DialectRegistry secondRegistry;
secondRegistry.insert<SecondTestDialect>();
secondRegistry.addExtension(
- "SECOND_TEST", +[](MLIRContext *ctx, SecondTestDialect *dialect) {
+ +[](MLIRContext *ctx, SecondTestDialect *dialect) {
dialect->addInterfaces<SecondTestDialectInterface>();
});
context.appendDialectRegistry(secondRegistry);
@@ -114,10 +114,9 @@ TEST(Dialect, RepeatedDelayedRegistration) {
// Set up the delayed registration.
DialectRegistry registry;
registry.insert<TestDialect>();
- registry.addExtension(
- "TEST_DIALECT", +[](MLIRContext *ctx, TestDialect *dialect) {
- dialect->addInterfaces<TestDialectInterface>();
- });
+ registry.addExtension(+[](MLIRContext *ctx, TestDialect *dialect) {
+ dialect->addInterfaces<TestDialectInterface>();
+ });
MLIRContext context(registry);
// Load the TestDialect and check that the interface got registered for it.
@@ -130,10 +129,9 @@ TEST(Dialect, RepeatedDelayedRegistration) {
// on repeated interface registration.
DialectRegistry secondRegistry;
secondRegistry.insert<TestDialect>();
- secondRegistry.addExtension(
- "TEST_DIALECT", +[](MLIRContext *ctx, TestDialect *dialect) {
- dialect->addInterfaces<TestDialectInterface>();
- });
+ secondRegistry.addExtension(+[](MLIRContext *ctx, TestDialect *dialect) {
+ dialect->addInterfaces<TestDialectInterface>();
+ });
context.appendDialectRegistry(secondRegistry);
testDialectInterface = dyn_cast<TestDialectInterfaceBase>(testDialect);
EXPECT_TRUE(testDialectInterface != nullptr);
@@ -143,20 +141,21 @@ namespace {
/// A dummy extension that increases a counter when being applied and
/// recursively adds additional extensions.
struct DummyExtension : DialectExtension<DummyExtension, TestDialect> {
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(DummyExtension)
+
DummyExtension(int *counter, int numRecursive)
: DialectExtension(), counter(counter), numRecursive(numRecursive) {}
- inline static std::vector<std::string> extensionIDs;
-
void apply(MLIRContext *ctx, TestDialect *dialect) const final {
++(*counter);
DialectRegistry nestedRegistry;
- extensionIDs.reserve(extensionIDs.size() + numRecursive);
for (int i = 0; i < numRecursive; ++i) {
- extensionIDs.push_back("DUMMY_" + std::to_string(i));
- nestedRegistry.addExtension(
- extensionIDs.back(),
- std::make_unique<DummyExtension>(counter, /*numRecursive=*/0));
+ // Create unique TypeIDs for these recursive extensions so they don't get
+ // de-duplicated.
+ auto extension =
+ std::make_unique<DummyExtension>(counter, /*numRecursive=*/0);
+ auto typeID = TypeID::getFromOpaquePointer(extension.get());
+ nestedRegistry.addExtension(typeID, std::move(extension));
}
// Adding additional extensions may trigger a reallocation of the
// `extensions` vector in the dialect registry.
@@ -175,11 +174,11 @@ TEST(Dialect, NestedDialectExtension) {
// Add an extension that adds 100 more extensions.
int counter1 = 0;
- registry.addExtension("DUMMY",
+ registry.addExtension(TypeID::get<DummyExtension>(),
std::make_unique<DummyExtension>(&counter1, 100));
// Add one more extension. This should not crash.
int counter2 = 0;
- registry.addExtension("DUMMY2",
+ registry.addExtension(TypeID::get<DummyExtension>(),
std::make_unique<DummyExtension>(&counter2, 0));
// Load dialect and apply extensions.
@@ -187,7 +186,7 @@ TEST(Dialect, NestedDialectExtension) {
Dialect *testDialect = context.getOrLoadDialect<TestDialect>();
ASSERT_TRUE(testDialect != nullptr);
- // Extensions may be applied multiple times. Make sure that each expected
+ // Extensions are de-duplicated by typeID. Make sure that each expected
// extension was applied at least once.
EXPECT_GE(counter1, 101);
EXPECT_GE(counter2, 1);
@@ -203,19 +202,28 @@ TEST(Dialect, SubsetWithExtensions) {
ASSERT_TRUE(registry2.isSubsetOf(registry1));
// Add extensions to registry2.
- int counter;
- registry2.addExtension("EXT", std::make_unique<DummyExtension>(&counter, 0));
+ int counter = 0;
+ registry2.addExtension(TypeID::get<DummyExtension>(),
+ std::make_unique<DummyExtension>(&counter, 0));
// Expect that (1) is a subset of (2) but not the other way around.
ASSERT_TRUE(registry1.isSubsetOf(registry2));
ASSERT_FALSE(registry2.isSubsetOf(registry1));
// Add extensions to registry1.
- registry1.addExtension("EXT", std::make_unique<DummyExtension>(&counter, 0));
+ registry1.addExtension(TypeID::get<DummyExtension>(),
+ std::make_unique<DummyExtension>(&counter, 0));
// Expect that (1) and (2) are equivalent.
ASSERT_TRUE(registry1.isSubsetOf(registry2));
ASSERT_TRUE(registry2.isSubsetOf(registry1));
+
+ // Load dialect and apply extensions.
+ MLIRContext context(registry1);
+ context.getOrLoadDialect<TestDialect>();
+ context.appendDialectRegistry(registry2);
+ // Expect that the extension as only invoked once.
+ ASSERT_EQ(counter, 1);
}
} // namespace
diff --git a/mlir/unittests/IR/InterfaceAttachmentTest.cpp b/mlir/unittests/IR/InterfaceAttachmentTest.cpp
index a5b835153b636..b6066dd5685dc 100644
--- a/mlir/unittests/IR/InterfaceAttachmentTest.cpp
+++ b/mlir/unittests/IR/InterfaceAttachmentTest.cpp
@@ -103,11 +103,9 @@ TEST(InterfaceAttachment, TypeDelayedContextConstruct) {
// Put the interface in the registry.
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(
- "TYPE_DELAYED_CONSTRUCT",
- +[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::TestType::attachInterface<TestTypeModel>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::TestType::attachInterface<TestTypeModel>(*ctx);
+ });
// Check that when a context is constructed with the given registry, the type
// interface gets registered.
@@ -124,10 +122,9 @@ TEST(InterfaceAttachment, TypeDelayedContextAppend) {
// Put the interface in the registry.
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(
- "TYPE_DELAYED_APPEND", +[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::TestType::attachInterface<TestTypeModel>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::TestType::attachInterface<TestTypeModel>(*ctx);
+ });
// Check that when the registry gets appended to the context, the interface
// becomes available for objects in loaded dialects.
@@ -141,10 +138,9 @@ TEST(InterfaceAttachment, TypeDelayedContextAppend) {
TEST(InterfaceAttachment, RepeatedRegistration) {
DialectRegistry registry;
- registry.addExtension(
- "REPEATED", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- IntegerType::attachInterface<Model>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ IntegerType::attachInterface<Model>(*ctx);
+ });
MLIRContext context(registry);
// Should't fail on repeated registration through the dialect registry.
@@ -155,10 +151,9 @@ TEST(InterfaceAttachment, TypeBuiltinDelayed) {
// Builtin dialect needs to registration or loading, but delayed interface
// registration must still work.
DialectRegistry registry;
- registry.addExtension(
- "BUILTIN_DELAYED", +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- IntegerType::attachInterface<Model>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ IntegerType::attachInterface<Model>(*ctx);
+ });
MLIRContext context(registry);
IntegerType i16 = IntegerType::get(&context, 16);
@@ -251,10 +246,9 @@ TEST(InterfaceAttachmentTest, AttributeDelayed) {
// that the delayed registration work for attributes.
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(
- "ATTRIBUTE_DELAYED", +[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::SimpleAAttr::attachInterface<TestExternalSimpleAAttrModel>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::SimpleAAttr::attachInterface<TestExternalSimpleAAttrModel>(*ctx);
+ });
MLIRContext context(registry);
context.loadDialect<test::TestDialect>();
@@ -358,17 +352,13 @@ struct TestExternalTestOpModel
TEST(InterfaceAttachment, OperationDelayedContextConstruct) {
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(
- "OPERATION_DELAYED_BUILTIN",
- +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
- });
- registry.addExtension(
- "OPERATION_DELAYED_TEST",
- +[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
- test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
+ });
+ registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
+ test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
+ });
// Construct the context directly from a registry. The interfaces are
// expected to be readily available on operations.
@@ -393,17 +383,13 @@ TEST(InterfaceAttachment, OperationDelayedContextConstruct) {
TEST(InterfaceAttachment, OperationDelayedContextAppend) {
DialectRegistry registry;
registry.insert<test::TestDialect>();
- registry.addExtension(
- "OPERATION_DELAYED_BUILTIN",
- +[](MLIRContext *ctx, BuiltinDialect *dialect) {
- ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
- });
- registry.addExtension(
- "OPERATION_DELAYED_TEST",
- +[](MLIRContext *ctx, test::TestDialect *dialect) {
- test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
- test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
- });
+ registry.addExtension(+[](MLIRContext *ctx, BuiltinDialect *dialect) {
+ ModuleOp::attachInterface<TestExternalOpModel>(*ctx);
+ });
+ registry.addExtension(+[](MLIRContext *ctx, test::TestDialect *dialect) {
+ test::OpJ::attachInterface<TestExternalTestOpModel<test::OpJ>>(*ctx);
+ test::OpH::attachInterface<TestExternalTestOpModel<test::OpH>>(*ctx);
+ });
// Construct the context, create ops, and only then append the registry. The
// interfaces are expected to be available after appending the registry.
>From f5e17822005c133de10699b4b4db1c7c3ec5ad69 Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Thu, 1 Aug 2024 08:39:51 -0700
Subject: [PATCH 5/8] Use typeID for extension ID
---
mlir/include/mlir/IR/DialectRegistry.h | 31 +++++++++-----------------
mlir/lib/IR/Dialect.cpp | 3 ++-
2 files changed, 13 insertions(+), 21 deletions(-)
diff --git a/mlir/include/mlir/IR/DialectRegistry.h b/mlir/include/mlir/IR/DialectRegistry.h
index ce5a766843228..69a81231975d7 100644
--- a/mlir/include/mlir/IR/DialectRegistry.h
+++ b/mlir/include/mlir/IR/DialectRegistry.h
@@ -14,9 +14,9 @@
#define MLIR_IR_DIALECTREGISTRY_H
#include "mlir/IR/MLIRContext.h"
+#include "mlir/Support/TypeID.h"
#include "llvm/ADT/ArrayRef.h"
#include "llvm/ADT/SmallVector.h"
-#include "llvm/ADT/StringRef.h"
#include <map>
#include <tuple>
@@ -207,7 +207,7 @@ class DialectRegistry {
void applyExtensions(MLIRContext *ctx) const;
/// Add the given extension to the registry.
- void addExtension(StringRef extensionID,
+ void addExtension(TypeID extensionID,
std::unique_ptr<DialectExtensionBase> extension) {
extensions.emplace_back(extensionID, std::move(extension));
}
@@ -215,39 +215,29 @@ class DialectRegistry {
/// Add the given extensions to the registry.
template <typename... ExtensionsT>
void addExtensions() {
- (addExtension(ExtensionsT::extensionID, std::make_unique<ExtensionsT>()),
+ (addExtension(TypeID::get<ExtensionsT>(), std::make_unique<ExtensionsT>()),
...);
}
/// Add an extension function that requires the given dialects.
/// Note: This bare functor overload is provided in addition to the
/// std::function variant to enable dialect type deduction, e.g.:
- /// registry.addExtension("ID", +[](MLIRContext *ctx, MyDialect *dialect) {
+ /// registry.addExtension(+[](MLIRContext *ctx, MyDialect *dialect) {
/// ... })
///
/// is equivalent to:
/// registry.addExtension<MyDialect>(
- /// "ID",
/// [](MLIRContext *ctx, MyDialect *dialect){ ... }
/// )
template <typename... DialectsT>
- void addExtension(StringRef extensionID,
- void (*extensionFn)(MLIRContext *, DialectsT *...)) {
- addExtension<DialectsT...>(
- extensionID,
- std::function<void(MLIRContext *, DialectsT * ...)>(extensionFn));
- }
- template <typename... DialectsT>
- void
- addExtension(StringRef extensionID,
- std::function<void(MLIRContext *, DialectsT *...)> extensionFn) {
- using ExtensionFnT = std::function<void(MLIRContext *, DialectsT * ...)>;
+ void addExtension(void (*extensionFn)(MLIRContext *, DialectsT *...)) {
+ using ExtensionFnT = void (*)(MLIRContext *, DialectsT *...);
struct Extension : public DialectExtension<Extension, DialectsT...> {
Extension(const Extension &) = default;
Extension(ExtensionFnT extensionFn)
: DialectExtension<Extension, DialectsT...>(),
- extensionFn(std::move(extensionFn)) {}
+ extensionFn(extensionFn) {}
~Extension() override = default;
void apply(MLIRContext *context, DialectsT *...dialects) const final {
@@ -255,8 +245,9 @@ class DialectRegistry {
}
ExtensionFnT extensionFn;
};
- addExtension(extensionID,
- std::make_unique<Extension>(std::move(extensionFn)));
+ addExtension(TypeID::getFromOpaquePointer(
+ reinterpret_cast<const void *>(extensionFn)),
+ std::make_unique<Extension>(extensionFn));
}
/// Returns true if the current registry is a subset of 'rhs', i.e. if 'rhs'
@@ -266,7 +257,7 @@ class DialectRegistry {
private:
MapTy registry;
using KeyExtensionPair =
- std::pair<llvm::StringRef, std::unique_ptr<DialectExtensionBase>>;
+ std::pair<TypeID, std::unique_ptr<DialectExtensionBase>>;
llvm::SmallVector<KeyExtensionPair> extensions;
};
diff --git a/mlir/lib/IR/Dialect.cpp b/mlir/lib/IR/Dialect.cpp
index c8b55bf751091..30fdb14703e75 100644
--- a/mlir/lib/IR/Dialect.cpp
+++ b/mlir/lib/IR/Dialect.cpp
@@ -14,6 +14,7 @@
#include "mlir/IR/ExtensibleDialect.h"
#include "mlir/IR/MLIRContext.h"
#include "mlir/IR/Operation.h"
+#include "mlir/Support/TypeID.h"
#include "llvm/ADT/MapVector.h"
#include "llvm/ADT/Twine.h"
#include "llvm/Support/Debug.h"
@@ -292,7 +293,7 @@ void DialectRegistry::applyExtensions(MLIRContext *ctx) const {
bool DialectRegistry::isSubsetOf(const DialectRegistry &rhs) const {
// Check that all extension keys are present in 'rhs'.
- llvm::DenseSet<llvm::StringRef> rhsExtensionKeys;
+ llvm::DenseSet<TypeID> rhsExtensionKeys;
{
auto rhsKeys = llvm::map_range(rhs.extensions,
[](const auto &item) { return item.first; });
>From 06356494de46638aa9e7ac59ea3056a71718a107 Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Thu, 1 Aug 2024 11:54:59 -0700
Subject: [PATCH 6/8] cleaned up impl
---
mlir/include/mlir/IR/DialectRegistry.h | 22 ++++-----
mlir/lib/IR/Dialect.cpp | 63 +++++++++++++++++++-------
mlir/unittests/IR/DialectTest.cpp | 2 +-
3 files changed, 58 insertions(+), 29 deletions(-)
diff --git a/mlir/include/mlir/IR/DialectRegistry.h b/mlir/include/mlir/IR/DialectRegistry.h
index 69a81231975d7..d507e56e42ca8 100644
--- a/mlir/include/mlir/IR/DialectRegistry.h
+++ b/mlir/include/mlir/IR/DialectRegistry.h
@@ -16,7 +16,7 @@
#include "mlir/IR/MLIRContext.h"
#include "mlir/Support/TypeID.h"
#include "llvm/ADT/ArrayRef.h"
-#include "llvm/ADT/SmallVector.h"
+#include "llvm/ADT/MapVector.h"
#include <map>
#include <tuple>
@@ -187,8 +187,8 @@ class DialectRegistry {
nameAndRegistrationIt.second.second);
// Merge the extensions.
for (const auto &extension : extensions)
- destination.extensions.emplace_back(extension.first,
- extension.second->clone());
+ destination.extensions.try_emplace(extension.first,
+ extension.second->clone());
}
/// Return the names of dialects known to this registry.
@@ -207,9 +207,9 @@ class DialectRegistry {
void applyExtensions(MLIRContext *ctx) const;
/// Add the given extension to the registry.
- void addExtension(TypeID extensionID,
+ bool addExtension(TypeID extensionID,
std::unique_ptr<DialectExtensionBase> extension) {
- extensions.emplace_back(extensionID, std::move(extension));
+ return extensions.try_emplace(extensionID, std::move(extension)).second;
}
/// Add the given extensions to the registry.
@@ -230,7 +230,7 @@ class DialectRegistry {
/// [](MLIRContext *ctx, MyDialect *dialect){ ... }
/// )
template <typename... DialectsT>
- void addExtension(void (*extensionFn)(MLIRContext *, DialectsT *...)) {
+ bool addExtension(void (*extensionFn)(MLIRContext *, DialectsT *...)) {
using ExtensionFnT = void (*)(MLIRContext *, DialectsT *...);
struct Extension : public DialectExtension<Extension, DialectsT...> {
@@ -245,9 +245,9 @@ class DialectRegistry {
}
ExtensionFnT extensionFn;
};
- addExtension(TypeID::getFromOpaquePointer(
- reinterpret_cast<const void *>(extensionFn)),
- std::make_unique<Extension>(extensionFn));
+ return addExtension(TypeID::getFromOpaquePointer(
+ reinterpret_cast<const void *>(extensionFn)),
+ std::make_unique<Extension>(extensionFn));
}
/// Returns true if the current registry is a subset of 'rhs', i.e. if 'rhs'
@@ -256,9 +256,7 @@ class DialectRegistry {
private:
MapTy registry;
- using KeyExtensionPair =
- std::pair<TypeID, std::unique_ptr<DialectExtensionBase>>;
- llvm::SmallVector<KeyExtensionPair> extensions;
+ llvm::MapVector<TypeID, std::unique_ptr<DialectExtensionBase>> extensions;
};
} // namespace mlir
diff --git a/mlir/lib/IR/Dialect.cpp b/mlir/lib/IR/Dialect.cpp
index 30fdb14703e75..84b4e529346a8 100644
--- a/mlir/lib/IR/Dialect.cpp
+++ b/mlir/lib/IR/Dialect.cpp
@@ -11,15 +11,20 @@
#include "mlir/IR/Diagnostics.h"
#include "mlir/IR/DialectImplementation.h"
#include "mlir/IR/DialectInterface.h"
+#include "mlir/IR/DialectRegistry.h"
#include "mlir/IR/ExtensibleDialect.h"
#include "mlir/IR/MLIRContext.h"
#include "mlir/IR/Operation.h"
#include "mlir/Support/TypeID.h"
#include "llvm/ADT/MapVector.h"
+#include "llvm/ADT/SetOperations.h"
+#include "llvm/ADT/SmallVector.h"
+#include "llvm/ADT/SmallVectorExtras.h"
#include "llvm/ADT/Twine.h"
#include "llvm/Support/Debug.h"
#include "llvm/Support/ManagedStatic.h"
#include "llvm/Support/Regex.h"
+#include <memory>
#define DEBUG_TYPE "dialect"
@@ -174,6 +179,42 @@ bool dialect_extension_detail::hasPromisedInterface(Dialect &dialect,
// DialectRegistry
//===----------------------------------------------------------------------===//
+namespace {
+template <typename Fn>
+void applyExtensionsFn(
+ Fn &&applyExtension,
+ const llvm::MapVector<TypeID, std::unique_ptr<DialectExtensionBase>>
+ &extensions) {
+ // Note: Additional extensions may be added while applying an extension.
+ // The iterators will be invalidated if extensions are added so we'll keep
+ // a copy of the extensions for ourselves.
+
+ const auto extractExtension =
+ [](const auto &entry) -> DialectExtensionBase * {
+ return entry.second.get();
+ };
+
+ auto startIt = extensions.begin(), endIt = extensions.end();
+ size_t count = 0;
+ while (startIt != endIt) {
+ count += endIt - startIt;
+
+ // Grab the subset of extensions we'll apply in this iteration.
+ const auto subset =
+ llvm::map_to_vector(llvm::make_range(startIt, endIt), extractExtension);
+
+ // Apply!
+ for (const auto *ext : subset) {
+ applyExtension(*ext);
+ }
+
+ // Book-keep for the next iteration.
+ startIt = extensions.begin() + count;
+ endIt = extensions.end();
+ }
+}
+} // namespace
+
DialectRegistry::DialectRegistry() { insert<BuiltinDialect>(); }
DialectAllocatorFunctionRef
@@ -259,9 +300,7 @@ void DialectRegistry::applyExtensions(Dialect *dialect) const {
extension.apply(ctx, requiredDialects);
};
- // Note: Additional extensions may be added while applying an extension.
- for (int i = 0; i < static_cast<int>(extensions.size()); ++i)
- applyExtension(*extensions[i].second);
+ applyExtensionsFn(applyExtension, extensions);
}
void DialectRegistry::applyExtensions(MLIRContext *ctx) const {
@@ -286,23 +325,15 @@ void DialectRegistry::applyExtensions(MLIRContext *ctx) const {
extension.apply(ctx, requiredDialects);
};
- // Note: Additional extensions may be added while applying an extension.
- for (int i = 0; i < static_cast<int>(extensions.size()); ++i)
- applyExtension(*extensions[i].second);
+ applyExtensionsFn(applyExtension, extensions);
}
bool DialectRegistry::isSubsetOf(const DialectRegistry &rhs) const {
// Check that all extension keys are present in 'rhs'.
- llvm::DenseSet<TypeID> rhsExtensionKeys;
- {
- auto rhsKeys = llvm::map_range(rhs.extensions,
- [](const auto &item) { return item.first; });
- rhsExtensionKeys.insert(rhsKeys.begin(), rhsKeys.end());
- }
-
- if (!llvm::all_of(extensions, [&rhsExtensionKeys](const auto &extension) {
- return rhsExtensionKeys.contains(extension.first);
- }))
+ const auto hasExtension = [&](const auto &key) {
+ return rhs.extensions.contains(key);
+ };
+ if (!llvm::all_of(make_first_range(extensions), hasExtension))
return false;
// Check that the current dialects fully overlap with the dialects in 'rhs'.
diff --git a/mlir/unittests/IR/DialectTest.cpp b/mlir/unittests/IR/DialectTest.cpp
index 4be392faa31d2..7dd6a01c3389b 100644
--- a/mlir/unittests/IR/DialectTest.cpp
+++ b/mlir/unittests/IR/DialectTest.cpp
@@ -178,7 +178,7 @@ TEST(Dialect, NestedDialectExtension) {
std::make_unique<DummyExtension>(&counter1, 100));
// Add one more extension. This should not crash.
int counter2 = 0;
- registry.addExtension(TypeID::get<DummyExtension>(),
+ registry.addExtension(TypeID::getFromOpaquePointer(&counter2),
std::make_unique<DummyExtension>(&counter2, 0));
// Load dialect and apply extensions.
>From 9f7fa2e8a2dc9160f6d006d8bd12dc44c76698ba Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nkalra at apple.com>
Date: Thu, 1 Aug 2024 12:08:56 -0700
Subject: [PATCH 7/8] add type id definitions
---
mlir/examples/transform/Ch2/lib/MyExtension.cpp | 3 +++
mlir/examples/transform/Ch3/lib/MyExtension.cpp | 3 +++
mlir/examples/transform/Ch4/lib/MyExtension.cpp | 3 +++
mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp | 2 ++
mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp | 2 ++
.../Bufferization/TransformOps/BufferizationTransformOps.cpp | 3 +++
mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp | 2 ++
mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp | 2 ++
mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp | 2 ++
mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp | 2 ++
mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp | 2 ++
mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp | 2 ++
.../SparseTensor/TransformOps/SparseTensorTransformOps.cpp | 3 +++
mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp | 2 ++
mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp | 2 ++
mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp | 2 ++
mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp | 2 ++
mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp | 2 ++
mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp | 2 ++
.../lib/Dialect/Transform/TestTransformDialectExtension.cpp | 2 ++
.../TilingInterface/TestTilingInterfaceTransformOps.cpp | 3 +++
mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp | 2 ++
22 files changed, 50 insertions(+)
diff --git a/mlir/examples/transform/Ch2/lib/MyExtension.cpp b/mlir/examples/transform/Ch2/lib/MyExtension.cpp
index 68d538a098018..b4b27e97d266e 100644
--- a/mlir/examples/transform/Ch2/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch2/lib/MyExtension.cpp
@@ -29,6 +29,9 @@
class MyExtension
: public ::mlir::transform::TransformDialectExtension<MyExtension> {
public:
+ // The TypeID of this extension.
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(MyExtension)
+
// The extension must derive the base constructor.
using Base::Base;
diff --git a/mlir/examples/transform/Ch3/lib/MyExtension.cpp b/mlir/examples/transform/Ch3/lib/MyExtension.cpp
index f7a99423a52ee..4b2123fa71d31 100644
--- a/mlir/examples/transform/Ch3/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch3/lib/MyExtension.cpp
@@ -35,6 +35,9 @@
class MyExtension
: public ::mlir::transform::TransformDialectExtension<MyExtension> {
public:
+ // The TypeID of this extension.
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(MyExtension)
+
// The extension must derive the base constructor.
using Base::Base;
diff --git a/mlir/examples/transform/Ch4/lib/MyExtension.cpp b/mlir/examples/transform/Ch4/lib/MyExtension.cpp
index 38c8ca1125a24..fa0ffc9dc2e8a 100644
--- a/mlir/examples/transform/Ch4/lib/MyExtension.cpp
+++ b/mlir/examples/transform/Ch4/lib/MyExtension.cpp
@@ -31,6 +31,9 @@
class MyExtension
: public ::mlir::transform::TransformDialectExtension<MyExtension> {
public:
+ // The TypeID of this extension.
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(MyExtension)
+
// The extension must derive the base constructor.
using Base::Base;
diff --git a/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp b/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
index 6135117348a5b..b2407a258c271 100644
--- a/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
+++ b/mlir/lib/Conversion/ConvertToLLVM/ConvertToLLVMPass.cpp
@@ -35,6 +35,8 @@ namespace {
/// starting a pass pipeline that involves dialect conversion to LLVM.
class LoadDependentDialectExtension : public DialectExtensionBase {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(LoadDependentDialectExtension)
+
LoadDependentDialectExtension() : DialectExtensionBase(/*dialectNames=*/{}) {}
void apply(MLIRContext *context,
diff --git a/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp b/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
index 6457655cfe416..eb52297940722 100644
--- a/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
+++ b/mlir/lib/Dialect/Affine/TransformOps/AffineTransformOps.cpp
@@ -157,6 +157,8 @@ class AffineTransformDialectExtension
: public transform::TransformDialectExtension<
AffineTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(AffineTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp b/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
index e10c7bd914e35..a1d7bb995fc73 100644
--- a/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
+++ b/mlir/lib/Dialect/Bufferization/TransformOps/BufferizationTransformOps.cpp
@@ -150,6 +150,9 @@ class BufferizationTransformDialectExtension
: public transform::TransformDialectExtension<
BufferizationTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(
+ BufferizationTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp b/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
index b632b25d0cc67..2728936bf33fd 100644
--- a/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
+++ b/mlir/lib/Dialect/Func/TransformOps/FuncTransformOps.cpp
@@ -236,6 +236,8 @@ class FuncTransformDialectExtension
: public transform::TransformDialectExtension<
FuncTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(FuncTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp b/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
index 3661c5dea4525..1528da914d546 100644
--- a/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
+++ b/mlir/lib/Dialect/GPU/TransformOps/GPUTransformOps.cpp
@@ -924,6 +924,8 @@ class GPUTransformDialectExtension
: public transform::TransformDialectExtension<
GPUTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(GPUTransformDialectExtension)
+
GPUTransformDialectExtension() {
declareGeneratedDialect<scf::SCFDialect>();
declareGeneratedDialect<arith::ArithDialect>();
diff --git a/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp b/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
index f4244ca962232..4591802ce74ac 100644
--- a/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
+++ b/mlir/lib/Dialect/Linalg/TransformOps/DialectExtension.cpp
@@ -30,6 +30,8 @@ class LinalgTransformDialectExtension
: public transform::TransformDialectExtension<
LinalgTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(LinalgTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp b/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
index 8469e84c668cb..89640ac323b68 100644
--- a/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
+++ b/mlir/lib/Dialect/MemRef/TransformOps/MemRefTransformOps.cpp
@@ -309,6 +309,8 @@ class MemRefTransformDialectExtension
: public transform::TransformDialectExtension<
MemRefTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(MemRefTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp b/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
index 7d3d868b326c6..8f8cb7a8de25a 100644
--- a/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
+++ b/mlir/lib/Dialect/NVGPU/TransformOps/NVGPUTransformOps.cpp
@@ -1135,6 +1135,8 @@ class NVGPUTransformDialectExtension
: public transform::TransformDialectExtension<
NVGPUTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(NVGPUTransformDialectExtension)
+
NVGPUTransformDialectExtension() {
declareGeneratedDialect<arith::ArithDialect>();
declareGeneratedDialect<affine::AffineDialect>();
diff --git a/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp b/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
index c4a55c302d0a3..551411bb14765 100644
--- a/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
+++ b/mlir/lib/Dialect/SCF/TransformOps/SCFTransformOps.cpp
@@ -613,6 +613,8 @@ class SCFTransformDialectExtension
: public transform::TransformDialectExtension<
SCFTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(SCFTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp b/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
index ca19259ebffa6..bdec43825ddc2 100644
--- a/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
+++ b/mlir/lib/Dialect/SparseTensor/TransformOps/SparseTensorTransformOps.cpp
@@ -38,6 +38,9 @@ class SparseTensorTransformDialectExtension
: public transform::TransformDialectExtension<
SparseTensorTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(
+ SparseTensorTransformDialectExtension)
+
SparseTensorTransformDialectExtension() {
declareGeneratedDialect<sparse_tensor::SparseTensorDialect>();
registerTransformOps<
diff --git a/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp b/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
index 33016f84056e9..f911619d71227 100644
--- a/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
+++ b/mlir/lib/Dialect/Tensor/TransformOps/TensorTransformOps.cpp
@@ -236,6 +236,8 @@ class TensorTransformDialectExtension
: public transform::TransformDialectExtension<
TensorTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(TensorTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp b/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
index e369daddb00cb..d69535169f956 100644
--- a/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
+++ b/mlir/lib/Dialect/Transform/DebugExtension/DebugExtension.cpp
@@ -20,6 +20,8 @@ namespace {
class DebugExtension
: public transform::TransformDialectExtension<DebugExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(DebugExtension)
+
void init() {
registerTransformOps<
#define GET_OP_LIST
diff --git a/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp b/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
index 94004365b8a1a..9dc95490b14bb 100644
--- a/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
+++ b/mlir/lib/Dialect/Transform/IRDLExtension/IRDLExtension.cpp
@@ -18,6 +18,8 @@ namespace {
class IRDLExtension
: public transform::TransformDialectExtension<IRDLExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(IRDLExtension)
+
void init() {
registerTransformOps<
#define GET_OP_LIST
diff --git a/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp b/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
index b33288fd7b991..0a099b5bc75ab 100644
--- a/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
+++ b/mlir/lib/Dialect/Transform/LoopExtension/LoopExtension.cpp
@@ -20,6 +20,8 @@ namespace {
class LoopExtension
: public transform::TransformDialectExtension<LoopExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(LoopExtension)
+
void init() {
registerTransformOps<
#define GET_OP_LIST
diff --git a/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp b/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
index 2c770abd56d52..27c5dc332a428 100644
--- a/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
+++ b/mlir/lib/Dialect/Transform/PDLExtension/PDLExtension.cpp
@@ -38,6 +38,8 @@ namespace {
/// with Transform dialect operations.
class PDLExtension : public transform::TransformDialectExtension<PDLExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(PDLExtension)
+
void init() {
registerTransformOps<
#define GET_OP_LIST
diff --git a/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp b/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
index 2e9aa88011825..bc423a3781bf0 100644
--- a/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
+++ b/mlir/lib/Dialect/Vector/TransformOps/VectorTransformOps.cpp
@@ -212,6 +212,8 @@ class VectorTransformDialectExtension
: public transform::TransformDialectExtension<
VectorTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(VectorTransformDialectExtension)
+
VectorTransformDialectExtension() {
declareGeneratedDialect<vector::VectorDialect>();
declareGeneratedDialect<LLVM::LLVMDialect>();
diff --git a/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp b/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
index b8a4b9470d736..c023aad4a3ee7 100644
--- a/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
+++ b/mlir/test/lib/Dialect/Transform/TestTransformDialectExtension.cpp
@@ -874,6 +874,8 @@ class TestTransformDialectExtension
: public transform::TransformDialectExtension<
TestTransformDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(TestTransformDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp b/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
index 8f206d9077272..02c336de0be99 100644
--- a/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
+++ b/mlir/test/lib/Interfaces/TilingInterface/TestTilingInterfaceTransformOps.cpp
@@ -386,6 +386,9 @@ class TestTilingInterfaceDialectExtension
: public transform::TransformDialectExtension<
TestTilingInterfaceDialectExtension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(
+ TestTilingInterfaceDialectExtension)
+
using Base::Base;
void init() {
diff --git a/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp b/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
index 40fb752ffd6eb..d2a4999594a9e 100644
--- a/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
+++ b/mlir/unittests/Dialect/Transform/BuildOnlyExtensionTest.cpp
@@ -18,6 +18,8 @@ using namespace mlir::transform;
namespace {
class Extension : public TransformDialectExtension<Extension> {
public:
+ MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID(Extension)
+
using Base::Base;
void init() { declareGeneratedDialect<func::FuncDialect>(); }
};
>From 004d5be5c99ff6f8e9e3a466707e53f5971c7701 Mon Sep 17 00:00:00 2001
From: Nikhil Kalra <nikhil.kalra at gmail.com>
Date: Thu, 1 Aug 2024 16:10:47 -0700
Subject: [PATCH 8/8] Update mlir/lib/IR/Dialect.cpp
Co-authored-by: Mehdi Amini <joker.eph at gmail.com>
---
mlir/lib/IR/Dialect.cpp | 4 +---
1 file changed, 1 insertion(+), 3 deletions(-)
diff --git a/mlir/lib/IR/Dialect.cpp b/mlir/lib/IR/Dialect.cpp
index 84b4e529346a8..cc80677a4078f 100644
--- a/mlir/lib/IR/Dialect.cpp
+++ b/mlir/lib/IR/Dialect.cpp
@@ -203,10 +203,8 @@ void applyExtensionsFn(
const auto subset =
llvm::map_to_vector(llvm::make_range(startIt, endIt), extractExtension);
- // Apply!
- for (const auto *ext : subset) {
+ for (const auto *ext : subset)
applyExtension(*ext);
- }
// Book-keep for the next iteration.
startIt = extensions.begin() + count;
More information about the flang-commits
mailing list