[PATCH] D98757: [AMX] Not fold constant bitcast into amx intrisic

[Xiang Zhang via Phabricator via cfe-commits]

xiangzhangllvm created this revision.
xiangzhangllvm added reviewers: LuoYuanke, pengfei, LiuChen3, yubing.
Herald added a subscriber: hiraditya.
xiangzhangllvm requested review of this revision.
Herald added projects: clang, LLVM.
Herald added subscribers: llvm-commits, cfe-commits.

We won't fold bitcast for tile type, becasue there is no way to
assignee a tmm reg from a constant. We manually generate tilestore
and tileload at pass "Lower AMX type".


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D98757

Files:
  clang/test/CodeGen/X86/amx_api.c
  llvm/lib/Analysis/ConstantFolding.cpp
  llvm/lib/Transforms/Scalar/SCCP.cpp


Index: llvm/lib/Transforms/Scalar/SCCP.cpp
===================================================================
--- llvm/lib/Transforms/Scalar/SCCP.cpp
+++ llvm/lib/Transforms/Scalar/SCCP.cpp
@@ -826,7 +826,7 @@
   if (Constant *OpC = getConstant(OpSt)) {
     // Fold the constant as we build.
     Constant *C = ConstantFoldCastOperand(I.getOpcode(), OpC, I.getType(), DL);
-    if (isa<UndefValue>(C))
+    if (!C || isa<UndefValue>(C))
       return;
     // Propagate constant value
     markConstant(&I, C);
Index: llvm/lib/Analysis/ConstantFolding.cpp
===================================================================
--- llvm/lib/Analysis/ConstantFolding.cpp
+++ llvm/lib/Analysis/ConstantFolding.cpp
@@ -104,10 +104,16 @@
   assert(CastInst::castIsValid(Instruction::BitCast, C, DestTy) &&
          "Invalid constantexpr bitcast!");
 
+  // We won't fold bitcast for tile type, becasue there is no way to
+  // assigne a tmm reg from a constant. We manually generate tilestore
+  // and tileload at pass "Lower AMX type".
+  if (DestTy->isX86_AMXTy())
+    return nullptr;
+
   // Catch the obvious splat cases.
-  if (C->isNullValue() && !DestTy->isX86_MMXTy() && !DestTy->isX86_AMXTy())
+  if (C->isNullValue() && !DestTy->isX86_MMXTy())
     return Constant::getNullValue(DestTy);
-  if (C->isAllOnesValue() && !DestTy->isX86_MMXTy() && !DestTy->isX86_AMXTy() &&
+  if (C->isAllOnesValue() && !DestTy->isX86_MMXTy() &&
       !DestTy->isPtrOrPtrVectorTy()) // Don't get ones for ptr types!
     return Constant::getAllOnesValue(DestTy);
 
Index: clang/test/CodeGen/X86/amx_api.c
===================================================================
--- clang/test/CodeGen/X86/amx_api.c
+++ clang/test/CodeGen/X86/amx_api.c
@@ -1,6 +1,9 @@
 // RUN: %clang_cc1 %s -flax-vector-conversions=none -ffreestanding -triple=x86_64-unknown-unknown  -target-feature +avx512f  -target-feature +amx-int8  \
 // RUN: -target-feature +amx-bf16 -emit-llvm -o - -Werror -pedantic | FileCheck %s --check-prefixes=CHECK
 
+// RUN: %clang_cc1 %s -flax-vector-conversions=none -ffreestanding -triple=x86_64-unknown-unknown  -target-feature +avx512f  -target-feature +amx-int8  \
+// RUN: -target-feature +amx-bf16 -O2 -emit-llvm -o - -Werror -pedantic | FileCheck %s --check-prefixes=CHECK2
+
 #include <immintrin.h>
 
 char buf[1024];
@@ -31,6 +34,15 @@
   __tile_stored(buf, STRIDE, c);
 }
 
+// Not fold the bitcast const vector into amx intrisic.
+void test_tile_init(short row, short col) {
+  __tile1024i c = {row, col, {1, 2, 3}};
+  __tile_stored(buf, STRIDE, c);
+  //CHECK2-LABEL: @test_tile_init
+  //CHECK2: {{%.*}} = bitcast <256 x i32> <i32 1, i32 2, i32 3, i32 0,
+  //CHECK2-NEXT: call void @llvm.x86.tilestored64.internal({{.*}}, x86_amx {{%.*}})
+}
+
 void test_tile_loadd(short row, short col) {
   //CHECK-LABEL: @test_tile_loadd
   //CHECK: call x86_amx @llvm.x86.tileloadd64.internal


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D98757.331148.patch
Type: text/x-patch
Size: 2897 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20210317/f0bd164f/attachment-0001.bin>