Skip to content

Commit 79034e9

Browse files
justinfargnolijustinfargnoli
committed
Use isNarrowingProfitable instead of shouldReduceRegisterPressure
1 parent 8702dd1 commit 79034e9

File tree

8 files changed

+50
-58
lines changed

8 files changed

+50
-58
lines changed

‎llvm/include/llvm/CodeGen/TargetLowering.h‎

Lines changed: 0 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -2979,10 +2979,6 @@ class TargetLoweringBase {
29792979
return isTruncateFree(Val.getValueType(), VT2);
29802980
}
29812981

2982-
// Return true if the target will accepts tradeoffs (e.g. increase the number
2983-
// of instructions) to reduce register pressure.
2984-
virtual bool shouldReduceRegisterPressure() const { return false; }
2985-
29862982
virtual bool isProfitableToHoist(Instruction *I) const { return true; }
29872983

29882984
/// Return true if the extension represented by \p I is free.

‎llvm/lib/CodeGen/SelectionDAG/DAGCombiner.cpp‎

Lines changed: 9 additions & 9 deletions
Original file line numberDiff line numberDiff line change
@@ -5786,11 +5786,7 @@ SDValue DAGCombiner::hoistLogicOpWithSameOpcodeHands(SDNode *N) {
57865786
}
57875787

