[LLVM Pulldown] Bump LLVM to rev f2e244f443c99f814216bef37a82f1b8d3e5… (#1094)

Co-authored-by: Sang Ik Lee <[email protected]>

Author: Garra1980

build_tools/llvm_version.txt

@@ -1 +1 @@
-6b7e1b97f4bbb6dd6fca04ee4caccbf54ec92f09
+f2e244f443c99f814216bef37a82f1b8d3e5975d

build_tools/patches/0001-Add-support-for-VectorAnyINTEL-capability.patch

@@ -1,6 +1,6 @@
-From 46f1b01b6cde4956f4b08985adc59dd530788d4a Mon Sep 17 00:00:00 2001
+From 4167e203a75627ca13d8ea7560aaea9a6bb506f0 Mon Sep 17 00:00:00 2001
 From: Garra1980 <[email protected]>
-Date: Fri, 13 Jun 2025 19:12:11 +0200
+Date: Sat, 12 Jul 2025 00:39:57 +0200
 Subject: [PATCH] Add support for VectorAnyINTEL capability
 
 ---
@@ -24,10 +24,10 @@ Subject: [PATCH] Add support for VectorAnyINTEL capability
  17 files changed, 322 insertions(+), 69 deletions(-)
 
 diff --git a/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVBase.td b/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVBase.td
-index d2ba76cdad90..ac491f6068a0 100644
+index 910418f1706a..29af93d8e752 100644
 --- a/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVBase.td
 +++ b/mlir/include/mlir/Dialect/SPIRV/IR/SPIRVBase.td
-@@ -4194,7 +4194,14 @@ def SPIRV_BFloat16KHR : TypeAlias<BF16, "BFloat16">;
+@@ -4217,7 +4217,14 @@ def SPIRV_BFloat16KHR : TypeAlias<BF16, "BFloat16">;
  def SPIRV_Float : FloatOfWidths<[16, 32, 64]>;
  def SPIRV_Float16or32 : FloatOfWidths<[16, 32]>;
  def SPIRV_AnyFloat : AnyTypeOf<[SPIRV_Float, SPIRV_BFloat16KHR]>;
@@ -43,7 +43,7 @@ index d2ba76cdad90..ac491f6068a0 100644
                                         [SPIRV_Bool, SPIRV_Integer, SPIRV_AnyFloat]>;
  // Component type check is done in the type parser for the following SPIR-V
  // dialect-specific types so we use "Any" here.
-@@ -4245,7 +4252,7 @@ class SPIRV_MatrixOfType<list<Type> allowedTypes> :
+@@ -4270,7 +4277,7 @@ class SPIRV_MatrixOfType<list<Type> allowedTypes> :
      "Matrix">;
 
  class SPIRV_VectorOf<Type type> :
@@ -150,7 +150,7 @@ index 45ec1846580f..6ca59f91eee9 100644
  // Negative values for `n` index in reverse.
  class ShapedTypeWithNthDimOfSize<int n, list<int> allowedSizes> : Type<
 diff --git a/mlir/lib/Dialect/SPIRV/IR/SPIRVDialect.cpp b/mlir/lib/Dialect/SPIRV/IR/SPIRVDialect.cpp
-index a21acef1c4b4..a7c60f6bf1cb 100644
+index 88c7adf3dfcb..d29c88a1fd53 100644
 --- a/mlir/lib/Dialect/SPIRV/IR/SPIRVDialect.cpp
 +++ b/mlir/lib/Dialect/SPIRV/IR/SPIRVDialect.cpp
 @@ -188,9 +188,12 @@ static Type parseAndVerifyType(SPIRVDialect const &dialect,
@@ -169,10 +169,10 @@ index a21acef1c4b4..a7c60f6bf1cb 100644
        return Type();
      }
 diff --git a/mlir/lib/Dialect/SPIRV/IR/SPIRVTypes.cpp b/mlir/lib/Dialect/SPIRV/IR/SPIRVTypes.cpp
-index 93e0c9b33c54..a349da00027e 100644
+index 2b90df42af5c..34f25f2b3bc9 100644
 --- a/mlir/lib/Dialect/SPIRV/IR/SPIRVTypes.cpp
 +++ b/mlir/lib/Dialect/SPIRV/IR/SPIRVTypes.cpp
-@@ -100,9 +100,10 @@ bool CompositeType::classof(Type type) {
+@@ -101,9 +101,10 @@ bool CompositeType::classof(Type type) {
  }
 
  bool CompositeType::isValid(VectorType type) {
@@ -186,7 +186,7 @@ index 93e0c9b33c54..a349da00027e 100644
  }
 
  Type CompositeType::getElementType(unsigned index) const {
-@@ -164,7 +165,21 @@ void CompositeType::getCapabilities(
+@@ -174,7 +175,21 @@ void CompositeType::getCapabilities(
        .Case<VectorType>([&](VectorType type) {
          auto vecSize = getNumElements();
          if (vecSize == 8 || vecSize == 16) {
@@ -210,7 +210,7 @@ index 93e0c9b33c54..a349da00027e 100644
            capabilities.push_back(ref);
          }
 diff --git a/mlir/lib/Dialect/SPIRV/Transforms/SPIRVConversion.cpp b/mlir/lib/Dialect/SPIRV/Transforms/SPIRVConversion.cpp
-index f5a58c58e05d..406e81235e10 100644
+index 1e7bb046d375..24e633da72aa 100644
 --- a/mlir/lib/Dialect/SPIRV/Transforms/SPIRVConversion.cpp
 +++ b/mlir/lib/Dialect/SPIRV/Transforms/SPIRVConversion.cpp
 @@ -87,9 +87,13 @@ static std::optional<SmallVector<int64_t>> getTargetShape(VectorType vecType) {
@@ -499,15 +499,15 @@ index 1737f4a906bf..13f4e17167ef 100644
  } // end module
 
 diff --git a/mlir/test/Dialect/SPIRV/IR/arithmetic-ops.mlir b/mlir/test/Dialect/SPIRV/IR/arithmetic-ops.mlir
-index d58c27598f2b..4c22244c08e0 100644
+index 3adafc15c79f..f75fd6cb0d39 100644
 --- a/mlir/test/Dialect/SPIRV/IR/arithmetic-ops.mlir
 +++ b/mlir/test/Dialect/SPIRV/IR/arithmetic-ops.mlir
-@@ -339,7 +339,7 @@ func.func @dot(%arg0: vector<4xf32>, %arg1: vector<4xf32>) -> f16 {
+@@ -348,7 +348,7 @@ func.func @dot(%arg0: vector<4xf32>, %arg1: vector<4xf32>) -> f16 {
  // -----
 
  func.func @dot(%arg0: vector<4xi32>, %arg1: vector<4xi32>) -> i32 {
--  // expected-error @+1 {{'spirv.Dot' op operand #0 must be vector of 16/32/64-bit float values of length 2/3/4/8/16}}
-+  // expected-error @+1 {{op operand #0 must be vector of 16/32/64-bit float values of length 2-4294967295, but got 'vector<4xi32>'}}
+-  // expected-error @+1 {{'spirv.Dot' op operand #0 must be vector of 16/32/64-bit float or BFloat16 values of length 2/3/4/8/16}}
++  // expected-error @+1 {{op operand #0 must be vector of 16/32/64-bit float or BFloat16 values of length 2-4294967295, but got 'vector<4xi32>'}}
    %0 = spirv.Dot %arg0, %arg1 : vector<4xi32> -> i32
    return %0 : i32
  }
@@ -543,7 +543,7 @@ index f3f0ebf60f46..1138f38bcef2 100644
    return %0 : f16
  }
 diff --git a/mlir/test/Dialect/SPIRV/IR/gl-ops.mlir b/mlir/test/Dialect/SPIRV/IR/gl-ops.mlir
-index 642346cc40b0..10ede222ada7 100644
+index 5c5d94c40e57..8edaa3762c23 100644
 --- a/mlir/test/Dialect/SPIRV/IR/gl-ops.mlir
 +++ b/mlir/test/Dialect/SPIRV/IR/gl-ops.mlir
 @@ -27,7 +27,7 @@ func.func @exp(%arg0 : i32) -> () {
@@ -722,7 +722,7 @@ index 8f021ed3d663..21558b9607f8 100644
    // expected-error @+1 {{expected ':'}}
    %2 = spirv.CL.s_abs %arg0, %arg1 : i32
 diff --git a/mlir/test/Target/SPIRV/arithmetic-ops.mlir b/mlir/test/Target/SPIRV/arithmetic-ops.mlir
-index b1ea13c6854f..90144afc6f3a 100644
+index b80e17f979da..32103f7b9c57 100644
 --- a/mlir/test/Target/SPIRV/arithmetic-ops.mlir
 +++ b/mlir/test/Target/SPIRV/arithmetic-ops.mlir
 @@ -6,9 +6,9 @@ spirv.module Logical GLSL450 requires #spirv.vce<v1.0, [Shader], []> {

include/imex/Conversion/Passes.h

@@ -28,7 +28,6 @@
 #include <imex/Conversion/XeGPUToVC/XeGPUToVC.h>
 #include <imex/Conversion/XeGPUToXeVM/XeGPUToXeVM.h>
 #include <imex/Conversion/XeTileToXeGPU/XeTileToXeGPU.h>
-#include <imex/Conversion/XeVMToLLVM/XeVMToLLVM.h>
 
 namespace imex {
 

include/imex/Conversion/Passes.td

@@ -466,17 +466,6 @@ def ConvertArithToVC : Pass<"convert-arith-to-vc", "::mlir::gpu::GPUModuleOp"> {
   let constructor = "imex::createConvertArithToVCPass()";
 }
 
-//===----------------------------------------------------------------------===//
-// XeVMToLLVM
-//===----------------------------------------------------------------------===//
-
-def ConvertXeVMToLLVMPass : Pass<"convert-xevm-to-llvm"> {
-  let summary = "Convert XeVM to LLVM dialect";
-  let dependentDialects = [
-    "imex::xevm::XeVMDialect",
-  ];
-}
-
 //===----------------------------------------------------------------------===//
 // XeGPUToXeVM
 //===----------------------------------------------------------------------===//
@@ -485,7 +474,7 @@ def ConvertXeGPUToXeVMPass : Pass<"convert-xegpu-to-xevm"> {
   let summary = "Convert XeGPU to XeVM dialect";
   let dependentDialects = [
     "::mlir::xegpu::XeGPUDialect",
-    "::imex::xevm::XeVMDialect",
+    "::mlir::xevm::XeVMDialect",
     "::mlir::vector::VectorDialect",
     "::mlir::memref::MemRefDialect",
     "::mlir::arith::ArithDialect",

include/imex/Conversion/XeVMToLLVM/Mangling.h

@@ -2,5 +2,4 @@ add_subdirectory(DistRuntime)
 add_subdirectory(NDArray)
 add_subdirectory(Region)
 add_subdirectory(GPUX)
-add_subdirectory(LLVMIR)
 add_subdirectory(XeTile)