[Mlir-commits] [mlir] [libc] [llvm] [mlir][Tensor] Remove unnecessary `#include` from TestTensorTransforms.cpp (PR #80383)
llvmlistbot at llvm.org
llvmlistbot at llvm.org
Thu Feb 1 20:08:01 PST 2024
llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT-->
@llvm/pr-subscribers-mlir
Author: Justin Fargnoli (justinfargnoli)
<details>
<summary>Changes</summary>
---
Patch is 27.78 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/80383.diff
3 Files Affected:
- (modified) llvm/lib/Target/NVPTX/NVPTXInstrInfo.td (-13)
- (modified) llvm/test/CodeGen/NVPTX/i8x4-instructions.ll (+291-171)
- (modified) mlir/test/lib/Dialect/Tensor/TestTensorTransforms.cpp (-1)
``````````diff
diff --git a/llvm/lib/Target/NVPTX/NVPTXInstrInfo.td b/llvm/lib/Target/NVPTX/NVPTXInstrInfo.td
index e1cced3275449..351d8d5371756 100644
--- a/llvm/lib/Target/NVPTX/NVPTXInstrInfo.td
+++ b/llvm/lib/Target/NVPTX/NVPTXInstrInfo.td
@@ -2258,19 +2258,6 @@ def: Pat<(setlt (sext_inreg (trunc Int32Regs:$a), i8), (sext_inreg (trunc Int32R
def: Pat<(setle (sext_inreg (trunc Int32Regs:$a), i8), (sext_inreg (trunc Int32Regs:$b), i8)),
(SETP_s32rr Int32Regs:$a, Int32Regs:$b, CmpLE)>;
-def: Pat<(setugt (i16 (and (trunc Int32Regs:$a), 255)), (i16 (and (trunc Int32Regs:$b), 255))),
- (SETP_u32rr Int32Regs:$a, Int32Regs:$b, CmpHI)>;
-def: Pat<(setuge (i16 (and (trunc Int32Regs:$a), 255)), (i16 (and (trunc Int32Regs:$b), 255))),
- (SETP_u32rr Int32Regs:$a, Int32Regs:$b, CmpHS)>;
-def: Pat<(setult (i16 (and (trunc Int32Regs:$a), 255)), (i16 (and (trunc Int32Regs:$b), 255))),
- (SETP_u32rr Int32Regs:$a, Int32Regs:$b, CmpLO)>;
-def: Pat<(setule (i16 (and (trunc Int32Regs:$a), 255)), (i16 (and (trunc Int32Regs:$b), 255))),
- (SETP_u32rr Int32Regs:$a, Int32Regs:$b, CmpLS)>;
-def: Pat<(seteq (i16 (and (trunc Int32Regs:$a), 255)), (i16 (and (trunc Int32Regs:$b), 255))),
- (SETP_u32rr Int32Regs:$a, Int32Regs:$b, CmpEQ)>;
-def: Pat<(setne (i16 (and (trunc Int32Regs:$a), 255)), (i16 (and (trunc Int32Regs:$b), 255))),
- (SETP_u32rr Int32Regs:$a, Int32Regs:$b, CmpNE)>;
-
// i1 compare -> i32
def : Pat<(i32 (setne Int1Regs:$a, Int1Regs:$b)),
(SELP_u32ii -1, 0, (XORb1rr Int1Regs:$a, Int1Regs:$b))>;
diff --git a/llvm/test/CodeGen/NVPTX/i8x4-instructions.ll b/llvm/test/CodeGen/NVPTX/i8x4-instructions.ll
index 1ec68b4a271ba..c03081ce68362 100644
--- a/llvm/test/CodeGen/NVPTX/i8x4-instructions.ll
+++ b/llvm/test/CodeGen/NVPTX/i8x4-instructions.ll
@@ -285,33 +285,51 @@ define <4 x i8> @test_smax(<4 x i8> %a, <4 x i8> %b) #0 {
define <4 x i8> @test_umax(<4 x i8> %a, <4 x i8> %b) #0 {
; CHECK-LABEL: test_umax(
; CHECK: {
-; CHECK-NEXT: .reg .pred %p<5>;
-; CHECK-NEXT: .reg .b32 %r<19>;
+; CHECK-NEXT: .reg .pred %p<5>;
+; CHECK-NEXT: .reg .b16 %rs<17>;
+; CHECK-NEXT: .reg .b32 %r<19>;
; CHECK-EMPTY:
; CHECK-NEXT: // %bb.0:
-; CHECK-NEXT: ld.param.u32 %r2, [test_umax_param_1];
-; CHECK-NEXT: ld.param.u32 %r1, [test_umax_param_0];
-; CHECK-NEXT: bfe.s32 %r3, %r1, 24, 8;
-; CHECK-NEXT: bfe.s32 %r4, %r2, 24, 8;
-; CHECK-NEXT: setp.hi.u32 %p1, %r3, %r4;
-; CHECK-NEXT: bfe.s32 %r5, %r1, 16, 8;
-; CHECK-NEXT: bfe.s32 %r6, %r2, 16, 8;
-; CHECK-NEXT: setp.hi.u32 %p2, %r5, %r6;
-; CHECK-NEXT: bfe.s32 %r7, %r1, 8, 8;
-; CHECK-NEXT: bfe.s32 %r8, %r2, 8, 8;
-; CHECK-NEXT: setp.hi.u32 %p3, %r7, %r8;
-; CHECK-NEXT: bfe.s32 %r9, %r1, 0, 8;
-; CHECK-NEXT: bfe.s32 %r10, %r2, 0, 8;
-; CHECK-NEXT: setp.hi.u32 %p4, %r9, %r10;
-; CHECK-NEXT: selp.b32 %r11, %r9, %r10, %p4;
-; CHECK-NEXT: selp.b32 %r12, %r7, %r8, %p3;
-; CHECK-NEXT: bfi.b32 %r13, %r12, %r11, 8, 8;
-; CHECK-NEXT: selp.b32 %r14, %r5, %r6, %p2;
-; CHECK-NEXT: bfi.b32 %r15, %r14, %r13, 16, 8;
-; CHECK-NEXT: selp.b32 %r16, %r3, %r4, %p1;
-; CHECK-NEXT: bfi.b32 %r17, %r16, %r15, 24, 8;
-; CHECK-NEXT: st.param.b32 [func_retval0+0], %r17;
-; CHECK-NEXT: ret;
+; CHECK-NEXT: ld.param.u32 %r1, [test_umax_param_0];
+; CHECK-NEXT: ld.param.u32 %r2, [test_umax_param_1];
+; CHECK-NEXT: bfe.s32 %r3, %r2, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs1, %r3;
+; CHECK-NEXT: and.b16 %rs2, %rs1, 255;
+; CHECK-NEXT: bfe.s32 %r4, %r1, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs3, %r4;
+; CHECK-NEXT: and.b16 %rs4, %rs3, 255;
+; CHECK-NEXT: setp.gt.u16 %p1, %rs4, %rs2;
+; CHECK-NEXT: bfe.s32 %r5, %r2, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs5, %r5;
+; CHECK-NEXT: and.b16 %rs6, %rs5, 255;
+; CHECK-NEXT: bfe.s32 %r6, %r1, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs7, %r6;
+; CHECK-NEXT: and.b16 %rs8, %rs7, 255;
+; CHECK-NEXT: setp.gt.u16 %p2, %rs8, %rs6;
+; CHECK-NEXT: bfe.s32 %r7, %r2, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs9, %r7;
+; CHECK-NEXT: and.b16 %rs10, %rs9, 255;
+; CHECK-NEXT: bfe.s32 %r8, %r1, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs11, %r8;
+; CHECK-NEXT: and.b16 %rs12, %rs11, 255;
+; CHECK-NEXT: setp.gt.u16 %p3, %rs12, %rs10;
+; CHECK-NEXT: bfe.s32 %r9, %r2, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs13, %r9;
+; CHECK-NEXT: and.b16 %rs14, %rs13, 255;
+; CHECK-NEXT: bfe.s32 %r10, %r1, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs15, %r10;
+; CHECK-NEXT: and.b16 %rs16, %rs15, 255;
+; CHECK-NEXT: setp.gt.u16 %p4, %rs16, %rs14;
+; CHECK-NEXT: selp.b32 %r11, %r10, %r9, %p4;
+; CHECK-NEXT: selp.b32 %r12, %r8, %r7, %p3;
+; CHECK-NEXT: bfi.b32 %r13, %r12, %r11, 8, 8;
+; CHECK-NEXT: selp.b32 %r14, %r6, %r5, %p2;
+; CHECK-NEXT: bfi.b32 %r15, %r14, %r13, 16, 8;
+; CHECK-NEXT: selp.b32 %r16, %r4, %r3, %p1;
+; CHECK-NEXT: bfi.b32 %r17, %r16, %r15, 24, 8;
+; CHECK-NEXT: st.param.b32 [func_retval0+0], %r17;
+; CHECK-NEXT: ret;
+
%cmp = icmp ugt <4 x i8> %a, %b
%r = select <4 x i1> %cmp, <4 x i8> %a, <4 x i8> %b
ret <4 x i8> %r
@@ -355,32 +373,49 @@ define <4 x i8> @test_smin(<4 x i8> %a, <4 x i8> %b) #0 {
define <4 x i8> @test_umin(<4 x i8> %a, <4 x i8> %b) #0 {
; CHECK-LABEL: test_umin(
; CHECK: {
-; CHECK-NEXT: .reg .pred %p<5>;
-; CHECK-NEXT: .reg .b32 %r<19>;
+; CHECK-NEXT: .reg .pred %p<5>;
+; CHECK-NEXT: .reg .b16 %rs<17>;
+; CHECK-NEXT: .reg .b32 %r<19>;
; CHECK-EMPTY:
; CHECK-NEXT: // %bb.0:
; CHECK-NEXT: ld.param.u32 %r2, [test_umin_param_1];
; CHECK-NEXT: ld.param.u32 %r1, [test_umin_param_0];
-; CHECK-NEXT: bfe.s32 %r3, %r1, 24, 8;
-; CHECK-NEXT: bfe.s32 %r4, %r2, 24, 8;
-; CHECK-NEXT: setp.ls.u32 %p1, %r3, %r4;
-; CHECK-NEXT: bfe.s32 %r5, %r1, 16, 8;
-; CHECK-NEXT: bfe.s32 %r6, %r2, 16, 8;
-; CHECK-NEXT: setp.ls.u32 %p2, %r5, %r6;
-; CHECK-NEXT: bfe.s32 %r7, %r1, 8, 8;
-; CHECK-NEXT: bfe.s32 %r8, %r2, 8, 8;
-; CHECK-NEXT: setp.ls.u32 %p3, %r7, %r8;
-; CHECK-NEXT: bfe.s32 %r9, %r1, 0, 8;
-; CHECK-NEXT: bfe.s32 %r10, %r2, 0, 8;
-; CHECK-NEXT: setp.ls.u32 %p4, %r9, %r10;
-; CHECK-NEXT: selp.b32 %r11, %r9, %r10, %p4;
-; CHECK-NEXT: selp.b32 %r12, %r7, %r8, %p3;
-; CHECK-NEXT: bfi.b32 %r13, %r12, %r11, 8, 8;
-; CHECK-NEXT: selp.b32 %r14, %r5, %r6, %p2;
-; CHECK-NEXT: bfi.b32 %r15, %r14, %r13, 16, 8;
-; CHECK-NEXT: selp.b32 %r16, %r3, %r4, %p1;
-; CHECK-NEXT: bfi.b32 %r17, %r16, %r15, 24, 8;
-; CHECK-NEXT: st.param.b32 [func_retval0+0], %r17;
+; CHECK-NEXT: bfe.s32 %r3, %r2, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs1, %r3;
+; CHECK-NEXT: and.b16 %rs2, %rs1, 255;
+; CHECK-NEXT: bfe.s32 %r4, %r1, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs3, %r4;
+; CHECK-NEXT: and.b16 %rs4, %rs3, 255;
+; CHECK-NEXT: setp.le.u16 %p1, %rs4, %rs2;
+; CHECK-NEXT: bfe.s32 %r5, %r2, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs5, %r5;
+; CHECK-NEXT: and.b16 %rs6, %rs5, 255;
+; CHECK-NEXT: bfe.s32 %r6, %r1, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs7, %r6;
+; CHECK-NEXT: and.b16 %rs8, %rs7, 255;
+; CHECK-NEXT: setp.le.u16 %p2, %rs8, %rs6;
+; CHECK-NEXT: bfe.s32 %r7, %r2, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs9, %r7;
+; CHECK-NEXT: and.b16 %rs10, %rs9, 255;
+; CHECK-NEXT: bfe.s32 %r8, %r1, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs11, %r8;
+; CHECK-NEXT: and.b16 %rs12, %rs11, 255;
+; CHECK-NEXT: setp.le.u16 %p3, %rs12, %rs10;
+; CHECK-NEXT: bfe.s32 %r9, %r2, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs13, %r9;
+; CHECK-NEXT: and.b16 %rs14, %rs13, 255;
+; CHECK-NEXT: bfe.s32 %r10, %r1, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs15, %r10;
+; CHECK-NEXT: and.b16 %rs16, %rs15, 255;
+; CHECK-NEXT: setp.le.u16 %p4, %rs16, %rs14;
+; CHECK-NEXT: selp.b32 %r11, %r10, %r9, %p4;
+; CHECK-NEXT: selp.b32 %r12, %r8, %r7, %p3;
+; CHECK-NEXT: bfi.b32 %r13, %r12, %r11, 8, 8;
+; CHECK-NEXT: selp.b32 %r14, %r6, %r5, %p2;
+; CHECK-NEXT: bfi.b32 %r15, %r14, %r13, 16, 8;
+; CHECK-NEXT: selp.b32 %r16, %r4, %r3, %p1;
+; CHECK-NEXT: bfi.b32 %r17, %r16, %r15, 24, 8;
+; CHECK-NEXT: st.param.b32 [func_retval0+0], %r17;
; CHECK-NEXT: ret;
%cmp = icmp ule <4 x i8> %a, %b
%r = select <4 x i1> %cmp, <4 x i8> %a, <4 x i8> %b
@@ -390,37 +425,54 @@ define <4 x i8> @test_umin(<4 x i8> %a, <4 x i8> %b) #0 {
define <4 x i8> @test_eq(<4 x i8> %a, <4 x i8> %b, <4 x i8> %c) #0 {
; CHECK-LABEL: test_eq(
; CHECK: {
-; CHECK-NEXT: .reg .pred %p<5>;
-; CHECK-NEXT: .reg .b32 %r<24>;
+; CHECK-NEXT: .reg .pred %p<5>;
+; CHECK-NEXT: .reg .b16 %rs<17>;
+; CHECK-NEXT: .reg .b32 %r<24>;
; CHECK-EMPTY:
; CHECK-NEXT: // %bb.0:
; CHECK-NEXT: ld.param.u32 %r3, [test_eq_param_2];
; CHECK-NEXT: ld.param.u32 %r2, [test_eq_param_1];
; CHECK-NEXT: ld.param.u32 %r1, [test_eq_param_0];
-; CHECK-NEXT: bfe.s32 %r4, %r2, 24, 8;
-; CHECK-NEXT: bfe.s32 %r5, %r1, 24, 8;
-; CHECK-NEXT: setp.eq.u32 %p1, %r5, %r4;
-; CHECK-NEXT: bfe.s32 %r6, %r2, 16, 8;
-; CHECK-NEXT: bfe.s32 %r7, %r1, 16, 8;
-; CHECK-NEXT: setp.eq.u32 %p2, %r7, %r6;
-; CHECK-NEXT: bfe.s32 %r8, %r2, 8, 8;
-; CHECK-NEXT: bfe.s32 %r9, %r1, 8, 8;
-; CHECK-NEXT: setp.eq.u32 %p3, %r9, %r8;
-; CHECK-NEXT: bfe.s32 %r10, %r2, 0, 8;
-; CHECK-NEXT: bfe.s32 %r11, %r1, 0, 8;
-; CHECK-NEXT: setp.eq.u32 %p4, %r11, %r10;
-; CHECK-NEXT: bfe.s32 %r12, %r3, 0, 8;
-; CHECK-NEXT: selp.b32 %r13, %r11, %r12, %p4;
-; CHECK-NEXT: bfe.s32 %r14, %r3, 8, 8;
-; CHECK-NEXT: selp.b32 %r15, %r9, %r14, %p3;
-; CHECK-NEXT: bfi.b32 %r16, %r15, %r13, 8, 8;
-; CHECK-NEXT: bfe.s32 %r17, %r3, 16, 8;
-; CHECK-NEXT: selp.b32 %r18, %r7, %r17, %p2;
-; CHECK-NEXT: bfi.b32 %r19, %r18, %r16, 16, 8;
-; CHECK-NEXT: bfe.s32 %r20, %r3, 24, 8;
-; CHECK-NEXT: selp.b32 %r21, %r5, %r20, %p1;
-; CHECK-NEXT: bfi.b32 %r22, %r21, %r19, 24, 8;
-; CHECK-NEXT: st.param.b32 [func_retval0+0], %r22;
+; CHECK-NEXT: bfe.s32 %r4, %r2, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs1, %r4;
+; CHECK-NEXT: and.b16 %rs2, %rs1, 255;
+; CHECK-NEXT: bfe.s32 %r5, %r1, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs3, %r5;
+; CHECK-NEXT: and.b16 %rs4, %rs3, 255;
+; CHECK-NEXT: setp.eq.s16 %p1, %rs4, %rs2;
+; CHECK-NEXT: bfe.s32 %r6, %r2, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs5, %r6;
+; CHECK-NEXT: and.b16 %rs6, %rs5, 255;
+; CHECK-NEXT: bfe.s32 %r7, %r1, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs7, %r7;
+; CHECK-NEXT: and.b16 %rs8, %rs7, 255;
+; CHECK-NEXT: setp.eq.s16 %p2, %rs8, %rs6;
+; CHECK-NEXT: bfe.s32 %r8, %r2, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs9, %r8;
+; CHECK-NEXT: and.b16 %rs10, %rs9, 255;
+; CHECK-NEXT: bfe.s32 %r9, %r1, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs11, %r9;
+; CHECK-NEXT: and.b16 %rs12, %rs11, 255;
+; CHECK-NEXT: setp.eq.s16 %p3, %rs12, %rs10;
+; CHECK-NEXT: bfe.s32 %r10, %r2, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs13, %r10;
+; CHECK-NEXT: and.b16 %rs14, %rs13, 255;
+; CHECK-NEXT: bfe.s32 %r11, %r1, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs15, %r11;
+; CHECK-NEXT: and.b16 %rs16, %rs15, 255;
+; CHECK-NEXT: setp.eq.s16 %p4, %rs16, %rs14;
+; CHECK-NEXT: bfe.s32 %r12, %r3, 0, 8;
+; CHECK-NEXT: selp.b32 %r13, %r11, %r12, %p4;
+; CHECK-NEXT: bfe.s32 %r14, %r3, 8, 8;
+; CHECK-NEXT: selp.b32 %r15, %r9, %r14, %p3;
+; CHECK-NEXT: bfi.b32 %r16, %r15, %r13, 8, 8;
+; CHECK-NEXT: bfe.s32 %r17, %r3, 16, 8;
+; CHECK-NEXT: selp.b32 %r18, %r7, %r17, %p2;
+; CHECK-NEXT: bfi.b32 %r19, %r18, %r16, 16, 8;
+; CHECK-NEXT: bfe.s32 %r20, %r3, 24, 8;
+; CHECK-NEXT: selp.b32 %r21, %r5, %r20, %p1;
+; CHECK-NEXT: bfi.b32 %r22, %r21, %r19, 24, 8;
+; CHECK-NEXT: st.param.b32 [func_retval0+0], %r22;
; CHECK-NEXT: ret;
%cmp = icmp eq <4 x i8> %a, %b
%r = select <4 x i1> %cmp, <4 x i8> %a, <4 x i8> %c
@@ -430,37 +482,54 @@ define <4 x i8> @test_eq(<4 x i8> %a, <4 x i8> %b, <4 x i8> %c) #0 {
define <4 x i8> @test_ne(<4 x i8> %a, <4 x i8> %b, <4 x i8> %c) #0 {
; CHECK-LABEL: test_ne(
; CHECK: {
-; CHECK-NEXT: .reg .pred %p<5>;
-; CHECK-NEXT: .reg .b32 %r<24>;
+; CHECK-NEXT: .reg .pred %p<5>;
+; CHECK-NEXT: .reg .b16 %rs<17>;
+; CHECK-NEXT: .reg .b32 %r<24>;
; CHECK-EMPTY:
; CHECK-NEXT: // %bb.0:
; CHECK-NEXT: ld.param.u32 %r3, [test_ne_param_2];
; CHECK-NEXT: ld.param.u32 %r2, [test_ne_param_1];
; CHECK-NEXT: ld.param.u32 %r1, [test_ne_param_0];
-; CHECK-NEXT: bfe.s32 %r4, %r2, 24, 8;
-; CHECK-NEXT: bfe.s32 %r5, %r1, 24, 8;
-; CHECK-NEXT: setp.ne.u32 %p1, %r5, %r4;
-; CHECK-NEXT: bfe.s32 %r6, %r2, 16, 8;
-; CHECK-NEXT: bfe.s32 %r7, %r1, 16, 8;
-; CHECK-NEXT: setp.ne.u32 %p2, %r7, %r6;
-; CHECK-NEXT: bfe.s32 %r8, %r2, 8, 8;
-; CHECK-NEXT: bfe.s32 %r9, %r1, 8, 8;
-; CHECK-NEXT: setp.ne.u32 %p3, %r9, %r8;
-; CHECK-NEXT: bfe.s32 %r10, %r2, 0, 8;
-; CHECK-NEXT: bfe.s32 %r11, %r1, 0, 8;
-; CHECK-NEXT: setp.ne.u32 %p4, %r11, %r10;
-; CHECK-NEXT: bfe.s32 %r12, %r3, 0, 8;
-; CHECK-NEXT: selp.b32 %r13, %r11, %r12, %p4;
-; CHECK-NEXT: bfe.s32 %r14, %r3, 8, 8;
-; CHECK-NEXT: selp.b32 %r15, %r9, %r14, %p3;
-; CHECK-NEXT: bfi.b32 %r16, %r15, %r13, 8, 8;
-; CHECK-NEXT: bfe.s32 %r17, %r3, 16, 8;
-; CHECK-NEXT: selp.b32 %r18, %r7, %r17, %p2;
-; CHECK-NEXT: bfi.b32 %r19, %r18, %r16, 16, 8;
-; CHECK-NEXT: bfe.s32 %r20, %r3, 24, 8;
-; CHECK-NEXT: selp.b32 %r21, %r5, %r20, %p1;
-; CHECK-NEXT: bfi.b32 %r22, %r21, %r19, 24, 8;
-; CHECK-NEXT: st.param.b32 [func_retval0+0], %r22;
+; CHECK-NEXT: bfe.s32 %r4, %r2, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs1, %r4;
+; CHECK-NEXT: and.b16 %rs2, %rs1, 255;
+; CHECK-NEXT: bfe.s32 %r5, %r1, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs3, %r5;
+; CHECK-NEXT: and.b16 %rs4, %rs3, 255;
+; CHECK-NEXT: setp.ne.s16 %p1, %rs4, %rs2;
+; CHECK-NEXT: bfe.s32 %r6, %r2, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs5, %r6;
+; CHECK-NEXT: and.b16 %rs6, %rs5, 255;
+; CHECK-NEXT: bfe.s32 %r7, %r1, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs7, %r7;
+; CHECK-NEXT: and.b16 %rs8, %rs7, 255;
+; CHECK-NEXT: setp.ne.s16 %p2, %rs8, %rs6;
+; CHECK-NEXT: bfe.s32 %r8, %r2, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs9, %r8;
+; CHECK-NEXT: and.b16 %rs10, %rs9, 255;
+; CHECK-NEXT: bfe.s32 %r9, %r1, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs11, %r9;
+; CHECK-NEXT: and.b16 %rs12, %rs11, 255;
+; CHECK-NEXT: setp.ne.s16 %p3, %rs12, %rs10;
+; CHECK-NEXT: bfe.s32 %r10, %r2, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs13, %r10;
+; CHECK-NEXT: and.b16 %rs14, %rs13, 255;
+; CHECK-NEXT: bfe.s32 %r11, %r1, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs15, %r11;
+; CHECK-NEXT: and.b16 %rs16, %rs15, 255;
+; CHECK-NEXT: setp.ne.s16 %p4, %rs16, %rs14;
+; CHECK-NEXT: bfe.s32 %r12, %r3, 0, 8;
+; CHECK-NEXT: selp.b32 %r13, %r11, %r12, %p4;
+; CHECK-NEXT: bfe.s32 %r14, %r3, 8, 8;
+; CHECK-NEXT: selp.b32 %r15, %r9, %r14, %p3;
+; CHECK-NEXT: bfi.b32 %r16, %r15, %r13, 8, 8;
+; CHECK-NEXT: bfe.s32 %r17, %r3, 16, 8;
+; CHECK-NEXT: selp.b32 %r18, %r7, %r17, %p2;
+; CHECK-NEXT: bfi.b32 %r19, %r18, %r16, 16, 8;
+; CHECK-NEXT: bfe.s32 %r20, %r3, 24, 8;
+; CHECK-NEXT: selp.b32 %r21, %r5, %r20, %p1;
+; CHECK-NEXT: bfi.b32 %r22, %r21, %r19, 24, 8;
+; CHECK-NEXT: st.param.b32 [func_retval0+0], %r22;
; CHECK-NEXT: ret;
%cmp = icmp ne <4 x i8> %a, %b
%r = select <4 x i1> %cmp, <4 x i8> %a, <4 x i8> %c
@@ -912,42 +981,59 @@ define <4 x i8> @test_select(<4 x i8> %a, <4 x i8> %b, i1 zeroext %c) #0 {
define <4 x i8> @test_select_cc(<4 x i8> %a, <4 x i8> %b, <4 x i8> %c, <4 x i8> %d) #0 {
; CHECK-LABEL: test_select_cc(
; CHECK: {
-; CHECK-NEXT: .reg .pred %p<5>;
-; CHECK-NEXT: .reg .b32 %r<29>;
+; CHECK-NEXT: .reg .pred %p<5>;
+; CHECK-NEXT: .reg .b16 %rs<17>;
+; CHECK-NEXT: .reg .b32 %r<29>;
; CHECK-EMPTY:
; CHECK-NEXT: // %bb.0:
; CHECK-NEXT: ld.param.u32 %r4, [test_select_cc_param_3];
; CHECK-NEXT: ld.param.u32 %r3, [test_select_cc_param_2];
; CHECK-NEXT: ld.param.u32 %r2, [test_select_cc_param_1];
; CHECK-NEXT: ld.param.u32 %r1, [test_select_cc_param_0];
-; CHECK-NEXT: bfe.s32 %r5, %r4, 24, 8;
-; CHECK-NEXT: bfe.s32 %r6, %r3, 24, 8;
-; CHECK-NEXT: setp.ne.u32 %p1, %r6, %r5;
-; CHECK-NEXT: bfe.s32 %r7, %r4, 16, 8;
-; CHECK-NEXT: bfe.s32 %r8, %r3, 16, 8;
-; CHECK-NEXT: setp.ne.u32 %p2, %r8, %r7;
-; CHECK-NEXT: bfe.s32 %r9, %r4, 8, 8;
-; CHECK-NEXT: bfe.s32 %r10, %r3, 8, 8;
-; CHECK-NEXT: setp.ne.u32 %p3, %r10, %r9;
-; CHECK-NEXT: bfe.s32 %r11, %r4, 0, 8;
-; CHECK-NEXT: bfe.s32 %r12, %r3, 0, 8;
-; CHECK-NEXT: setp.ne.u32 %p4, %r12, %r11;
-; CHECK-NEXT: bfe.s32 %r13, %r2, 0, 8;
-; CHECK-NEXT: bfe.s32 %r14, %r1, 0, 8;
-; CHECK-NEXT: selp.b32 %r15, %r14, %r13, %p4;
-; CHECK-NEXT: bfe.s32 %r16, %r2, 8, 8;
-; CHECK-NEXT: bfe.s32 %r17, %r1, 8, 8;
-; CHECK-NEXT: selp.b32 %r18, %r17, %r16, %p3;
-; CHECK-NEXT: bfi.b32 %r19, %r18, %r15, 8, 8;
-; CHECK-NEXT: bfe.s32 %r20, %r2, 16, 8;
-; CHECK-NEXT: bfe.s32 %r21, %r1, 16, 8;
-; CHECK-NEXT: selp.b32 %r22, %r21, %r20, %p2;
-; CHECK-NEXT: bfi.b32 %r23, %r22, %r19, 16, 8;
-; CHECK-NEXT: bfe.s32 %r24, %r2, 24, 8;
-; CHECK-NEXT: bfe.s32 %r25, %r1, 24, 8;
-; CHECK-NEXT: selp.b32 %r26, %r25, %r24, %p1;
-; CHECK-NEXT: bfi.b32 %r27, %r26, %r23, 24, 8;
-; CHECK-NEXT: st.param.b32 [func_retval0+0], %r27;
+; CHECK-NEXT: bfe.s32 %r5, %r4, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs1, %r5;
+; CHECK-NEXT: and.b16 %rs2, %rs1, 255;
+; CHECK-NEXT: bfe.s32 %r6, %r3, 24, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs3, %r6;
+; CHECK-NEXT: and.b16 %rs4, %rs3, 255;
+; CHECK-NEXT: setp.ne.s16 %p1, %rs4, %rs2;
+; CHECK-NEXT: bfe.s32 %r7, %r4, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs5, %r7;
+; CHECK-NEXT: and.b16 %rs6, %rs5, 255;
+; CHECK-NEXT: bfe.s32 %r8, %r3, 16, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs7, %r8;
+; CHECK-NEXT: and.b16 %rs8, %rs7, 255;
+; CHECK-NEXT: setp.ne.s16 %p2, %rs8, %rs6;
+; CHECK-NEXT: bfe.s32 %r9, %r4, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs9, %r9;
+; CHECK-NEXT: and.b16 %rs10, %rs9, 255;
+; CHECK-NEXT: bfe.s32 %r10, %r3, 8, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs11, %r10;
+; CHECK-NEXT: and.b16 %rs12, %rs11, 255;
+; CHECK-NEXT: setp.ne.s16 %p3, %rs12, %rs10;
+; CHECK-NEXT: bfe.s32 %r11, %r4, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs13, %r11;
+; CHECK-NEXT: and.b16 %rs14, %rs13, 255;
+; CHECK-NEXT: bfe.s32 %r12, %r3, 0, 8;
+; CHECK-NEXT: cvt.u16.u32 %rs15, %r12;
+; CHECK-NEXT: and.b16 %rs16, %rs15, 255;
+; CHECK-NEXT: setp.ne.s16 %p4, %rs16, %rs14;
+; CHECK-NEXT: bfe.s32 %r13, %r2, 0, 8;
+; CHECK-NEXT: bfe.s32 %r14, %r1, 0, 8;
+; CHECK-NEXT: selp.b32 %r15, %r14, %r13, %p4;
+; CHECK-NEXT: bfe.s32 %r16, %r2, 8, 8;
+; CHECK-NEXT: bfe.s32 %r17, %r1, 8, 8;
+; CHECK-NEXT: selp.b32 %r18, %r17, %r16, %p3;
+; CHECK-NEXT: bfi.b32 %r19, %r18, %r15, 8, 8;
+; CHECK-NEXT: bfe.s32 %r20, %r2, 16, 8;
+; CHECK-NEXT: bfe.s32 %r21, %r1, 16, 8;
+; CHECK-NEXT: selp.b32 %r22, %r21, %r20, %p2;
+; CHECK-NEXT: bfi.b32 %r23, %r22, %r19, 16, 8;
+; CHECK-NEXT: bfe.s32 %r24, %r2, 24, 8;
+; CHECK-NEXT: b...
[truncated]
``````````
</details>
https://github.com/llvm/llvm-project/pull/80383
More information about the Mlir-commits
mailing list