Skip to content

Commit 9a6ad28

Browse files
committed
[inline] Clone return range attribute on the callsite intoinlined call
1 parent 6dd513a commit 9a6ad28

File tree

3 files changed

+18
-7
lines changed

3 files changed

+18
-7
lines changed

clang/test/Headers/__clang_hip_math.hip

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -231,7 +231,7 @@ extern "C" __device__ uint64_t test___make_mantissa(const char *p) {
231231

232232
// CHECK-LABEL: @test_abs(
233233
// CHECK-NEXT: entry:
234-
// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef i32 @llvm.abs.i32(i32 [[X:%.*]], i1 true)
234+
// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef range(i32 0, -2147483648) i32 @llvm.abs.i32(i32 [[X:%.*]], i1 true)
235235
// CHECK-NEXT: ret i32 [[TMP0]]
236236
//
237237
extern "C" __device__ int test_abs(int x) {
@@ -240,7 +240,7 @@ extern "C" __device__ int test_abs(int x) {
240240

241241
// CHECK-LABEL: @test_labs(
242242
// CHECK-NEXT: entry:
243-
// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true)
243+
// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef range(i64 0, -9223372036854775808) i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true)
244244
// CHECK-NEXT: ret i64 [[TMP0]]
245245
//
246246
extern "C" __device__ long test_labs(long x) {
@@ -249,7 +249,7 @@ extern "C" __device__ long test_labs(long x) {
249249

250250
// CHECK-LABEL: @test_llabs(
251251
// CHECK-NEXT: entry:
252-
// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true)
252+
// CHECK-NEXT: [[TMP0:%.*]] = tail call noundef range(i64 0, -9223372036854775808) i64 @llvm.abs.i64(i64 [[X:%.*]], i1 true)
253253
// CHECK-NEXT: ret i64 [[TMP0]]
254254
//
255255
extern "C" __device__ long long test_llabs(long x) {

llvm/lib/Transforms/Utils/InlineFunction.cpp

Lines changed: 12 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -30,11 +30,12 @@
3030
#include "llvm/Analysis/ProfileSummaryInfo.h"
3131
#include "llvm/Analysis/ValueTracking.h"
3232
#include "llvm/Analysis/VectorUtils.h"
33-
#include "llvm/IR/AttributeMask.h"
3433
#include "llvm/IR/Argument.h"
34+
#include "llvm/IR/AttributeMask.h"
3535
#include "llvm/IR/BasicBlock.h"
3636
#include "llvm/IR/CFG.h"
3737
#include "llvm/IR/Constant.h"
38+
#include "llvm/IR/ConstantRange.h"
3839
#include "llvm/IR/Constants.h"
3940
#include "llvm/IR/DataLayout.h"
4041
#include "llvm/IR/DebugInfo.h"
@@ -1444,6 +1445,8 @@ static AttrBuilder IdentifyValidPoisonGeneratingAttributes(CallBase &CB) {
14441445
Valid.addAttribute(Attribute::NonNull);
14451446
if (CB.hasRetAttr(Attribute::Alignment))
14461447
Valid.addAlignmentAttr(CB.getRetAlign());
1448+
if (CB.hasRetAttr(Attribute::Range))
1449+
Valid.addRangeAttr(*CB.getRange());
14471450
return Valid;
14481451
}
14491452

@@ -1535,6 +1538,14 @@ static void AddReturnAttributes(CallBase &CB, ValueToValueMapTy &VMap) {
15351538
if (ValidPG.getAlignment().valueOrOne() < AL.getRetAlignment().valueOrOne())
15361539
ValidPG.removeAttribute(Attribute::Alignment);
15371540
if (ValidPG.hasAttributes()) {
1541+
Attribute CBRange = ValidPG.getAttribute(Attribute::Range);
1542+
if (CBRange.isValid()) {
1543+
Attribute NewRange = AL.getRetAttr(Attribute::Range);
1544+
if (NewRange.isValid()) {
1545+
ValidPG.addRangeAttr(
1546+
CBRange.getRange().intersectWith(NewRange.getRange()));
1547+
}
1548+
}
15381549
// Three checks.
15391550
// If the callsite has `noundef`, then a poison due to violating the
15401551
// return attribute will create UB anyways so we can always propagate.

llvm/test/Transforms/Inline/ret_attr_align_and_noundef.ll

Lines changed: 3 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -351,7 +351,7 @@ define i8 @callee13() {
351351

352352
define i8 @caller13_okay_use_after_poison_anyways() {
353353
; CHECK-LABEL: define i8 @caller13_okay_use_after_poison_anyways() {
354-
; CHECK-NEXT: [[R_I:%.*]] = call i8 @val8()
354+
; CHECK-NEXT: [[R_I:%.*]] = call range(i8 0, 10) i8 @val8()
355355
; CHECK-NEXT: call void @use.val(i8 [[R_I]])
356356
; CHECK-NEXT: ret i8 [[R_I]]
357357
;
@@ -382,7 +382,7 @@ define i8 @caller14_fail_creates_ub() {
382382

383383
define i8 @caller14_okay_is_ub_anyways() {
384384
; CHECK-LABEL: define i8 @caller14_okay_is_ub_anyways() {
385-
; CHECK-NEXT: [[R_I:%.*]] = call noundef i8 @val8()
385+
; CHECK-NEXT: [[R_I:%.*]] = call noundef range(i8 0, 10) i8 @val8()
386386
; CHECK-NEXT: call void @use.val(i8 [[R_I]])
387387
; CHECK-NEXT: ret i8 [[R_I]]
388388
;
@@ -402,7 +402,7 @@ define i8 @callee15() {
402402

403403
define i8 @caller15_okay_intersect_ranges() {
404404
; CHECK-LABEL: define i8 @caller15_okay_intersect_ranges() {
405-
; CHECK-NEXT: [[R_I:%.*]] = call range(i8 5, 10) i8 @val8()
405+
; CHECK-NEXT: [[R_I:%.*]] = call range(i8 5, 7) i8 @val8()
406406
; CHECK-NEXT: call void @use.val(i8 [[R_I]])
407407
; CHECK-NEXT: ret i8 [[R_I]]
408408
;

0 commit comments

Comments
 (0)