[MLIR][SPIRVToLLVM] Remove typed pointer support #70568

Dinistro · 2023-10-28T19:42:25Z

This commit removes the support for lowering SPIRV to LLVM dialect with typed pointers. Typed pointers have been deprecated for a while now and it's planned to soon remove them from the LLVM dialect.

Related PSA: https://discourse.llvm.org/t/psa-removal-of-typed-pointers-from-the-llvm-dialect/74502

This commit removes the support for lowering SPIRV to LLVM dialect with typed pointers. Typed pointers are deprecated for a while now and it's planned to soon remove them from the LLVM dialect.

llvmbot · 2023-10-28T20:02:06Z

@llvm/pr-subscribers-mlir-llvm
@llvm/pr-subscribers-mlir-spirv

@llvm/pr-subscribers-mlir

Author: Christian Ulmann (Dinistro)

Changes

This commit removes the support for lowering SPIRV to LLVM dialect with typed pointers. Typed pointers have been deprecated for a while now and it's planned to soon remove them from the LLVM dialect.

I'm making this a draft for now, as we first require a PSA for this change.

Full diff: https://github.com/llvm/llvm-project/pull/70568.diff

22 Files Affected:

(modified) mlir/include/mlir/Conversion/Passes.td (-3)
(modified) mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVM.cpp (+2-2)
(modified) mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVMPass.cpp (-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/arithmetic-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/bitwise-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/cast-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/comparison-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/constant-op-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/control-flow-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/func-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/gl-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/logical-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls_fail.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/memory-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/misc-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/module-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/shift-ops-to-llvm.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/spirv-storage-class-mapping.mlir (+2-2)
(modified) mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm-invalid.mlir (+1-1)
(modified) mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm.mlir (+1-1)
(removed) mlir/test/Conversion/SPIRVToLLVM/typed-pointers.mlir (-18)

diff --git a/mlir/include/mlir/Conversion/Passes.td b/mlir/include/mlir/Conversion/Passes.td
index 336f0d3af951b9a..cf6e545749ffc64 100644
--- a/mlir/include/mlir/Conversion/Passes.td
+++ b/mlir/include/mlir/Conversion/Passes.td
@@ -1019,9 +1019,6 @@ def ConvertSPIRVToLLVMPass : Pass<"convert-spirv-to-llvm", "ModuleOp"> {
   let dependentDialects = ["LLVM::LLVMDialect"];
 
   let options = [
-    Option<"useOpaquePointers", "use-opaque-pointers", "bool",
-                 /*default=*/"true", "Generate LLVM IR using opaque pointers "
-                 "instead of typed pointers">,
     Option<"clientAPI", "client-api", "::mlir::spirv::ClientAPI",
 	   /*default=*/"::mlir::spirv::ClientAPI::Unknown",
 	   "Derive StorageClass to address space mapping from the client API",
diff --git a/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVM.cpp b/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVM.cpp
index 60f34f413f587d4..a54163db9e1050e 100644
--- a/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVM.cpp
+++ b/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVM.cpp
@@ -1379,8 +1379,8 @@ class BitcastConversionPattern
     if (!dstType)
       return failure();
 
-    if (typeConverter.useOpaquePointers() &&
-        isa<LLVM::LLVMPointerType>(dstType)) {
+    // LLVM's opaque pointers do not require bitcasts.
+    if (isa<LLVM::LLVMPointerType>(dstType)) {
       rewriter.replaceOp(bitcastOp, adaptor.getOperand());
       return success();
     }
diff --git a/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVMPass.cpp b/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVMPass.cpp
index 40798e9eb9dcbae..38091e449c56ee8 100644
--- a/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVMPass.cpp
+++ b/mlir/lib/Conversion/SPIRVToLLVM/SPIRVToLLVMPass.cpp
@@ -42,7 +42,6 @@ void ConvertSPIRVToLLVMPass::runOnOperation() {
   ModuleOp module = getOperation();
 
   LowerToLLVMOptions options(&getContext());
-  options.useOpaquePointers = useOpaquePointers;
 
   LLVMTypeConverter converter(&getContext(), options);
 
diff --git a/mlir/test/Conversion/SPIRVToLLVM/arithmetic-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/arithmetic-ops-to-llvm.mlir
index 925443758c43db1..dbbf8610afb4dfb 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/arithmetic-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/arithmetic-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.IAdd
diff --git a/mlir/test/Conversion/SPIRVToLLVM/bitwise-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/bitwise-ops-to-llvm.mlir
index af232bb2c6387d0..a0afe0dafcaa248 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/bitwise-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/bitwise-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.BitCount
diff --git a/mlir/test/Conversion/SPIRVToLLVM/cast-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/cast-ops-to-llvm.mlir
index 4f6c1eaaf50487d..2026027a6dacee7 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/cast-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/cast-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.Bitcast
diff --git a/mlir/test/Conversion/SPIRVToLLVM/comparison-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/comparison-ops-to-llvm.mlir
index 272ee309fdf6d78..52359db3be7bdc0 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/comparison-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/comparison-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.IEqual
diff --git a/mlir/test/Conversion/SPIRVToLLVM/constant-op-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/constant-op-to-llvm.mlir
index f66b0768a614360..2d74022b344066a 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/constant-op-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/constant-op-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.Constant
diff --git a/mlir/test/Conversion/SPIRVToLLVM/control-flow-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/control-flow-ops-to-llvm.mlir
index 54ef71f75f528fc..3557830e779e240 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/control-flow-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/control-flow-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' -split-input-file -verify-diagnostics %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm -split-input-file -verify-diagnostics %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.Branch
diff --git a/mlir/test/Conversion/SPIRVToLLVM/func-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/func-ops-to-llvm.mlir
index 1dff2c48cefeead..5b3d8ba5ca59587 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/func-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/func-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.Return
diff --git a/mlir/test/Conversion/SPIRVToLLVM/gl-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/gl-ops-to-llvm.mlir
index 851d164284c48d9..e1936e2fd8abea0 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/gl-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/gl-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.GL.Ceil
diff --git a/mlir/test/Conversion/SPIRVToLLVM/logical-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/logical-ops-to-llvm.mlir
index 65db67314361f5a..6d93480d3ed142e 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/logical-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/logical-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.LogicalEqual
diff --git a/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls.mlir b/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls.mlir
index 71149f138940744..61944b9d047e49b 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt --lower-host-to-llvm='use-opaque-pointers=1' %s -split-input-file | FileCheck %s
+// RUN: mlir-opt --lower-host-to-llvm %s -split-input-file | FileCheck %s
 
 module attributes {gpu.container_module, spirv.target_env = #spirv.target_env<#spirv.vce<v1.0, [Shader], [SPV_KHR_variable_pointers]>, #spirv.resource_limits<max_compute_workgroup_invocations = 128, max_compute_workgroup_size = [128, 128, 64]>>} {
 
diff --git a/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls_fail.mlir b/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls_fail.mlir
index 05ed60825261365..e36d30b434fc76e 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls_fail.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/lower-host-to-llvm-calls_fail.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt --lower-host-to-llvm='use-opaque-pointers=1' %s -verify-diagnostics
+// RUN: mlir-opt --lower-host-to-llvm %s -verify-diagnostics
 
 module {
 // expected-error @+1 {{The module must contain exactly one entry point function}}
diff --git a/mlir/test/Conversion/SPIRVToLLVM/memory-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/memory-ops-to-llvm.mlir
index 04357a1b00a5b86..1847975b279afae 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/memory-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/memory-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.AccessChain
diff --git a/mlir/test/Conversion/SPIRVToLLVM/misc-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/misc-ops-to-llvm.mlir
index b90022b25265dc7..13bde6e6fc56305 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/misc-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/misc-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.CompositeExtract
diff --git a/mlir/test/Conversion/SPIRVToLLVM/module-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/module-ops-to-llvm.mlir
index d4cb88db720ca5f..894de8b9e90a924 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/module-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/module-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.module
diff --git a/mlir/test/Conversion/SPIRVToLLVM/shift-ops-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/shift-ops-to-llvm.mlir
index db8f207193c9a98..da4def7aeda82ae 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/shift-ops-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/shift-ops-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' %s | FileCheck %s
+// RUN: mlir-opt -convert-spirv-to-llvm %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spirv.ShiftRightArithmetic
diff --git a/mlir/test/Conversion/SPIRVToLLVM/spirv-storage-class-mapping.mlir b/mlir/test/Conversion/SPIRVToLLVM/spirv-storage-class-mapping.mlir
index 989ada93cf36ee7..b3991cbdbe8af1e 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/spirv-storage-class-mapping.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/spirv-storage-class-mapping.mlir
@@ -1,5 +1,5 @@
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1' -verify-diagnostics %s | FileCheck %s --check-prefixes=CHECK-UNKNOWN,CHECK-ALL
-// RUN: mlir-opt -convert-spirv-to-llvm='use-opaque-pointers=1 client-api=OpenCL' -verify-diagnostics %s | FileCheck %s --check-prefixes=CHECK-OPENCL,CHECK-ALL
+// RUN: mlir-opt -convert-spirv-to-llvm -verify-diagnostics %s | FileCheck %s --check-prefixes=CHECK-UNKNOWN,CHECK-ALL
+// RUN: mlir-opt -convert-spirv-to-llvm='client-api=OpenCL' -verify-diagnostics %s | FileCheck %s --check-prefixes=CHECK-OPENCL,CHECK-ALL
 
 // CHECK-OPENCL:         llvm.func @pointerUniformConstant(!llvm.ptr<2>)
 // CHECK-UNKNOWN:        llvm.func @pointerUniformConstant(!llvm.ptr)
diff --git a/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm-invalid.mlir b/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm-invalid.mlir
index 3965c47ec199fcb..084652aff9583d6 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm-invalid.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm-invalid.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt %s -convert-spirv-to-llvm='use-opaque-pointers=1' -verify-diagnostics -split-input-file
+// RUN: mlir-opt %s -convert-spirv-to-llvm -verify-diagnostics -split-input-file
 
 // expected-error@+1 {{failed to legalize operation 'spirv.func' that was explicitly marked illegal}}
 spirv.func @array_with_unnatural_stride(%arg: !spirv.array<4 x f32, stride=8>) -> () "None" {
diff --git a/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm.mlir b/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm.mlir
index 167ad021a5fa1dc..0f2dbf8ef115587 100644
--- a/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm.mlir
+++ b/mlir/test/Conversion/SPIRVToLLVM/spirv-types-to-llvm.mlir
@@ -1,4 +1,4 @@
-// RUN: mlir-opt -split-input-file -convert-spirv-to-llvm='use-opaque-pointers=1' -verify-diagnostics %s | FileCheck %s
+// RUN: mlir-opt -split-input-file -convert-spirv-to-llvm -verify-diagnostics %s | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // Array type
diff --git a/mlir/test/Conversion/SPIRVToLLVM/typed-pointers.mlir b/mlir/test/Conversion/SPIRVToLLVM/typed-pointers.mlir
deleted file mode 100644
index 2c56f42a1fd523a..000000000000000
--- a/mlir/test/Conversion/SPIRVToLLVM/typed-pointers.mlir
+++ /dev/null
@@ -1,18 +0,0 @@
-// RUN: mlir-opt -split-input-file -convert-spirv-to-llvm='use-opaque-pointers=0' %s | FileCheck %s
-
-//===----------------------------------------------------------------------===//
-// Pointer type
-//===----------------------------------------------------------------------===//
-
-// CHECK-LABEL: @pointer_scalar(!llvm.ptr<i1>, !llvm.ptr<f32>)
-spirv.func @pointer_scalar(!spirv.ptr<i1, Uniform>, !spirv.ptr<f32, Private>) "None"
-
-// CHECK-LABEL: @pointer_vector(!llvm.ptr<vector<4xi32>>)
-spirv.func @pointer_vector(!spirv.ptr<vector<4xi32>, Function>) "None"
-
-// CHECK-LABEL: @bitcast_pointer
-spirv.func @bitcast_pointer(%arg0: !spirv.ptr<f32, Function>) "None" {
-  // CHECK: llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<i32>
-  %0 = spirv.Bitcast %arg0 : !spirv.ptr<f32, Function> to !spirv.ptr<i32, Function>
-  spirv.Return
-}

zero9178

LGTM

Also adding some spir-v reviewers for visibility

antiagainst

Cool, thanks!

llvm#70568 removed the support for lowering SPIRV to LLVM dialect. We now need to stop using enableOpaquePointers with ConvertSPIRVToLLVMPassOptions.

#70568 removed the support for lowering SPIRV to LLVM dialect. We now need to stop using enableOpaquePointers with ConvertSPIRVToLLVMPassOptions.

[MLIR][SPIRVToLLVM] Remove typed pointer support

d2415d6

This commit removes the support for lowering SPIRV to LLVM dialect with typed pointers. Typed pointers are deprecated for a while now and it's planned to soon remove them from the LLVM dialect.

Dinistro requested review from zero9178 and victor-eds October 28, 2023 19:42

Dinistro added mlir mlir:spirv mlir:llvm labels Oct 28, 2023

Dinistro marked this pull request as ready for review October 29, 2023 09:57

zero9178 approved these changes Oct 29, 2023

View reviewed changes

zero9178 requested review from antiagainst and kuhar October 29, 2023 12:48

antiagainst approved these changes Oct 29, 2023

View reviewed changes

kuhar approved these changes Oct 30, 2023

View reviewed changes

Dinistro merged commit 41f3b83 into llvm:main Oct 30, 2023

Dinistro deleted the spirv-to-llvm-remove-typed-ptr branch October 30, 2023 07:57

tsitdikov added a commit to tsitdikov/llvm-project that referenced this pull request Oct 30, 2023

Update mlir-spirv-cpu-runner.cpp

b6bbb8e

llvm#70568 removed the support for lowering SPIRV to LLVM dialect. We now need to stop using enableOpaquePointers with ConvertSPIRVToLLVMPassOptions.

tsitdikov mentioned this pull request Oct 30, 2023

Update mlir-spirv-cpu-runner.cpp #70649

Merged

akuegel pushed a commit that referenced this pull request Oct 30, 2023

Update mlir-spirv-cpu-runner.cpp (#70649)

97f0595

#70568 removed the support for lowering SPIRV to LLVM dialect. We now need to stop using enableOpaquePointers with ConvertSPIRVToLLVMPassOptions.

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Uh oh!

[MLIR][SPIRVToLLVM] Remove typed pointer support #70568

[MLIR][SPIRVToLLVM] Remove typed pointer support #70568

Uh oh!

Dinistro commented Oct 28, 2023 •

edited

Loading

Uh oh!

llvmbot commented Oct 28, 2023 •

edited

Loading

Uh oh!

zero9178 left a comment

Uh oh!

antiagainst left a comment

Uh oh!

Uh oh!

[MLIR][SPIRVToLLVM] Remove typed pointer support #70568

[MLIR][SPIRVToLLVM] Remove typed pointer support #70568

Uh oh!

Conversation

Dinistro commented Oct 28, 2023 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

llvmbot commented Oct 28, 2023 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Uh oh!

zero9178 left a comment

Choose a reason for hiding this comment

Uh oh!

antiagainst left a comment

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Dinistro commented Oct 28, 2023 •

edited

Loading

llvmbot commented Oct 28, 2023 •

edited

Loading