[Mlir-commits] [mlir] [mlir][acc] Add attributes for parallelism dimensions (PR #182209)

Razvan Lupusoru llvmlistbot at llvm.org
Wed Feb 18 19:02:13 PST 2026


================
@@ -119,6 +123,80 @@ static void addResultEffect(
   effects.emplace_back(EffectTy::get(), mlir::cast<mlir::OpResult>(result));
 }
 
+static int64_t gpuProcessorIndex(gpu::Processor p) {
+  switch (p) {
+  case gpu::Processor::Sequential:
+    return 0;
+  case gpu::Processor::ThreadX:
+    return 1;
+  case gpu::Processor::ThreadY:
+    return 2;
+  case gpu::Processor::ThreadZ:
+    return 3;
+  case gpu::Processor::BlockX:
+    return 4;
+  case gpu::Processor::BlockY:
+    return 5;
+  case gpu::Processor::BlockZ:
+    return 6;
+  }
+  llvm_unreachable("unhandled gpu::Processor");
+}
+
+static gpu::Processor indexToGpuProcessor(int64_t idx) {
+  switch (idx) {
+  case 0:
+    return gpu::Processor::Sequential;
+  case 1:
+    return gpu::Processor::ThreadX;
+  case 2:
+    return gpu::Processor::ThreadY;
+  case 3:
+    return gpu::Processor::ThreadZ;
+  case 4:
+    return gpu::Processor::BlockX;
+  case 5:
+    return gpu::Processor::BlockY;
+  case 6:
+    return gpu::Processor::BlockZ;
+  default:
+    return gpu::Processor::Sequential;
+  }
+}
+
+static GPUParallelDimAttr intToParDim(MLIRContext *context, int64_t dimInt) {
+  return GPUParallelDimAttr::get(
+      context, IntegerAttr::get(IndexType::get(context), dimInt));
+}
+
+static GPUParallelDimAttr processorParDim(MLIRContext *context,
+                                          gpu::Processor proc) {
+  return GPUParallelDimAttr::get(
+      context,
+      IntegerAttr::get(IndexType::get(context), gpuProcessorIndex(proc)));
+}
+
+static ParseResult parseProcessorValue(AsmParser &parser,
+                                       GPUParallelDimAttr &dim) {
+  std::string keyword;
+  llvm::SMLoc loc = parser.getCurrentLocation();
+  if (failed(parser.parseKeywordOrString(&keyword)))
+    return failure();
+  auto maybeProcessor = gpu::symbolizeProcessor(keyword);
+  if (!maybeProcessor) {
+    return parser.emitError(loc)
+           << "expected one of ::mlir::gpu::Processor enum names";
+  }
----------------
razvanlupusoru wrote:

Done.

https://github.com/llvm/llvm-project/pull/182209


More information about the Mlir-commits mailing list