diff --git a/clang/test/Headers/__clang_hip_math.hip b/clang/test/Headers/__clang_hip_math.hip index 1271868a53b866..26da82843c5124 100644 --- a/clang/test/Headers/__clang_hip_math.hip +++ b/clang/test/Headers/__clang_hip_math.hip @@ -231,7 +231,7 @@ extern "C" __device__ uint64_t test___make_mantissa(const char *p) { // CHECK-LABEL: @test_abs( // CHECK-NEXT: entry: -// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef i32 @llvm.abs.i32(i32 [[X:%.*]], i1 true) +// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef range(i32 0, -2147483648) i32 @llvm.abs.i32(i32 [[X:%.*]], i1 true) // CHECK-NEXT: ret i32 [[TMP0]] // extern "C" __device__ int test_abs(int x) { @@ -240,7 +240,7 @@ extern "C" __device__ int test_abs(int x) { // CHECK-LABEL: @test_labs( // CHECK-NEXT: entry: -// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true) +// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef range(i64 0, -9223372036854775808) i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true) // CHECK-NEXT: ret i64 [[TMP0]] // extern "C" __device__ long test_labs(long x) { @@ -249,7 +249,7 @@ extern "C" __device__ long test_labs(long x) { // CHECK-LABEL: @test_llabs( // CHECK-NEXT: entry: -// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true) +// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef range(i64 0, -9223372036854775808) i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true) // CHECK-NEXT: ret i64 [[TMP0]] // extern "C" __device__ long long test_llabs(long x) { diff --git a/llvm/lib/Transforms/Utils/InlineFunction.cpp b/llvm/lib/Transforms/Utils/InlineFunction.cpp index 7b846f2d2d72d6..eb471b259c7d4e 100644 --- a/llvm/lib/Transforms/Utils/InlineFunction.cpp +++ b/llvm/lib/Transforms/Utils/InlineFunction.cpp @@ -30,11 +30,12 @@ #include "llvm/Analysis/ProfileSummaryInfo.h" #include "llvm/Analysis/ValueTracking.h" #include "llvm/Analysis/VectorUtils.h" -#include "llvm/IR/AttributeMask.h" #include "llvm/IR/Argument.h" +#include "llvm/IR/AttributeMask.h" #include "llvm/IR/BasicBlock.h" #include "llvm/IR/CFG.h" #include "llvm/IR/Constant.h" +#include "llvm/IR/ConstantRange.h" #include "llvm/IR/Constants.h" #include "llvm/IR/DataLayout.h" #include "llvm/IR/DebugInfo.h" @@ -1450,6 +1451,8 @@ static AttrBuilder IdentifyValidPoisonGeneratingAttributes(CallBase &CB) { Valid.addAttribute(Attribute::NonNull); if (CB.hasRetAttr(Attribute::Alignment)) Valid.addAlignmentAttr(CB.getRetAlign()); + if (std::optional Range = CB.getRange()) + Valid.addRangeAttr(*Range); return Valid; } @@ -1541,6 +1544,14 @@ static void AddReturnAttributes(CallBase &CB, ValueToValueMapTy &VMap) { if (ValidPG.getAlignment().valueOrOne() < AL.getRetAlignment().valueOrOne()) ValidPG.removeAttribute(Attribute::Alignment); if (ValidPG.hasAttributes()) { + Attribute CBRange = ValidPG.getAttribute(Attribute::Range); + if (CBRange.isValid()) { + Attribute NewRange = AL.getRetAttr(Attribute::Range); + if (NewRange.isValid()) { + ValidPG.addRangeAttr( + CBRange.getRange().intersectWith(NewRange.getRange())); + } + } // Three checks. // If the callsite has `noundef`, then a poison due to violating the // return attribute will create UB anyways so we can always propagate. diff --git a/llvm/test/Transforms/Inline/ret_attr_align_and_noundef.ll b/llvm/test/Transforms/Inline/ret_attr_align_and_noundef.ll index 7e76401c0b4de0..f4cebf1fcb5da0 100644 --- a/llvm/test/Transforms/Inline/ret_attr_align_and_noundef.ll +++ b/llvm/test/Transforms/Inline/ret_attr_align_and_noundef.ll @@ -351,7 +351,7 @@ define i8 @callee13() { define i8 @caller13_okay_use_after_poison_anyways() { ; CHECK-LABEL: define i8 @caller13_okay_use_after_poison_anyways() { -; CHECK-NEXT: [[R_I:%.*]] = call i8 @val8() +; CHECK-NEXT: [[R_I:%.*]] = call range(i8 0, 10) i8 @val8() ; CHECK-NEXT: call void @use.val(i8 [[R_I]]) ; CHECK-NEXT: ret i8 [[R_I]] ; @@ -382,7 +382,7 @@ define i8 @caller14_fail_creates_ub() { define i8 @caller14_okay_is_ub_anyways() { ; CHECK-LABEL: define i8 @caller14_okay_is_ub_anyways() { -; CHECK-NEXT: [[R_I:%.*]] = call noundef i8 @val8() +; CHECK-NEXT: [[R_I:%.*]] = call noundef range(i8 0, 10) i8 @val8() ; CHECK-NEXT: call void @use.val(i8 [[R_I]]) ; CHECK-NEXT: ret i8 [[R_I]] ; @@ -402,7 +402,7 @@ define i8 @callee15() { define i8 @caller15_okay_intersect_ranges() { ; CHECK-LABEL: define i8 @caller15_okay_intersect_ranges() { -; CHECK-NEXT: [[R_I:%.*]] = call range(i8 5, 10) i8 @val8() +; CHECK-NEXT: [[R_I:%.*]] = call range(i8 5, 7) i8 @val8() ; CHECK-NEXT: call void @use.val(i8 [[R_I]]) ; CHECK-NEXT: ret i8 [[R_I]] ;