[CHERI] Allow @llvm.returnaddress to return a pointer in any address space.#188464
Merged
Conversation
Member
|
@llvm/pr-subscribers-backend-amdgpu @llvm/pr-subscribers-compiler-rt-sanitizer Author: Owen Anderson (resistor) ChangesClang now constructs calls to it using the default program address space from the DataLayout. Patch is 90.01 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/188464.diff 56 Files Affected:
diff --git a/clang/lib/CodeGen/CGBuiltin.cpp b/clang/lib/CodeGen/CGBuiltin.cpp
index df03e84ce9f81..51c5d970a0f84 100644
--- a/clang/lib/CodeGen/CGBuiltin.cpp
+++ b/clang/lib/CodeGen/CGBuiltin.cpp
@@ -4824,11 +4824,13 @@ RValue CodeGenFunction::EmitBuiltinExpr(const GlobalDecl GD, unsigned BuiltinID,
case Builtin::BI__builtin_return_address: {
Value *Depth = ConstantEmitter(*this).emitAbstract(E->getArg(0),
getContext().UnsignedIntTy);
- Function *F = CGM.getIntrinsic(Intrinsic::returnaddress);
+ Function *F =
+ CGM.getIntrinsic(Intrinsic::returnaddress, {CGM.ProgramPtrTy});
return RValue::get(Builder.CreateCall(F, Depth));
}
case Builtin::BI_ReturnAddress: {
- Function *F = CGM.getIntrinsic(Intrinsic::returnaddress);
+ Function *F =
+ CGM.getIntrinsic(Intrinsic::returnaddress, {CGM.ProgramPtrTy});
return RValue::get(Builder.CreateCall(F, Builder.getInt32(0)));
}
case Builtin::BI__builtin_frame_address: {
diff --git a/clang/lib/CodeGen/CodeGenModule.cpp b/clang/lib/CodeGen/CodeGenModule.cpp
index daaa846bf42bc..3fcd6f5f904db 100644
--- a/clang/lib/CodeGen/CodeGenModule.cpp
+++ b/clang/lib/CodeGen/CodeGenModule.cpp
@@ -451,6 +451,8 @@ CodeGenModule::CodeGenModule(ASTContext &C,
llvm::PointerType::get(LLVMContext, DL.getAllocaAddrSpace());
GlobalsInt8PtrTy =
llvm::PointerType::get(LLVMContext, DL.getDefaultGlobalsAddressSpace());
+ ProgramPtrTy =
+ llvm::PointerType::get(LLVMContext, DL.getProgramAddressSpace());
ConstGlobalsPtrTy = llvm::PointerType::get(
LLVMContext, C.getTargetAddressSpace(GetGlobalConstantAddressSpace()));
ASTAllocaAddressSpace = getTargetCodeGenInfo().getASTAllocaAddressSpace();
diff --git a/clang/lib/CodeGen/CodeGenTypeCache.h b/clang/lib/CodeGen/CodeGenTypeCache.h
index 015306bb97373..39ea8a681dc42 100644
--- a/clang/lib/CodeGen/CodeGenTypeCache.h
+++ b/clang/lib/CodeGen/CodeGenTypeCache.h
@@ -72,6 +72,9 @@ struct CodeGenTypeCache {
llvm::PointerType *GlobalsInt8PtrTy;
};
+ /// Pointer in program address space
+ llvm::PointerType *ProgramPtrTy;
+
/// void* in the address space for constant globals
llvm::PointerType *ConstGlobalsPtrTy;
diff --git a/clang/test/CodeGen/ms-intrinsics.c b/clang/test/CodeGen/ms-intrinsics.c
index 6528a63e380c2..271aced5e0b7c 100644
--- a/clang/test/CodeGen/ms-intrinsics.c
+++ b/clang/test/CodeGen/ms-intrinsics.c
@@ -134,7 +134,7 @@ void *test_ReturnAddress(void) {
return _ReturnAddress();
}
// CHECK-LABEL: define{{.*}}ptr @test_ReturnAddress()
-// CHECK: = tail call ptr @llvm.returnaddress(i32 0)
+// CHECK: = tail call ptr @llvm.returnaddress.p0(i32 0)
// CHECK: ret ptr
#if defined(__i386__) || defined(__x86_64__) || defined (__aarch64__)
diff --git a/clang/test/Headers/hip-header.hip b/clang/test/Headers/hip-header.hip
index 146a43b643dba..b4bd32dc9e3b0 100644
--- a/clang/test/Headers/hip-header.hip
+++ b/clang/test/Headers/hip-header.hip
@@ -169,7 +169,7 @@ __device__ double test_isnan() {
// MALLOC: call i64 @__ockl_dm_alloc
// NOMALLOC: call void @llvm.trap
// MALLOC-ASAN-LABEL: define weak {{.*}}ptr @malloc(i64
-// MALLOC-ASAN: call ptr @llvm.returnaddress(i32 0)
+// MALLOC-ASAN: call ptr @llvm.returnaddress.p0(i32 0)
// MALLOC-ASAN: call i64 @__asan_malloc_impl(i64 {{.*}}, i64 {{.*}})
__device__ void test_malloc(void *a) {
a = malloc(42);
@@ -183,7 +183,7 @@ __device__ void test_malloc(void *a) {
// MALLOC: call void @__ockl_dm_dealloc
// NOMALLOC: call void @llvm.trap
// MALLOC-ASAN-LABEL: define weak {{.*}}void @free(ptr
-// MALLOC-ASAN: call ptr @llvm.returnaddress(i32 0)
+// MALLOC-ASAN: call ptr @llvm.returnaddress.p0(i32 0)
// MALLOC-ASAN: call void @__asan_free_impl(i64 {{.*}}, i64 {{.*}})
__device__ void test_free(void *a) {
free(a);
diff --git a/llvm/include/llvm/IR/Intrinsics.td b/llvm/include/llvm/IR/Intrinsics.td
index 4469ff155b854..6d4b9bd4415ae 100644
--- a/llvm/include/llvm/IR/Intrinsics.td
+++ b/llvm/include/llvm/IR/Intrinsics.td
@@ -885,7 +885,7 @@ def int_swift_async_context_addr : Intrinsic<[llvm_ptr_ty], [], []>;
//===--------------------- Code Generator Intrinsics ----------------------===//
//
-def int_returnaddress : DefaultAttrsIntrinsic<[llvm_ptr_ty], [llvm_i32_ty],
+def int_returnaddress : DefaultAttrsIntrinsic<[llvm_anyptr_ty], [llvm_i32_ty],
[IntrNoMem, ImmArg<ArgIndex<0>>]>;
def int_addressofreturnaddress : DefaultAttrsIntrinsic<[llvm_anyptr_ty], [], [IntrNoMem]>;
def int_frameaddress : DefaultAttrsIntrinsic<[llvm_anyptr_ty], [llvm_i32_ty],
diff --git a/llvm/lib/Target/AMDGPU/AMDGPUSwLowerLDS.cpp b/llvm/lib/Target/AMDGPU/AMDGPUSwLowerLDS.cpp
index 362c221aa1392..04383855b946b 100644
--- a/llvm/lib/Target/AMDGPU/AMDGPUSwLowerLDS.cpp
+++ b/llvm/lib/Target/AMDGPU/AMDGPUSwLowerLDS.cpp
@@ -867,8 +867,9 @@ void AMDGPUSwLowerLDS::lowerKernelLDSAccesses(Function *Func,
// Create a call to malloc function which does device global memory allocation
// with size equals to all LDS global accesses size in this kernel.
- Value *ReturnAddress =
- IRB.CreateIntrinsic(Intrinsic::returnaddress, {IRB.getInt32(0)});
+ Value *ReturnAddress = IRB.CreateIntrinsic(
+ Intrinsic::returnaddress, IRB.getPtrTy(AMDGPUAS::FLAT_ADDRESS),
+ {IRB.getInt32(0)});
FunctionCallee MallocFunc = M.getOrInsertFunction(
StringRef("__asan_malloc_impl"),
FunctionType::get(Int64Ty, {Int64Ty, Int64Ty}, false));
@@ -933,8 +934,9 @@ void AMDGPUSwLowerLDS::lowerKernelLDSAccesses(Function *Func,
FunctionCallee AsanFreeFunc = M.getOrInsertFunction(
StringRef("__asan_free_impl"),
FunctionType::get(IRB.getVoidTy(), {Int64Ty, Int64Ty}, false));
- Value *ReturnAddr =
- IRB.CreateIntrinsic(Intrinsic::returnaddress, IRB.getInt32(0));
+ Value *ReturnAddr = IRB.CreateIntrinsic(Intrinsic::returnaddress,
+ IRB.getPtrTy(AMDGPUAS::FLAT_ADDRESS),
+ IRB.getInt32(0));
Value *RAPToInt = IRB.CreatePtrToInt(ReturnAddr, Int64Ty);
Value *MallocPtrToInt = IRB.CreatePtrToInt(LoadMallocPtr, Int64Ty);
IRB.CreateCall(AsanFreeFunc, {MallocPtrToInt, RAPToInt});
diff --git a/llvm/lib/Transforms/Instrumentation/ThreadSanitizer.cpp b/llvm/lib/Transforms/Instrumentation/ThreadSanitizer.cpp
index 811911644106b..f05efd863fb74 100644
--- a/llvm/lib/Transforms/Instrumentation/ThreadSanitizer.cpp
+++ b/llvm/lib/Transforms/Instrumentation/ThreadSanitizer.cpp
@@ -577,8 +577,10 @@ bool ThreadSanitizer::sanitizeFunction(Function &F,
if ((Res || HasCalls) && ClInstrumentFuncEntryExit) {
InstrumentationIRBuilder IRB(&F.getEntryBlock(),
F.getEntryBlock().getFirstNonPHIIt());
- Value *ReturnAddress =
- IRB.CreateIntrinsic(Intrinsic::returnaddress, IRB.getInt32(0));
+ auto ProgramAsPtrTy = PointerType::get(F.getParent()->getContext(),
+ DL.getProgramAddressSpace());
+ Value *ReturnAddress = IRB.CreateIntrinsic(
+ Intrinsic::returnaddress, {ProgramAsPtrTy}, IRB.getInt32(0));
IRB.CreateCall(TsanFuncEntry, ReturnAddress);
EscapeEnumerator EE(F, "tsan_cleanup", ClHandleCxxExceptions);
diff --git a/llvm/lib/Transforms/Utils/EntryExitInstrumenter.cpp b/llvm/lib/Transforms/Utils/EntryExitInstrumenter.cpp
index 29c17ffc41a74..71a32664c7e77 100644
--- a/llvm/lib/Transforms/Utils/EntryExitInstrumenter.cpp
+++ b/llvm/lib/Transforms/Utils/EntryExitInstrumenter.cpp
@@ -53,8 +53,11 @@ static void insertCall(Function &CurFn, StringRef Func,
// On RISC-V, AArch64, and LoongArch, the `_mcount` function takes
// `__builtin_return_address(0)` as an argument since
// `__builtin_return_address(1)` is not available on these platforms.
+ auto ProgASPtr =
+ PointerType::get(C, M.getDataLayout().getProgramAddressSpace());
Instruction *RetAddr = CallInst::Create(
- Intrinsic::getOrInsertDeclaration(&M, Intrinsic::returnaddress),
+ Intrinsic::getOrInsertDeclaration(&M, Intrinsic::returnaddress,
+ {ProgASPtr}),
ConstantInt::get(Type::getInt32Ty(C), 0), "", InsertionPt);
RetAddr->setDebugLoc(DL);
@@ -77,13 +80,16 @@ static void insertCall(Function &CurFn, StringRef Func,
}
if (Func == "__cyg_profile_func_enter" || Func == "__cyg_profile_func_exit") {
- Type *ArgTypes[] = {PointerType::getUnqual(C), PointerType::getUnqual(C)};
+ auto ProgASPtr =
+ PointerType::get(C, M.getDataLayout().getProgramAddressSpace());
+ Type *ArgTypes[] = {ProgASPtr, ProgASPtr};
FunctionCallee Fn = M.getOrInsertFunction(
Func, FunctionType::get(Type::getVoidTy(C), ArgTypes, false));
Instruction *RetAddr = CallInst::Create(
- Intrinsic::getOrInsertDeclaration(&M, Intrinsic::returnaddress),
+ Intrinsic::getOrInsertDeclaration(&M, Intrinsic::returnaddress,
+ {ProgASPtr}),
ArrayRef<Value *>(ConstantInt::get(Type::getInt32Ty(C), 0)), "",
InsertionPt);
RetAddr->setDebugLoc(DL);
diff --git a/llvm/test/Bitcode/compatibility-3.6.ll b/llvm/test/Bitcode/compatibility-3.6.ll
index 2148e013126b3..62b5a88d085c6 100644
--- a/llvm/test/Bitcode/compatibility-3.6.ll
+++ b/llvm/test/Bitcode/compatibility-3.6.ll
@@ -1112,7 +1112,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility-3.7.ll b/llvm/test/Bitcode/compatibility-3.7.ll
index fed9cce2a0091..61cc50ef4dead 100644
--- a/llvm/test/Bitcode/compatibility-3.7.ll
+++ b/llvm/test/Bitcode/compatibility-3.7.ll
@@ -1143,7 +1143,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility-3.8.ll b/llvm/test/Bitcode/compatibility-3.8.ll
index 92695b9a41b80..19a5c0f7a4e1f 100644
--- a/llvm/test/Bitcode/compatibility-3.8.ll
+++ b/llvm/test/Bitcode/compatibility-3.8.ll
@@ -1298,7 +1298,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility-3.9.ll b/llvm/test/Bitcode/compatibility-3.9.ll
index aa11917332e11..b29463940424a 100644
--- a/llvm/test/Bitcode/compatibility-3.9.ll
+++ b/llvm/test/Bitcode/compatibility-3.9.ll
@@ -1369,7 +1369,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility-4.0.ll b/llvm/test/Bitcode/compatibility-4.0.ll
index cefccdc02c08c..0d3a024af511d 100644
--- a/llvm/test/Bitcode/compatibility-4.0.ll
+++ b/llvm/test/Bitcode/compatibility-4.0.ll
@@ -1369,7 +1369,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility-5.0.ll b/llvm/test/Bitcode/compatibility-5.0.ll
index ae3e2e8ffbb0f..c59701c5915aa 100644
--- a/llvm/test/Bitcode/compatibility-5.0.ll
+++ b/llvm/test/Bitcode/compatibility-5.0.ll
@@ -1381,7 +1381,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility-6.0.ll b/llvm/test/Bitcode/compatibility-6.0.ll
index cfb5ff7b350a2..f0b18a8c8145e 100644
--- a/llvm/test/Bitcode/compatibility-6.0.ll
+++ b/llvm/test/Bitcode/compatibility-6.0.ll
@@ -1391,7 +1391,7 @@ declare void @llvm.instrprof_increment(i8*, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call i8* @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call i8* @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/Bitcode/compatibility.ll b/llvm/test/Bitcode/compatibility.ll
index c87159fe960f3..f2df4c68404fb 100644
--- a/llvm/test/Bitcode/compatibility.ll
+++ b/llvm/test/Bitcode/compatibility.ll
@@ -1887,7 +1887,7 @@ declare void @llvm.instrprof_increment(ptr, i64, i32, i32)
!10 = !{!"rax"}
define void @intrinsics.codegen() {
call ptr @llvm.returnaddress(i32 1)
- ; CHECK: call ptr @llvm.returnaddress(i32 1)
+ ; CHECK: call ptr @llvm.returnaddress.p0(i32 1)
call ptr @llvm.frameaddress(i32 1)
; CHECK: call ptr @llvm.frameaddress.p0(i32 1)
diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access-asan.ll
index 4e53df3924985..e3a28b6379077 100644
--- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access-asan.ll
+++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access-asan.ll
@@ -122,7 +122,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: store i32 [[TMP19]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 4, i32 2), align 4
; CHECK-NEXT: [[TMP28:%.*]] = add i32 [[TMP15]], [[TMP19]]
; CHECK-NEXT: [[TMP26:%.*]] = zext i32 [[TMP28]] to i64
-; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP23:%.*]] = ptrtoint ptr [[TMP22]] to i64
; CHECK-NEXT: [[TMP35:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP26]], i64 [[TMP23]])
; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP35]] to ptr addrspace(1)
@@ -227,7 +227,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: call void @llvm.amdgcn.s.barrier()
; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]]
; CHECK: Free:
-; CHECK-NEXT: [[TMP32:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP32:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP33:%.*]] = ptrtoint ptr [[TMP32]] to i64
; CHECK-NEXT: [[TMP34:%.*]] = ptrtoint ptr addrspace(1) [[TMP31]] to i64
; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP34]], i64 [[TMP33]])
diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access.ll
index 8cbeb80d62335..c155a99ccca80 100644
--- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access.ll
+++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-indirect-access.ll
@@ -76,7 +76,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: store i32 [[TMP19]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 4, i32 2), align 4
; CHECK-NEXT: [[TMP28:%.*]] = add i32 [[TMP15]], [[TMP19]]
; CHECK-NEXT: [[TMP26:%.*]] = zext i32 [[TMP28]] to i64
-; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP23:%.*]] = ptrtoint ptr [[TMP22]] to i64
; CHECK-NEXT: [[TMP35:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP26]], i64 [[TMP23]])
; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP35]] to ptr addrspace(1)
@@ -112,7 +112,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: call void @llvm.amdgcn.s.barrier()
; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]]
; CHECK: Free:
-; CHECK-NEXT: [[TMP32:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP32:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP33:%.*]] = ptrtoint ptr [[TMP32]] to i64
; CHECK-NEXT: [[TMP34:%.*]] = ptrtoint ptr addrspace(1) [[TMP31]] to i64
; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP34]], i64 [[TMP33]])
diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test-asan.ll
index 32601422c7e67..12dcc92f49dc6 100644
--- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test-asan.ll
+++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test-asan.ll
@@ -36,7 +36,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: store i32 [[TMP16]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 1, i32 2), align 4
; CHECK-NEXT: [[TMP17:%.*]] = add i32 [[TMP24]], [[TMP16]]
; CHECK-NEXT: [[TMP21:%.*]] = zext i32 [[TMP17]] to i64
-; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP23:%.*]] = ptrtoint ptr [[TMP22]] to i64
; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP21]], i64 [[TMP23]])
; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1)
@@ -82,7 +82,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: call void @llvm.amdgcn.s.barrier()
; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]]
; CHECK: Free:
-; CHECK-NEXT: [[TMP25:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP25:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP26:%.*]] = ptrtoint ptr [[TMP25]] to i64
; CHECK-NEXT: [[TMP27:%.*]] = ptrtoint ptr addrspace(1) [[TMP28]] to i64
; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP27]], i64 [[TMP26]])
diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test.ll
index 5e90eb0b95219..f6876702dc0bb 100644
--- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test.ll
+++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-dynamic-lds-test.ll
@@ -36,7 +36,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: store i32 [[TMP16]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 1, i32 2), align 4
; CHECK-NEXT: [[TMP17:%.*]] = add i32 [[TMP24]], [[TMP16]]
; CHECK-NEXT: [[TMP21:%.*]] = zext i32 [[TMP17]] to i64
-; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP22:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEXT: [[TMP23:%.*]] = ptrtoint ptr [[TMP22]] to i64
; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP21]], i64 [[TMP23]])
; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1)
@@ -60,7 +60,7 @@ define amdgpu_kernel void @k0() sanitize_address {
; CHECK-NEXT: call void @llvm.amdgcn.s.barrier()
; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]]
; CHECK: Free:
-; CHECK-NEXT: [[TMP25:%.*]] = call ptr @llvm.returnaddress(i32 0)
+; CHECK-NEXT: [[TMP25:%.*]] = call ptr @llvm.returnaddress.p0(i32 0)
; CHECK-NEX...
[truncated]
|
arsenm
reviewed
Mar 25, 2026
arsenm
approved these changes
Mar 25, 2026
jrtc27
approved these changes
Mar 25, 2026
jrtc27
left a comment
Contributor
There was a problem hiding this comment.
Thanks. Do we have others like this still? I forget which have been upstreamed (or reinvented by others) over the years.
Contributor
Author
We have similar changes for It's my intent to get to those as well, but this one was the most invasive and felt like it deserved its own PR. |
1 similar comment
Contributor
Author
We have similar changes for It's my intent to get to those as well, but this one was the most invasive and felt like it deserved its own PR. |
…space. Clang now constructs calls to it using the default program address space from the DataLayout. Co-authored-by: Alex Richardson <alexrichardson@google.com>
andykaylor
added a commit
to andykaylor/llvm-project
that referenced
this pull request
Mar 25, 2026
A recent change to the handling of llvm.frameaddress intrinsics broke a couple of CIR tests. This updates the CIR test to match the new output. See llvm#188464
andykaylor
pushed a commit
that referenced
this pull request
Mar 25, 2026
Fix llvm.returnaddress call in tests after changes from #188464
Aadarsh-Keshri
pushed a commit
to Aadarsh-Keshri/llvm-project
that referenced
this pull request
Mar 28, 2026
…space. (llvm#188464) Clang now constructs calls to it using the default program address space from the DataLayout. Co-authored-by: Alex Richardson <alexrichardson@google.com>
Aadarsh-Keshri
pushed a commit
to Aadarsh-Keshri/llvm-project
that referenced
this pull request
Mar 28, 2026
Fix llvm.returnaddress call in tests after changes from llvm#188464
JonathanBrouwer
added a commit
to JonathanBrouwer/rust
that referenced
this pull request
May 11, 2026
LLVM 23: Specify `returnaddress` intrinsic return type llvm/llvm-project#188464 made the return type of the intrinsic generic to support different pointer address spaces. @rustbot label llvm-main
JonathanBrouwer
added a commit
to JonathanBrouwer/rust
that referenced
this pull request
May 11, 2026
LLVM 23: Specify `returnaddress` intrinsic return type llvm/llvm-project#188464 made the return type of the intrinsic generic to support different pointer address spaces. @rustbot label llvm-main
rust-timer
added a commit
to rust-lang/rust
that referenced
this pull request
May 12, 2026
Rollup merge of #156461 - TimNN:retty, r=nikic LLVM 23: Specify `returnaddress` intrinsic return type llvm/llvm-project#188464 made the return type of the intrinsic generic to support different pointer address spaces. @rustbot label llvm-main
WhySoBad
pushed a commit
to WhySoBad/miri
that referenced
this pull request
May 13, 2026
LLVM 23: Specify `returnaddress` intrinsic return type llvm/llvm-project#188464 made the return type of the intrinsic generic to support different pointer address spaces. @rustbot label llvm-main
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
4 participants
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Clang now constructs calls to it using the default program address space from the DataLayout.
Co-authored-by: Alex Richardson alexrichardson@google.com