57885788
// logic_op (truncate x), (truncate y) --> truncate (logic_op x, y)
5789-
//
5790-
// For targets that are particulaly senesitve to register pressure (e.g. GPUs)
5791-
// it's preferable to increase the number of truncate instructions in order to
5792-
// decrease the bit width of the logic_op.
5793-
if (HandOpcode == ISD::TRUNCATE && !TLI.shouldReduceRegisterPressure()) {
5789+
if (HandOpcode == ISD::TRUNCATE) {
57945790
// If both operands have other uses, this transform would create extra
57955791
// instructions without eliminating anything.
57965792
if (!N0.hasOneUse() && !N1.hasOneUse())
@@ -5802,9 +5798,14 @@ SDValue DAGCombiner::hoistLogicOpWithSameOpcodeHands(SDNode *N) {
58025798
if (LegalOperations && !TLI.isOperationLegal(LogicOpcode, XVT))
58035799
return SDValue();
58045800
// Be extra careful sinking truncate. If it's free, there's no benefit in
5805-
// widening a binop. Also, don't create a logic op on an illegal type.
5801+
// widening a binop.
58065802
if (TLI.isZExtFree(VT, XVT) && TLI.isTruncateFree(XVT, VT))
58075803
return SDValue();
5804+
// Prevent an infinite loop if the target preferts the inverse
5805+
// transformation.
5806+
if (TLI.isNarrowingProfitable(XVT, VT))
5807+
return SDValue();
5808+
// Don't create a logic op on an illegal type.
58085809
if (!TLI.isTypeLegal(XVT))
58095810
return SDValue();
58105811
SDValue Logic = DAG.getNode(LogicOpcode, DL, XVT, X, Y);
@@ -15368,16 +15369,15 @@ SDValue DAGCombiner::visitTRUNCATE(SDNode *N) {
1536815369
break;
1536915370
}
1537015371

15371-
if (TLI.shouldReduceRegisterPressure()) {
15372+
if (TLI.isNarrowingProfitable(SrcVT, VT)) {
1537215373
switch (N0.getOpcode()) {
1537315374
case ISD::ADD:
1537415375
case ISD::SUB:
1537515376
case ISD::MUL:
1537615377
case ISD::AND:
1537715378
case ISD::OR:
1537815379
case ISD::XOR:
15379-
if (!(N0.hasOneUse() && VT.isScalarInteger() &&
15380-
TLI.isTruncateFree(SrcVT, VT)))
15380+
if (!(N0.hasOneUse() && VT.isScalarInteger()))
1538115381
break;
1538215382
if (LegalOperations && !TLI.isOperationLegal(N0.getOpcode(), VT))
1538315383
break;

‎llvm/lib/Target/NVPTX/NVPTXISelLowering.h‎

Lines changed: 3 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -505,15 +505,13 @@ class NVPTXTargetLowering : public TargetLowering {
505505
DstTy->getPrimitiveSizeInBits() == 32;
506506
}
507507

508-
bool isTruncateFree(EVT FromVT, EVT ToVT) const override {
508+
bool isNarrowingProfitable(EVT SrcVT, EVT DestVT) const override {
509509
// Truncating 64-bit to 32-bit is free in SASS.
510-
if (!(FromVT.isScalarInteger() && ToVT.isScalarInteger()))
510+
if (!(SrcVT.isScalarInteger() && DestVT.isScalarInteger()))
511511
return false;
512-
return FromVT.getFixedSizeInBits() == 64 && ToVT.getFixedSizeInBits() == 32;
512+
return SrcVT.getFixedSizeInBits() == 64 && DestVT.getFixedSizeInBits() == 32;
513513
}
514514

515-
bool shouldReduceRegisterPressure() const override { return true; }
516-
517515
EVT getSetCCResultType(const DataLayout &DL, LLVMContext &Ctx,
518516
EVT VT) const override {
519517
if (VT.isVector())

‎llvm/test/CodeGen/NVPTX/boolean-patterns.ll‎

Lines changed: 16 additions & 24 deletions
Original file line numberDiff line numberDiff line change
@@ -43,20 +43,16 @@ define i1 @m2and_ri(i1 %a) {
4343
define i1 @select2or(i1 %a, i1 %b) {
4444
; CHECK-LABEL: select2or(
4545
; CHECK: {
46-
; CHECK-NEXT: .reg .pred %p<4>;
47-
; CHECK-NEXT: .reg .b16 %rs<5>;
48-
; CHECK-NEXT: .reg .b32 %r<2>;
46+
; CHECK-NEXT: .reg .b16 %rs<4>;
47+
; CHECK-NEXT: .reg .b32 %r<3>;
4948
; CHECK-EMPTY:
5049
; CHECK-NEXT: // %bb.0:
51-
; CHECK-NEXT: ld.param.u8 %rs1, [select2or_param_1];
52-
; CHECK-NEXT: and.b16 %rs2, %rs1, 1;
53-
; CHECK-NEXT: setp.eq.b16 %p1, %rs2, 1;
54-
; CHECK-NEXT: ld.param.u8 %rs3, [select2or_param_0];
55-
; CHECK-NEXT: and.b16 %rs4, %rs3, 1;
56-
; CHECK-NEXT: setp.eq.b16 %p2, %rs4, 1;
57-
; CHECK-NEXT: or.pred %p3, %p2, %p1;
58-
; CHECK-NEXT: selp.u32 %r1, 1, 0, %p3;
59-
; CHECK-NEXT: st.param.b32 [func_retval0+0], %r1;
50+
; CHECK-NEXT: ld.param.u8 %rs1, [select2or_param_0];
51+
; CHECK-NEXT: ld.param.u8 %rs2, [select2or_param_1];
52+
; CHECK-NEXT: or.b16 %rs3, %rs1, %rs2;
53+
; CHECK-NEXT: cvt.u32.u16 %r1, %rs3;
54+
; CHECK-NEXT: and.b32 %r2, %r1, 1;
55+
; CHECK-NEXT: st.param.b32 [func_retval0+0], %r2;
6056
; CHECK-NEXT: ret;
6157
%r = select i1 %a, i1 1, i1 %b
6258
ret i1 %r
@@ -65,20 +61,16 @@ define i1 @select2or(i1 %a, i1 %b) {
6561
define i1 @select2and(i1 %a, i1 %b) {
6662
; CHECK-LABEL: select2and(
6763
; CHECK: {
68-
; CHECK-NEXT: .reg .pred %p<4>;
69-
; CHECK-NEXT: .reg .b16 %rs<5>;
70-
; CHECK-NEXT: .reg .b32 %r<2>;
64+
; CHECK-NEXT: .reg .b16 %rs<4>;
65+
; CHECK-NEXT: .reg .b32 %r<3>;
7166
; CHECK-EMPTY:
7267
; CHECK-NEXT: // %bb.0:
73-
; CHECK-NEXT: ld.param.u8 %rs1, [select2and_param_1];
74-
; CHECK-NEXT: and.b16 %rs2, %rs1, 1;
75-
; CHECK-NEXT: setp.eq.b16 %p1, %rs2, 1;
76-
; CHECK-NEXT: ld.param.u8 %rs3, [select2and_param_0];
77-
; CHECK-NEXT: and.b16 %rs4, %rs3, 1;
78-
; CHECK-NEXT: setp.eq.b16 %p2, %rs4, 1;
79-
; CHECK-NEXT: and.pred %p3, %p2, %p1;
80-
; CHECK-NEXT: selp.u32 %r1, 1, 0, %p3;
81-
; CHECK-NEXT: st.param.b32 [func_retval0+0], %r1;
68+
; CHECK-NEXT: ld.param.u8 %rs1, [select2and_param_0];
69+
; CHECK-NEXT: ld.param.u8 %rs2, [select2and_param_1];
70+
; CHECK-NEXT: and.b16 %rs3, %rs1, %rs2;
71+
; CHECK-NEXT: cvt.u32.u16 %r1, %rs3;
72+
; CHECK-NEXT: and.b32 %r2, %r1, 1;
73+
; CHECK-NEXT: st.param.b32 [func_retval0+0], %r2;
8274
; CHECK-NEXT: ret;
8375
%r = select i1 %a, i1 %b, i1 0
8476
ret i1 %r

‎llvm/test/CodeGen/X86/apx/and.ll‎

Lines changed: 2 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -5,14 +5,12 @@
55
define i8 @and8rr(i8 noundef %a, i8 noundef %b) {
66
; CHECK-LABEL: and8rr:
77
; CHECK: # %bb.0: # %entry
8-
; CHECK-NEXT: andl %esi, %edi, %eax # encoding: [0x62,0xf4,0x7c,0x18,0x21,0xf7]
9-
; CHECK-NEXT: # kill: def $al killed $al killed $eax
8+
; CHECK-NEXT: andb %sil, %dil, %al # encoding: [0x62,0xf4,0x7c,0x18,0x20,0xf7]
109
; CHECK-NEXT: retq # encoding: [0xc3]
1110
;
1211
; NF-LABEL: and8rr:
1312
; NF: # %bb.0: # %entry
14-
; NF-NEXT: {nf} andl %esi, %edi, %eax # EVEX TO EVEX Compression encoding: [0x62,0xf4,0x7c,0x1c,0x21,0xf7]
15-
; NF-NEXT: # kill: def $al killed $al killed $eax
13+
; NF-NEXT: {nf} andb %sil, %dil, %al # EVEX TO EVEX Compression encoding: [0x62,0xf4,0x7c,0x1c,0x20,0xf7]
1614
; NF-NEXT: retq # encoding: [0xc3]
1715
entry:
1816
%and = and i8 %a, %b

‎llvm/test/CodeGen/X86/unfold-masked-merge-scalar-constmask-interleavedbits.ll‎

Lines changed: 8 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -114,19 +114,19 @@ define i64 @out64_constmask(i64 %x, i64 %y) {
114114
define i8 @in8_constmask(i8 %x, i8 %y) {
115115
; CHECK-NOBMI-LABEL: in8_constmask:
116116
; CHECK-NOBMI: # %bb.0:
117-
; CHECK-NOBMI-NEXT: movl %esi, %eax
118-
; CHECK-NOBMI-NEXT: xorl %esi, %edi
119-
; CHECK-NOBMI-NEXT: andb $85, %dil
120-
; CHECK-NOBMI-NEXT: xorb %dil, %al
117+
; CHECK-NOBMI-NEXT: movl %edi, %eax
118+
; CHECK-NOBMI-NEXT: xorb %sil, %al
119+
; CHECK-NOBMI-NEXT: andb $85, %al
120+
; CHECK-NOBMI-NEXT: xorb %sil, %al
121121
; CHECK-NOBMI-NEXT: # kill: def $al killed $al killed $eax
122122
; CHECK-NOBMI-NEXT: retq
123123
;
124124
; CHECK-BMI-LABEL: in8_constmask:
125125
; CHECK-BMI: # %bb.0:
126-
; CHECK-BMI-NEXT: movl %esi, %eax
127-
; CHECK-BMI-NEXT: xorl %esi, %edi
128-
; CHECK-BMI-NEXT: andb $85, %dil
129-
; CHECK-BMI-NEXT: xorb %dil, %al
126+
; CHECK-BMI-NEXT: movl %edi, %eax
127+
; CHECK-BMI-NEXT: xorb %sil, %al
128+
; CHECK-BMI-NEXT: andb $85, %al
129+
; CHECK-BMI-NEXT: xorb %sil, %al
130130
; CHECK-BMI-NEXT: # kill: def $al killed $al killed $eax
131131
; CHECK-BMI-NEXT: retq
132132
%n0 = xor i8 %x, %y

‎llvm/test/CodeGen/X86/vec_saddo.ll‎

Lines changed: 6 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1045,12 +1045,16 @@ define <4 x i32> @saddo_v4i1(<4 x i1> %a0, <4 x i1> %a1, ptr %p2) nounwind {
10451045
;
10461046
; AVX512-LABEL: saddo_v4i1:
10471047
; AVX512: # %bb.0:
1048-
; AVX512-NEXT: vpxor %xmm1, %xmm0, %xmm2
1049-
; AVX512-NEXT: vpslld $31, %xmm2, %xmm2
1048+
; AVX512-NEXT: vpslld $31, %xmm0, %xmm2
10501049
; AVX512-NEXT: vptestmd %xmm2, %xmm2, %k0
1050+
; AVX512-NEXT: vpslld $31, %xmm1, %xmm2
1051+
; AVX512-NEXT: vptestmd %xmm2, %xmm2, %k1
1052+
; AVX512-NEXT: kxorw %k1, %k0, %k0
10511053
; AVX512-NEXT: vpand %xmm1, %xmm0, %xmm0
10521054
; AVX512-NEXT: vpslld $31, %xmm0, %xmm0
10531055
; AVX512-NEXT: vpsrad $31, %xmm0, %xmm0
1056+
; AVX512-NEXT: kshiftlw $12, %k0, %k0
1057+
; AVX512-NEXT: kshiftrw $12, %k0, %k0
10541058
; AVX512-NEXT: kmovd %k0, %eax
10551059
; AVX512-NEXT: movb %al, (%rdi)
10561060
; AVX512-NEXT: retq

‎llvm/test/CodeGen/X86/vec_uaddo.ll‎

Lines changed: 6 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1098,12 +1098,16 @@ define <4 x i32> @uaddo_v4i1(<4 x i1> %a0, <4 x i1> %a1, ptr %p2) nounwind {
10981098
;
10991099
; AVX512-LABEL: uaddo_v4i1:
11001100
; AVX512: # %bb.0:
1101-
; AVX512-NEXT: vpxor %xmm1, %xmm0, %xmm2
1102-
; AVX512-NEXT: vpslld $31, %xmm2, %xmm2
1101+
; AVX512-NEXT: vpslld $31, %xmm0, %xmm2
11031102
; AVX512-NEXT: vptestmd %xmm2, %xmm2, %k0
1103+
; AVX512-NEXT: vpslld $31, %xmm1, %xmm2
1104+
; AVX512-NEXT: vptestmd %xmm2, %xmm2, %k1
1105+
; AVX512-NEXT: kxorw %k1, %k0, %k0
11041106
; AVX512-NEXT: vpand %xmm1, %xmm0, %xmm0
11051107
; AVX512-NEXT: vpslld $31, %xmm0, %xmm0
11061108
; AVX512-NEXT: vpsrad $31, %xmm0, %xmm0
1109+
; AVX512-NEXT: kshiftlw $12, %k0, %k0
1110+
; AVX512-NEXT: kshiftrw $12, %k0, %k0
11071111
; AVX512-NEXT: kmovd %k0, %eax
11081112
; AVX512-NEXT: movb %al, (%rdi)
11091113
; AVX512-NEXT: retq

0 commit comments

Comments
 (0)