[llvm] [clang] Reapply "InstCombine: Introduce SimplifyDemandedUseFPClass"" (PR #74056)
via cfe-commits
cfe-commits at lists.llvm.org
Fri Dec 1 01:29:11 PST 2023
llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT-->
@llvm/pr-subscribers-clang
@llvm/pr-subscribers-llvm-transforms
Author: Matt Arsenault (arsenm)
<details>
<summary>Changes</summary>
This reverts commit ef388334ee5a3584255b9ef5b3fefdb244fa3fd7.
The referenced issue violates the spec for finite-only math only by
using a return value for a constant infinity. If the interpretation
is results and arguments cannot violate nofpclass, then any
std::numeric_limits<T>::infinity() result is invalid under
-ffinite-math-only. Without this interpretation the utility of
nofpclass is slashed.
---
Patch is 53.86 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/74056.diff
6 Files Affected:
- (modified) clang/test/Headers/__clang_hip_math.hip (+44-12)
- (modified) llvm/include/llvm/Analysis/ValueTracking.h (+4)
- (modified) llvm/lib/Transforms/InstCombine/InstCombineInternal.h (+9)
- (modified) llvm/lib/Transforms/InstCombine/InstCombineSimplifyDemanded.cpp (+139-1)
- (modified) llvm/lib/Transforms/InstCombine/InstructionCombining.cpp (+16-2)
- (modified) llvm/test/Transforms/InstCombine/simplify-demanded-fpclass.ll (+74-129)
``````````diff
diff --git a/clang/test/Headers/__clang_hip_math.hip b/clang/test/Headers/__clang_hip_math.hip
index c0f4a06acbb8e32..9e15aec94dc28ab 100644
--- a/clang/test/Headers/__clang_hip_math.hip
+++ b/clang/test/Headers/__clang_hip_math.hip
@@ -2557,33 +2557,65 @@ extern "C" __device__ double test_nan(const char *tag) {
return nan(tag);
}
-// CHECK-LABEL: @test_nanf_emptystr(
-// CHECK-NEXT: entry:
-// CHECK-NEXT: ret float 0x7FF8000000000000
+// DEFAULT-LABEL: @test_nanf_emptystr(
+// DEFAULT-NEXT: entry:
+// DEFAULT-NEXT: ret float 0x7FF8000000000000
+//
+// FINITEONLY-LABEL: @test_nanf_emptystr(
+// FINITEONLY-NEXT: entry:
+// FINITEONLY-NEXT: ret float poison
+//
+// APPROX-LABEL: @test_nanf_emptystr(
+// APPROX-NEXT: entry:
+// APPROX-NEXT: ret float 0x7FF8000000000000
//
extern "C" __device__ float test_nanf_emptystr() {
return nanf("");
}
-// CHECK-LABEL: @test_nan_emptystr(
-// CHECK-NEXT: entry:
-// CHECK-NEXT: ret double 0x7FF8000000000000
+// DEFAULT-LABEL: @test_nan_emptystr(
+// DEFAULT-NEXT: entry:
+// DEFAULT-NEXT: ret double 0x7FF8000000000000
+//
+// FINITEONLY-LABEL: @test_nan_emptystr(
+// FINITEONLY-NEXT: entry:
+// FINITEONLY-NEXT: ret double poison
+//
+// APPROX-LABEL: @test_nan_emptystr(
+// APPROX-NEXT: entry:
+// APPROX-NEXT: ret double 0x7FF8000000000000
//
extern "C" __device__ double test_nan_emptystr() {
return nan("");
}
-// CHECK-LABEL: @test_nanf_fill(
-// CHECK-NEXT: entry:
-// CHECK-NEXT: ret float 0x7FF8000000000000
+// DEFAULT-LABEL: @test_nanf_fill(
+// DEFAULT-NEXT: entry:
+// DEFAULT-NEXT: ret float 0x7FF8000000000000
+//
+// FINITEONLY-LABEL: @test_nanf_fill(
+// FINITEONLY-NEXT: entry:
+// FINITEONLY-NEXT: ret float poison
+//
+// APPROX-LABEL: @test_nanf_fill(
+// APPROX-NEXT: entry:
+// APPROX-NEXT: ret float 0x7FF8000000000000
//
extern "C" __device__ float test_nanf_fill() {
return nanf("0x456");
}
-// CHECK-LABEL: @test_nan_fill(
-// CHECK-NEXT: entry:
-// CHECK-NEXT: ret double 0x7FF8000000000000
+// DEFAULT-LABEL: @test_nan_fill(
+// DEFAULT-NEXT: entry:
+// DEFAULT-NEXT: ret double 0x7FF8000000000000
+//
+// FINITEONLY-LABEL: @test_nan_fill(
+// FINITEONLY-NEXT: entry:
+// FINITEONLY-NEXT: ret double poison
+//
+// APPROX-LABEL: @test_nan_fill(
+// APPROX-NEXT: entry:
+// APPROX-NEXT: ret double 0x7FF8000000000000
//
extern "C" __device__ double test_nan_fill() {
return nan("0x123");
diff --git a/llvm/include/llvm/Analysis/ValueTracking.h b/llvm/include/llvm/Analysis/ValueTracking.h
index 82c87edd6297cdf..f9ce679bc74268f 100644
--- a/llvm/include/llvm/Analysis/ValueTracking.h
+++ b/llvm/include/llvm/Analysis/ValueTracking.h
@@ -243,6 +243,10 @@ struct KnownFPClass {
/// definitely set or false if the sign bit is definitely unset.
std::optional<bool> SignBit;
+ bool operator==(KnownFPClass Other) const {
+ return KnownFPClasses == Other.KnownFPClasses && SignBit == Other.SignBit;
+ }
+
/// Return true if it's known this can never be one of the mask entries.
bool isKnownNever(FPClassTest Mask) const {
return (KnownFPClasses & Mask) == fcNone;
diff --git a/llvm/lib/Transforms/InstCombine/InstCombineInternal.h b/llvm/lib/Transforms/InstCombine/InstCombineInternal.h
index 0bbb22be71569f6..9a66fb8f456f95b 100644
--- a/llvm/lib/Transforms/InstCombine/InstCombineInternal.h
+++ b/llvm/lib/Transforms/InstCombine/InstCombineInternal.h
@@ -551,6 +551,15 @@ class LLVM_LIBRARY_VISIBILITY InstCombinerImpl final
APInt &UndefElts, unsigned Depth = 0,
bool AllowMultipleUsers = false) override;
+ /// Attempts to replace V with a simpler value based on the demanded
+ /// floating-point classes
+ Value *SimplifyDemandedUseFPClass(Value *V, FPClassTest DemandedMask,
+ KnownFPClass &Known, unsigned Depth,
+ Instruction *CxtI);
+ bool SimplifyDemandedFPClass(Instruction *I, unsigned Op,
+ FPClassTest DemandedMask, KnownFPClass &Known,
+ unsigned Depth = 0);
+
/// Canonicalize the position of binops relative to shufflevector.
Instruction *foldVectorBinop(BinaryOperator &Inst);
Instruction *foldVectorSelect(SelectInst &Sel);
diff --git a/llvm/lib/Transforms/InstCombine/InstCombineSimplifyDemanded.cpp b/llvm/lib/Transforms/InstCombine/InstCombineSimplifyDemanded.cpp
index 4d19bd12d8f6f47..418155e90a2a7ed 100644
--- a/llvm/lib/Transforms/InstCombine/InstCombineSimplifyDemanded.cpp
+++ b/llvm/lib/Transforms/InstCombine/InstCombineSimplifyDemanded.cpp
@@ -466,7 +466,8 @@ Value *InstCombinerImpl::SimplifyDemandedUseBits(Value *V, APInt DemandedMask,
if (InputKnown.isNonNegative() ||
DemandedMask.getActiveBits() <= SrcBitWidth) {
// Convert to ZExt cast.
- CastInst *NewCast = new ZExtInst(I->getOperand(0), VTy, I->getName());
+ CastInst *NewCast = new ZExtInst(I->getOperand(0), VTy);
+ NewCast->takeName(I);
return InsertNewInstWith(NewCast, I->getIterator());
}
@@ -774,6 +775,7 @@ Value *InstCombinerImpl::SimplifyDemandedUseBits(Value *V, APInt DemandedMask,
BinaryOperator *LShr = BinaryOperator::CreateLShr(I->getOperand(0),
I->getOperand(1));
LShr->setIsExact(cast<BinaryOperator>(I)->isExact());
+ LShr->takeName(I);
return InsertNewInstWith(LShr, I->getIterator());
} else if (Known.One[BitWidth-ShiftAmt-1]) { // New bits are known one.
Known.One |= HighBits;
@@ -1849,3 +1851,139 @@ Value *InstCombinerImpl::SimplifyDemandedVectorElts(Value *V,
return MadeChange ? I : nullptr;
}
+
+/// For floating-point classes that resolve to a single bit pattern, return that
+/// value.
+static Constant *getFPClassConstant(Type *Ty, FPClassTest Mask) {
+ switch (Mask) {
+ case fcPosZero:
+ return ConstantFP::getZero(Ty);
+ case fcNegZero:
+ return ConstantFP::getZero(Ty, true);
+ case fcPosInf:
+ return ConstantFP::getInfinity(Ty);
+ case fcNegInf:
+ return ConstantFP::getInfinity(Ty, true);
+ case fcNone:
+ return PoisonValue::get(Ty);
+ default:
+ return nullptr;
+ }
+}
+
+Value *InstCombinerImpl::SimplifyDemandedUseFPClass(
+ Value *V, const FPClassTest DemandedMask, KnownFPClass &Known,
+ unsigned Depth, Instruction *CxtI) {
+ assert(Depth <= MaxAnalysisRecursionDepth && "Limit Search Depth");
+ Type *VTy = V->getType();
+
+ assert(Known == KnownFPClass() && "expected uninitialized state");
+
+ if (DemandedMask == fcNone)
+ return isa<UndefValue>(V) ? nullptr : PoisonValue::get(VTy);
+
+ if (Depth == MaxAnalysisRecursionDepth)
+ return nullptr;
+
+ Instruction *I = dyn_cast<Instruction>(V);
+ if (!I) {
+ // Handle constants and arguments
+ Known = computeKnownFPClass(V, fcAllFlags, CxtI, Depth + 1);
+ Value *FoldedToConst =
+ getFPClassConstant(VTy, DemandedMask & Known.KnownFPClasses);
+ return FoldedToConst == V ? nullptr : FoldedToConst;
+ }
+
+ if (!I->hasOneUse())
+ return nullptr;
+
+ // TODO: Should account for nofpclass/FastMathFlags on current instruction
+ switch (I->getOpcode()) {
+ case Instruction::FNeg: {
+ if (SimplifyDemandedFPClass(I, 0, llvm::fneg(DemandedMask), Known,
+ Depth + 1))
+ return I;
+ Known.fneg();
+ break;
+ }
+ case Instruction::Call: {
+ CallInst *CI = cast<CallInst>(I);
+ switch (CI->getIntrinsicID()) {
+ case Intrinsic::fabs:
+ if (SimplifyDemandedFPClass(I, 0, llvm::inverse_fabs(DemandedMask), Known,
+ Depth + 1))
+ return I;
+ Known.fabs();
+ break;
+ case Intrinsic::arithmetic_fence:
+ if (SimplifyDemandedFPClass(I, 0, DemandedMask, Known, Depth + 1))
+ return I;
+ break;
+ case Intrinsic::copysign: {
+ // Flip on more potentially demanded classes
+ const FPClassTest DemandedMaskAnySign = llvm::unknown_sign(DemandedMask);
+ if (SimplifyDemandedFPClass(I, 0, DemandedMaskAnySign, Known, Depth + 1))
+ return I;
+
+ if ((DemandedMask & fcPositive) == fcNone) {
+ // Roundabout way of replacing with fneg(fabs)
+ I->setOperand(1, ConstantFP::get(VTy, -1.0));
+ return I;
+ }
+
+ if ((DemandedMask & fcNegative) == fcNone) {
+ // Roundabout way of replacing with fabs
+ I->setOperand(1, ConstantFP::getZero(VTy));
+ return I;
+ }
+
+ KnownFPClass KnownSign =
+ computeKnownFPClass(I->getOperand(1), fcAllFlags, CxtI, Depth + 1);
+ Known.copysign(KnownSign);
+ break;
+ }
+ default:
+ Known = computeKnownFPClass(I, ~DemandedMask, CxtI, Depth + 1);
+ break;
+ }
+
+ break;
+ }
+ case Instruction::Select: {
+ KnownFPClass KnownLHS, KnownRHS;
+ if (SimplifyDemandedFPClass(I, 2, DemandedMask, KnownRHS, Depth + 1) ||
+ SimplifyDemandedFPClass(I, 1, DemandedMask, KnownLHS, Depth + 1))
+ return I;
+
+ if (KnownLHS.isKnownNever(DemandedMask))
+ return I->getOperand(2);
+ if (KnownRHS.isKnownNever(DemandedMask))
+ return I->getOperand(1);
+
+ // TODO: Recognize clamping patterns
+ Known = KnownLHS | KnownRHS;
+ break;
+ }
+ default:
+ Known = computeKnownFPClass(I, ~DemandedMask, CxtI, Depth + 1);
+ break;
+ }
+
+ return getFPClassConstant(VTy, DemandedMask & Known.KnownFPClasses);
+}
+
+bool InstCombinerImpl::SimplifyDemandedFPClass(Instruction *I, unsigned OpNo,
+ FPClassTest DemandedMask,
+ KnownFPClass &Known,
+ unsigned Depth) {
+ Use &U = I->getOperandUse(OpNo);
+ Value *NewVal =
+ SimplifyDemandedUseFPClass(U.get(), DemandedMask, Known, Depth, I);
+ if (!NewVal)
+ return false;
+ if (Instruction *OpInst = dyn_cast<Instruction>(U))
+ salvageDebugInfo(*OpInst);
+
+ replaceUse(U, NewVal);
+ return true;
+}
diff --git a/llvm/lib/Transforms/InstCombine/InstructionCombining.cpp b/llvm/lib/Transforms/InstCombine/InstructionCombining.cpp
index 26fdef672506a68..4f2f4edea6e09d3 100644
--- a/llvm/lib/Transforms/InstCombine/InstructionCombining.cpp
+++ b/llvm/lib/Transforms/InstCombine/InstructionCombining.cpp
@@ -2904,8 +2904,22 @@ Instruction *InstCombinerImpl::visitFree(CallInst &FI, Value *Op) {
}
Instruction *InstCombinerImpl::visitReturnInst(ReturnInst &RI) {
- // Nothing for now.
- return nullptr;
+ Value *RetVal = RI.getReturnValue();
+ if (!RetVal || !AttributeFuncs::isNoFPClassCompatibleType(RetVal->getType()))
+ return nullptr;
+
+ Function *F = RI.getFunction();
+ FPClassTest ReturnClass = F->getAttributes().getRetNoFPClass();
+ if (ReturnClass == fcNone)
+ return nullptr;
+
+ KnownFPClass KnownClass;
+ Value *Simplified =
+ SimplifyDemandedUseFPClass(RetVal, ~ReturnClass, KnownClass, 0, &RI);
+ if (!Simplified)
+ return nullptr;
+
+ return ReturnInst::Create(RI.getContext(), Simplified);
}
// WARNING: keep in sync with SimplifyCFGOpt::simplifyUnreachable()!
diff --git a/llvm/test/Transforms/InstCombine/simplify-demanded-fpclass.ll b/llvm/test/Transforms/InstCombine/simplify-demanded-fpclass.ll
index 9817b6e13ca8ae9..4f9396add2370b4 100644
--- a/llvm/test/Transforms/InstCombine/simplify-demanded-fpclass.ll
+++ b/llvm/test/Transforms/InstCombine/simplify-demanded-fpclass.ll
@@ -42,7 +42,7 @@ define nofpclass(inf) float @ret_nofpclass_inf_undef() {
define nofpclass(all) float @ret_nofpclass_all_var(float %arg) {
; CHECK-LABEL: define nofpclass(all) float @ret_nofpclass_all_var
; CHECK-SAME: (float [[ARG:%.*]]) {
-; CHECK-NEXT: ret float [[ARG]]
+; CHECK-NEXT: ret float poison
;
ret float %arg
}
@@ -51,7 +51,7 @@ define nofpclass(all) float @ret_nofpclass_all_var(float %arg) {
define nofpclass(all) <2 x float> @ret_nofpclass_all_var_vector(<2 x float> %arg) {
; CHECK-LABEL: define nofpclass(all) <2 x float> @ret_nofpclass_all_var_vector
; CHECK-SAME: (<2 x float> [[ARG:%.*]]) {
-; CHECK-NEXT: ret <2 x float> [[ARG]]
+; CHECK-NEXT: ret <2 x float> poison
;
ret <2 x float> %arg
}
@@ -65,14 +65,14 @@ define nofpclass(inf) float @ret_nofpclass_inf__0() {
define nofpclass(inf) float @ret_nofpclass_inf__pinf() {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__pinf() {
-; CHECK-NEXT: ret float 0x7FF0000000000000
+; CHECK-NEXT: ret float poison
;
ret float 0x7FF0000000000000
}
define nofpclass(pinf) float @ret_nofpclass_pinf__pinf() {
; CHECK-LABEL: define nofpclass(pinf) float @ret_nofpclass_pinf__pinf() {
-; CHECK-NEXT: ret float 0x7FF0000000000000
+; CHECK-NEXT: ret float poison
;
ret float 0x7FF0000000000000
}
@@ -86,7 +86,7 @@ define nofpclass(pinf) float @ret_nofpclass_pinf__ninf() {
define nofpclass(inf) float @ret_nofpclass_inf__ninf() {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__ninf() {
-; CHECK-NEXT: ret float 0xFFF0000000000000
+; CHECK-NEXT: ret float poison
;
ret float 0xFFF0000000000000
}
@@ -106,8 +106,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_inf_lhs(i1 %con
define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_arg_only_inf_lhs(i1 %cond, float nofpclass(nan norm zero sub) %x, float %y) {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_arg_only_inf_lhs
; CHECK-SAME: (i1 [[COND:%.*]], float nofpclass(nan zero sub norm) [[X:%.*]], float [[Y:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float [[X]], float [[Y]]
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float [[Y]]
;
%select = select i1 %cond, float %x, float %y
ret float %select
@@ -117,8 +116,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_arg_only_inf_lh
define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_arg_only_inf_rhs(i1 %cond, float %x, float nofpclass(nan norm zero sub) %y) {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_arg_only_inf_rhs
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]], float nofpclass(nan zero sub norm) [[Y:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float [[X]], float [[Y]]
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float [[X]]
;
%select = select i1 %cond, float %x, float %y
ret float %select
@@ -128,8 +126,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_nofpclass_arg_only_inf_rh
define nofpclass(inf) [3 x [2 x float]] @ret_float_array(i1 %cond, [3 x [2 x float]] nofpclass(nan norm zero sub) %x, [3 x [2 x float]] %y) {
; CHECK-LABEL: define nofpclass(inf) [3 x [2 x float]] @ret_float_array
; CHECK-SAME: (i1 [[COND:%.*]], [3 x [2 x float]] nofpclass(nan zero sub norm) [[X:%.*]], [3 x [2 x float]] [[Y:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], [3 x [2 x float]] [[X]], [3 x [2 x float]] [[Y]]
-; CHECK-NEXT: ret [3 x [2 x float]] [[SELECT]]
+; CHECK-NEXT: ret [3 x [2 x float]] [[Y]]
;
%select = select i1 %cond, [3 x [2 x float]] %x, [3 x [2 x float]] %y
ret [3 x [2 x float ]] %select
@@ -139,8 +136,7 @@ define nofpclass(inf) [3 x [2 x float]] @ret_float_array(i1 %cond, [3 x [2 x flo
define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_lhs(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_lhs
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0x7FF0000000000000, float [[X]]
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float [[X]]
;
%select = select i1 %cond, float 0x7FF0000000000000, float %x
ret float %select
@@ -150,8 +146,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_lhs(i1 %cond, float
define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_rhs(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_rhs
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float [[X]], float 0x7FF0000000000000
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float [[X]]
;
%select = select i1 %cond, float %x, float 0x7FF0000000000000
ret float %select
@@ -161,8 +156,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_rhs(i1 %cond, float
define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_or_ninf(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_or_ninf
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0x7FF0000000000000, float 0xFFF0000000000000
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float poison
;
%select = select i1 %cond, float 0x7FF0000000000000, float 0xFFF0000000000000
ret float %select
@@ -172,8 +166,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_pinf_or_ninf(i1 %cond, fl
define nofpclass(inf) float @ret_nofpclass_inf__select_ninf_or_pinf(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(inf) float @ret_nofpclass_inf__select_ninf_or_pinf
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0xFFF0000000000000, float 0x7FF0000000000000
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float poison
;
%select = select i1 %cond, float 0xFFF0000000000000, float 0x7FF0000000000000
ret float %select
@@ -183,8 +176,7 @@ define nofpclass(inf) float @ret_nofpclass_inf__select_ninf_or_pinf(i1 %cond, fl
define nofpclass(ninf) float @ret_nofpclass_ninf__select_ninf_or_pinf(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(ninf) float @ret_nofpclass_ninf__select_ninf_or_pinf
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0xFFF0000000000000, float 0x7FF0000000000000
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float 0x7FF0000000000000
;
%select = select i1 %cond, float 0xFFF0000000000000, float 0x7FF0000000000000
ret float %select
@@ -194,8 +186,7 @@ define nofpclass(ninf) float @ret_nofpclass_ninf__select_ninf_or_pinf(i1 %cond,
define nofpclass(pinf) float @ret_nofpclass_pinf__select_ninf_or_pinf(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(pinf) float @ret_nofpclass_pinf__select_ninf_or_pinf
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0xFFF0000000000000, float 0x7FF0000000000000
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float 0xFFF0000000000000
;
%select = select i1 %cond, float 0xFFF0000000000000, float 0x7FF0000000000000
ret float %select
@@ -205,8 +196,7 @@ define nofpclass(pinf) float @ret_nofpclass_pinf__select_ninf_or_pinf(i1 %cond,
define nofpclass(zero) float @ret_nofpclass_zero__select_pzero_or_nzero(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(zero) float @ret_nofpclass_zero__select_pzero_or_nzero
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0.000000e+00, float -0.000000e+00
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float poison
;
%select = select i1 %cond, float 0.0, float -0.0
ret float %select
@@ -216,8 +206,7 @@ define nofpclass(zero) float @ret_nofpclass_zero__select_pzero_or_nzero(i1 %cond
define nofpclass(nzero) float @ret_nofpclass_nzero__select_pzero_or_nzero(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(nzero) float @ret_nofpclass_nzero__select_pzero_or_nzero
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: [[SELECT:%.*]] = select i1 [[COND]], float 0.000000e+00, float -0.000000e+00
-; CHECK-NEXT: ret float [[SELECT]]
+; CHECK-NEXT: ret float 0.000000e+00
;
%select = select i1 %cond, float 0.0, float -0.0
ret float %select
@@ -227,8 +216,7 @@ define nofpclass(nzero) float @ret_nofpclass_nzero__select_pzero_or_nzero(i1 %co
define nofpclass(pzero) float @ret_nofpclass_pzero__select_pzero_or_nzero(i1 %cond, float %x) {
; CHECK-LABEL: define nofpclass(pzero) float @ret_nofpclass_pzero__select_pzero_or_nzero
; CHECK-SAME: (i1 [[COND:%.*]], float [[X:%.*]]) {
-; CHECK-NEXT: ...
[truncated]
``````````
</details>
https://github.com/llvm/llvm-project/pull/74056
More information about the cfe-commits
mailing list