https://github.com/resistor created https://github.com/llvm/llvm-project/pull/188464
Clang now constructs calls to it using the default program address space from the DataLayout. >From 18f057621ae802fe0dcefcac2779ed2ba0db0b8b Mon Sep 17 00:00:00 2001 From: Owen Anderson <[email protected]> Date: Wed, 25 Mar 2026 12:26:04 +0100 Subject: [PATCH] [CHERI] Allow @llvm.returnaddress to return a pointer in any address space. Clang now constructs calls to it using the default program address space from the DataLayout. --- clang/lib/CodeGen/CGBuiltin.cpp | 6 ++++-- clang/lib/CodeGen/CodeGenModule.cpp | 2 ++ clang/lib/CodeGen/CodeGenTypeCache.h | 3 +++ clang/test/CodeGen/ms-intrinsics.c | 2 +- clang/test/Headers/hip-header.hip | 4 ++-- llvm/include/llvm/IR/Intrinsics.td | 2 +- llvm/lib/Target/AMDGPU/AMDGPUSwLowerLDS.cpp | 10 ++++++---- .../Instrumentation/ThreadSanitizer.cpp | 6 ++++-- .../Transforms/Utils/EntryExitInstrumenter.cpp | 12 +++++++++--- llvm/test/Bitcode/compatibility-3.6.ll | 2 +- llvm/test/Bitcode/compatibility-3.7.ll | 2 +- llvm/test/Bitcode/compatibility-3.8.ll | 2 +- llvm/test/Bitcode/compatibility-3.9.ll | 2 +- llvm/test/Bitcode/compatibility-4.0.ll | 2 +- llvm/test/Bitcode/compatibility-5.0.ll | 2 +- llvm/test/Bitcode/compatibility-6.0.ll | 2 +- llvm/test/Bitcode/compatibility.ll | 2 +- ...sw-lower-lds-dynamic-indirect-access-asan.ll | 4 ++-- ...dgpu-sw-lower-lds-dynamic-indirect-access.ll | 4 ++-- ...amdgpu-sw-lower-lds-dynamic-lds-test-asan.ll | 4 ++-- .../amdgpu-sw-lower-lds-dynamic-lds-test.ll | 4 ++-- .../AMDGPU/amdgpu-sw-lower-lds-lower-all.ll | 8 ++++---- ...multi-static-dynamic-indirect-access-asan.ll | 8 ++++---- ...-lds-multi-static-dynamic-indirect-access.ll | 8 ++++---- ...-sw-lower-lds-multiple-blocks-return-asan.ll | 4 ++-- ...mdgpu-sw-lower-lds-multiple-blocks-return.ll | 5 +++-- ...mdgpu-sw-lower-lds-non-kernel-declaration.ll | 4 ++-- ...r-lds-static-dynamic-indirect-access-asan.ll | 4 ++-- ...-lower-lds-static-dynamic-indirect-access.ll | 4 ++-- ...sw-lower-lds-static-dynamic-lds-test-asan.ll | 4 ++-- ...dgpu-sw-lower-lds-static-dynamic-lds-test.ll | 5 +++-- ...-sw-lower-lds-static-indirect-access-asan.ll | 4 ++-- ...tatic-indirect-access-function-param-asan.ll | 4 ++-- ...lds-static-indirect-access-function-param.ll | 4 ++-- ...ower-lds-static-indirect-access-lower-all.ll | 4 ++-- ...er-lds-static-indirect-access-nested-asan.ll | 17 ++++++++--------- ...w-lower-lds-static-indirect-access-nested.ll | 16 ++++++++-------- ...s-static-indirect-access-no-kernel-lds-id.ll | 4 ++-- ...mdgpu-sw-lower-lds-static-indirect-access.ll | 4 ++-- .../AMDGPU/amdgpu-sw-lower-lds-static-lds-O0.ll | 4 ++-- ...mdgpu-sw-lower-lds-static-lds-no-heap-ptr.ll | 4 ++-- .../amdgpu-sw-lower-lds-static-lds-test-asan.ll | 4 ++-- ...r-lds-static-lds-test-atomic-cmpxchg-asan.ll | 4 ++-- ...-lower-lds-static-lds-test-atomicrmw-asan.ll | 4 ++-- .../amdgpu-sw-lower-lds-static-lds-test.ll | 4 ++-- ...mdgpu-sw-lower-lds-static-lds-vector-ptrs.ll | 4 ++-- .../ThreadSanitizer/atomic-non-integer.ll | 12 ++++++------ llvm/test/Instrumentation/ThreadSanitizer/eh.ll | 10 +++++----- .../ThreadSanitizer/no_sanitize_thread.ll | 2 +- .../sanitize-thread-no-checking.ll | 2 +- .../EntryExitInstrumenter/debug-info.ll | 4 ++-- .../EntryExitInstrumenter/mcount-with-frompc.ll | 2 +- .../Transforms/EntryExitInstrumenter/mcount.ll | 16 ++++++++-------- .../pre-inliner-instrumentation.ll | 4 ++-- .../test/Verifier/LoongArch/intrinsic-immarg.ll | 2 +- llvm/test/Verifier/intrinsic-immarg.ll | 2 +- 56 files changed, 146 insertions(+), 128 deletions(-) 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-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-lower-all.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-lower-all.ll index f30a382a62c6b..bb8e762bafe6e 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-lower-all.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-lower-all.ll @@ -22,7 +22,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -56,7 +56,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: [[TMP30:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP30:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP31:%.*]] = ptrtoint ptr [[TMP30]] to i64 ; CHECK-NEXT: [[TMP32:%.*]] = ptrtoint ptr addrspace(1) [[TMP21]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP32]], i64 [[TMP31]]) @@ -85,7 +85,7 @@ define amdgpu_kernel void @k1() { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K1_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k1.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -111,7 +111,7 @@ define amdgpu_kernel void @k1() { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP24:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP24:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP25:%.*]] = ptrtoint ptr [[TMP24]] to i64 ; CHECK-NEXT: [[TMP26:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP26]], i64 [[TMP25]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access-asan.ll index 91e0a9fc5018b..ca41da7ec42e9 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access-asan.ll @@ -210,7 +210,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 3, i32 2), align 4 ; CHECK-NEXT: [[TMP26:%.*]] = add i32 [[TMP15]], [[TMP19]] ; CHECK-NEXT: [[TMP27:%.*]] = zext i32 [[TMP26]] to i64 -; CHECK-NEXT: [[TMP28:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP28:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP33:%.*]] = ptrtoint ptr [[TMP28]] to i64 ; CHECK-NEXT: [[TMP24:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP27]], i64 [[TMP33]]) ; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP24]] to ptr addrspace(1) @@ -260,7 +260,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: [[TMP30:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP30:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP31:%.*]] = ptrtoint ptr [[TMP30]] to i64 ; CHECK-NEXT: [[TMP32:%.*]] = ptrtoint ptr addrspace(1) [[TMP29]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP32]], i64 [[TMP31]]) @@ -307,7 +307,7 @@ define amdgpu_kernel void @k1() sanitize_address { ; CHECK-NEXT: store i32 [[TMP24]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K1_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k1.md, i32 0, i32 4, i32 2), align 4 ; CHECK-NEXT: [[TMP25:%.*]] = add i32 [[TMP20]], [[TMP24]] ; CHECK-NEXT: [[TMP26:%.*]] = zext i32 [[TMP25]] to i64 -; CHECK-NEXT: [[TMP27:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP27:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP28:%.*]] = ptrtoint ptr [[TMP27]] to i64 ; CHECK-NEXT: [[TMP34:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP26]], i64 [[TMP28]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP34]] to ptr addrspace(1) @@ -361,7 +361,7 @@ define amdgpu_kernel void @k1() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; CHECK-NEXT: [[TMP35:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP35:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP36:%.*]] = ptrtoint ptr [[TMP35]] to i64 ; CHECK-NEXT: [[TMP37:%.*]] = ptrtoint ptr addrspace(1) [[TMP29]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP37]], i64 [[TMP36]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access.ll index d0caddb7934a7..f10bdd9c05ef0 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multi-static-dynamic-indirect-access.ll @@ -97,7 +97,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 3, i32 2), align 4 ; CHECK-NEXT: [[TMP26:%.*]] = add i32 [[TMP15]], [[TMP19]] ; CHECK-NEXT: [[TMP27:%.*]] = zext i32 [[TMP26]] to i64 -; CHECK-NEXT: [[TMP28:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP28:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP33:%.*]] = ptrtoint ptr [[TMP28]] to i64 ; CHECK-NEXT: [[TMP24:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP27]], i64 [[TMP33]]) ; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP24]] to ptr addrspace(1) @@ -125,7 +125,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: [[TMP30:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP30:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP31:%.*]] = ptrtoint ptr [[TMP30]] to i64 ; CHECK-NEXT: [[TMP32:%.*]] = ptrtoint ptr addrspace(1) [[TMP29]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP32]], i64 [[TMP31]]) @@ -172,7 +172,7 @@ define amdgpu_kernel void @k1() sanitize_address { ; CHECK-NEXT: store i32 [[TMP24]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K1_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k1.md, i32 0, i32 4, i32 2), align 4 ; CHECK-NEXT: [[TMP25:%.*]] = add i32 [[TMP20]], [[TMP24]] ; CHECK-NEXT: [[TMP26:%.*]] = zext i32 [[TMP25]] to i64 -; CHECK-NEXT: [[TMP27:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP27:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP28:%.*]] = ptrtoint ptr [[TMP27]] to i64 ; CHECK-NEXT: [[TMP34:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP26]], i64 [[TMP28]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP34]] to ptr addrspace(1) @@ -204,7 +204,7 @@ define amdgpu_kernel void @k1() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; CHECK-NEXT: [[TMP35:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP35:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP36:%.*]] = ptrtoint ptr [[TMP35]] to i64 ; CHECK-NEXT: [[TMP37:%.*]] = ptrtoint ptr addrspace(1) [[TMP29]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP37]], i64 [[TMP36]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return-asan.ll index 07baf90e370d1..44522c1d8ebbe 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return-asan.ll @@ -27,7 +27,7 @@ define amdgpu_kernel void @test_kernel() sanitize_address { ; CHECK-NEXT: [[TMP16:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_TEST_KERNEL_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.test_kernel.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP18:%.*]] = add i32 [[TMP15]], [[TMP16]] ; CHECK-NEXT: [[TMP17:%.*]] = zext i32 [[TMP18]] to i64 -; CHECK-NEXT: [[TMP14:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP14:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP19:%.*]] = ptrtoint ptr [[TMP14]] to i64 ; CHECK-NEXT: [[TMP20:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP17]], i64 [[TMP19]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP20]] to ptr addrspace(1) @@ -70,7 +70,7 @@ define amdgpu_kernel void @test_kernel() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; 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: [[TMP24:%.*]] = ptrtoint ptr addrspace(1) [[TMP21]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP24]], i64 [[TMP23]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return.ll index 6848e2c06c1e1..2cc5a2af75859 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-multiple-blocks-return.ll @@ -27,7 +27,7 @@ define amdgpu_kernel void @test_kernel() sanitize_address { ; CHECK-NEXT: [[TMP16:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_TEST_KERNEL_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.test_kernel.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP18:%.*]] = add i32 [[TMP15]], [[TMP16]] ; CHECK-NEXT: [[TMP17:%.*]] = zext i32 [[TMP18]] to i64 -; CHECK-NEXT: [[TMP14:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP14:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP19:%.*]] = ptrtoint ptr [[TMP14]] to i64 ; CHECK-NEXT: [[TMP20:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP17]], i64 [[TMP19]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP20]] to ptr addrspace(1) @@ -70,7 +70,7 @@ define amdgpu_kernel void @test_kernel() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; 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: [[TMP24:%.*]] = ptrtoint ptr addrspace(1) [[TMP21]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP24]], i64 [[TMP23]]) @@ -109,4 +109,5 @@ ret void ; CHECK: attributes #[[ATTR3:[0-9]+]] = { convergent nocallback nofree nounwind willreturn } ;. ; CHECK: [[META0]] = !{i32 0, i32 1} +; CHECK: [[META1:![0-9]+]] = !{i32 4, !"nosanitize_address", i32 1} ;. diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-non-kernel-declaration.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-non-kernel-declaration.ll index a6e6b84bba304..fb50a74ec55b5 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-non-kernel-declaration.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-non-kernel-declaration.ll @@ -50,7 +50,7 @@ define amdgpu_kernel void @k1() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K1_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k1.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -73,7 +73,7 @@ define amdgpu_kernel void @k1() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access-asan.ll index bad2d8e0fb5f4..310398ad6948c 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access-asan.ll @@ -123,7 +123,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) @@ -228,7 +228,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-static-dynamic-indirect-access.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access.ll index 0cc49c94e2279..3b4785872a675 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-indirect-access.ll @@ -77,7 +77,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) @@ -113,7 +113,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-static-dynamic-lds-test-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test-asan.ll index c5985e5cc4df8..e5c606addb3b0 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test-asan.ll @@ -46,7 +46,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: [[TMP32:%.*]] = add i32 [[TMP15]], [[TMP19]] ; CHECK-NEXT: [[TMP30:%.*]] = zext i32 [[TMP32]] 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: [[TMP39:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP30]], i64 [[TMP23]]) ; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP39]] to ptr addrspace(1) @@ -180,7 +180,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: [[TMP36:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP36:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP37:%.*]] = ptrtoint ptr [[TMP36]] to i64 ; CHECK-NEXT: [[TMP38:%.*]] = ptrtoint ptr addrspace(1) [[TMP35]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP38]], i64 [[TMP37]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test.ll index e0bfca0f63ca7..53a651e11bccc 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-dynamic-lds-test.ll @@ -46,7 +46,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: [[TMP32:%.*]] = add i32 [[TMP15]], [[TMP19]] ; CHECK-NEXT: [[TMP30:%.*]] = zext i32 [[TMP32]] 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: [[TMP39:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP30]], i64 [[TMP23]]) ; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP39]] to ptr addrspace(1) @@ -91,7 +91,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: [[TMP36:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP36:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP37:%.*]] = ptrtoint ptr [[TMP36]] to i64 ; CHECK-NEXT: [[TMP38:%.*]] = ptrtoint ptr addrspace(1) [[TMP35]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP38]], i64 [[TMP37]]) @@ -117,4 +117,5 @@ define amdgpu_kernel void @k0() sanitize_address { ;. ; CHECK: [[META0]] = !{i32 0, i32 1} ; CHECK: [[META1]] = !{i32 8, i32 9} +; CHECK: [[META2:![0-9]+]] = !{i32 4, !"nosanitize_address", i32 1} ;. diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-asan.ll index fa888a35cb8ba..9f2287bb54924 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-asan.ll @@ -86,7 +86,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP14:%.*]] = load i32, 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: [[TMP16:%.*]] = add i32 [[TMP13]], [[TMP14]] ; CHECK-NEXT: [[TMP15:%.*]] = zext i32 [[TMP16]] to i64 -; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP24:%.*]] = ptrtoint ptr [[TMP23]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP15]], i64 [[TMP24]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -196,7 +196,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: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param-asan.ll index a521d9d9d436b..6ca68376befb4 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param-asan.ll @@ -96,7 +96,7 @@ define amdgpu_kernel void @my_kernel() sanitize_address { ; CHECK-NEXT: [[TMP12:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_MY_KERNEL_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.my_kernel.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP14:%.*]] = add i32 [[TMP11]], [[TMP12]] ; CHECK-NEXT: [[TMP13:%.*]] = zext i32 [[TMP14]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP15:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP16:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP13]], i64 [[TMP15]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP16]] to ptr addrspace(1) @@ -121,7 +121,7 @@ define amdgpu_kernel void @my_kernel() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; CHECK-NEXT: [[TMP18:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP18:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP19:%.*]] = ptrtoint ptr [[TMP18]] to i64 ; CHECK-NEXT: [[TMP20:%.*]] = ptrtoint ptr addrspace(1) [[TMP17]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP20]], i64 [[TMP19]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param.ll index 55a36f85dc73a..07d4764a1852a 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-function-param.ll @@ -49,7 +49,7 @@ define amdgpu_kernel void @my_kernel() sanitize_address { ; CHECK-NEXT: [[TMP12:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_MY_KERNEL_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.my_kernel.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP14:%.*]] = add i32 [[TMP11]], [[TMP12]] ; CHECK-NEXT: [[TMP13:%.*]] = zext i32 [[TMP14]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP15:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP16:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP13]], i64 [[TMP15]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP16]] to ptr addrspace(1) @@ -74,7 +74,7 @@ define amdgpu_kernel void @my_kernel() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; CHECK-NEXT: [[TMP18:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP18:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP19:%.*]] = ptrtoint ptr [[TMP18]] to i64 ; CHECK-NEXT: [[TMP20:%.*]] = ptrtoint ptr addrspace(1) [[TMP17]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP20]], i64 [[TMP19]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-lower-all.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-lower-all.ll index 4625a7f626f9b..8acb379606ecc 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-lower-all.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-lower-all.ll @@ -62,7 +62,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, 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: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -103,7 +103,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: [[TMP34:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP34:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP35:%.*]] = ptrtoint ptr [[TMP34]] to i64 ; CHECK-NEXT: [[TMP36:%.*]] = ptrtoint ptr addrspace(1) [[TMP25]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP36]], i64 [[TMP35]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested-asan.ll index 255dda562c1ea..0e12aae72320d 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested-asan.ll @@ -6,7 +6,6 @@ @A = external addrspace(3) global [8 x ptr] @B = external addrspace(3) global [0 x i32] -;. ; @llvm.amdgcn.sw.lds.kernel_0 = internal addrspace(3) global ptr poison, no_sanitize_address, align 8, !absolute_symbol [[META0:![0-9]+]] ; @llvm.amdgcn.sw.lds.kernel_0.md = internal addrspace(1) global %llvm.amdgcn.sw.lds.kernel_0.md.type { %llvm.amdgcn.sw.lds.kernel_0.md.item { i32 0, i32 8, i32 32 }, %llvm.amdgcn.sw.lds.kernel_0.md.item { i32 32, i32 64, i32 96 } }, no_sanitize_address ; @llvm.amdgcn.sw.lds.kernel_2 = internal addrspace(3) global ptr poison, no_sanitize_address, align 8, !absolute_symbol [[META0]] @@ -36,7 +35,7 @@ define amdgpu_kernel void @kernel_0() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_0.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -58,7 +57,7 @@ define amdgpu_kernel void @kernel_0() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) @@ -96,7 +95,7 @@ define amdgpu_kernel void @kernel_1() sanitize_address { ; CHECK-NEXT: store i32 [[TMP14]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_1_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_1.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP15:%.*]] = add i32 [[TMP8]], [[TMP14]] ; CHECK-NEXT: [[TMP16:%.*]] = zext i32 [[TMP15]] to i64 -; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP18:%.*]] = ptrtoint ptr [[TMP17]] to i64 ; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP16]], i64 [[TMP18]]) ; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1) @@ -116,7 +115,7 @@ define amdgpu_kernel void @kernel_1() 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) [[TMP24]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP27]], i64 [[TMP26]]) @@ -144,7 +143,7 @@ define amdgpu_kernel void @kernel_2() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_2_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_2.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -166,7 +165,7 @@ define amdgpu_kernel void @kernel_2() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) @@ -204,7 +203,7 @@ define amdgpu_kernel void @kernel_3() sanitize_address { ; CHECK-NEXT: store i32 [[TMP14]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_3_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_3.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP15:%.*]] = add i32 [[TMP8]], [[TMP14]] ; CHECK-NEXT: [[TMP16:%.*]] = zext i32 [[TMP15]] to i64 -; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP18:%.*]] = ptrtoint ptr [[TMP17]] to i64 ; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP16]], i64 [[TMP18]]) ; CHECK-NEXT: [[TMP20:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1) @@ -224,7 +223,7 @@ define amdgpu_kernel void @kernel_3() 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) [[TMP24]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP27]], i64 [[TMP26]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested.ll index 7184ebbb8faa3..ada116e5bbda7 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-nested.ll @@ -36,7 +36,7 @@ define amdgpu_kernel void @kernel_0() sanitize_address { ; CHECK-NEXT: [[TMP10:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_0.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP12:%.*]] = add i32 [[TMP9]], [[TMP10]] ; CHECK-NEXT: [[TMP11:%.*]] = zext i32 [[TMP12]] to i64 -; CHECK-NEXT: [[TMP13:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP13:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP14:%.*]] = ptrtoint ptr [[TMP13]] to i64 ; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP11]], i64 [[TMP14]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1) @@ -58,7 +58,7 @@ define amdgpu_kernel void @kernel_0() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP16:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP16:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP17:%.*]] = ptrtoint ptr [[TMP16]] to i64 ; CHECK-NEXT: [[TMP18:%.*]] = ptrtoint ptr addrspace(1) [[TMP15]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP18]], i64 [[TMP17]]) @@ -96,7 +96,7 @@ define amdgpu_kernel void @kernel_1() sanitize_address { ; CHECK-NEXT: store i32 [[TMP11]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_1_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_1.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP15:%.*]] = add i32 [[TMP21]], [[TMP11]] ; CHECK-NEXT: [[TMP16:%.*]] = zext i32 [[TMP15]] to i64 -; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP18:%.*]] = ptrtoint ptr [[TMP17]] to i64 ; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP16]], i64 [[TMP18]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1) @@ -116,7 +116,7 @@ define amdgpu_kernel void @kernel_1() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP24:%.*]] = ptrtoint ptr [[TMP23]] to i64 ; CHECK-NEXT: [[TMP25:%.*]] = ptrtoint ptr addrspace(1) [[TMP22]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP25]], i64 [[TMP24]]) @@ -144,7 +144,7 @@ define amdgpu_kernel void @kernel_2() sanitize_address { ; CHECK-NEXT: [[TMP10:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_2_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_2.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP12:%.*]] = add i32 [[TMP9]], [[TMP10]] ; CHECK-NEXT: [[TMP11:%.*]] = zext i32 [[TMP12]] to i64 -; CHECK-NEXT: [[TMP13:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP13:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP14:%.*]] = ptrtoint ptr [[TMP13]] to i64 ; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP11]], i64 [[TMP14]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1) @@ -166,7 +166,7 @@ define amdgpu_kernel void @kernel_2() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP16:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP16:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP17:%.*]] = ptrtoint ptr [[TMP16]] to i64 ; CHECK-NEXT: [[TMP18:%.*]] = ptrtoint ptr addrspace(1) [[TMP15]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP18]], i64 [[TMP17]]) @@ -204,7 +204,7 @@ define amdgpu_kernel void @kernel_3() sanitize_address { ; CHECK-NEXT: store i32 [[TMP11]], ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_KERNEL_3_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.kernel_3.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP15:%.*]] = add i32 [[TMP21]], [[TMP11]] ; CHECK-NEXT: [[TMP16:%.*]] = zext i32 [[TMP15]] to i64 -; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP17:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP18:%.*]] = ptrtoint ptr [[TMP17]] to i64 ; CHECK-NEXT: [[TMP19:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP16]], i64 [[TMP18]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP19]] to ptr addrspace(1) @@ -224,7 +224,7 @@ define amdgpu_kernel void @kernel_3() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP24:%.*]] = ptrtoint ptr [[TMP23]] to i64 ; CHECK-NEXT: [[TMP25:%.*]] = ptrtoint ptr addrspace(1) [[TMP22]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP25]], i64 [[TMP24]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-no-kernel-lds-id.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-no-kernel-lds-id.ll index 704bc9e635294..13e4f28c4a88f 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-no-kernel-lds-id.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access-no-kernel-lds-id.ll @@ -63,7 +63,7 @@ define amdgpu_kernel void @k0() sanitize_address #1 { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, 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: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -104,7 +104,7 @@ define amdgpu_kernel void @k0() sanitize_address #1 { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP34:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP34:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP35:%.*]] = ptrtoint ptr [[TMP34]] to i64 ; CHECK-NEXT: [[TMP36:%.*]] = ptrtoint ptr addrspace(1) [[TMP25]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP36]], i64 [[TMP35]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access.ll index 8f5abe962f8eb..24bce7406e1b0 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-indirect-access.ll @@ -62,7 +62,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP14:%.*]] = load i32, 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: [[TMP16:%.*]] = add i32 [[TMP13]], [[TMP14]] ; CHECK-NEXT: [[TMP15:%.*]] = zext i32 [[TMP16]] to i64 -; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP24:%.*]] = ptrtoint ptr [[TMP23]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP15]], i64 [[TMP24]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -103,7 +103,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: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-O0.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-O0.ll index 1973a0acf4659..934ae1634d722 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-O0.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-O0.ll @@ -22,7 +22,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, 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: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -50,7 +50,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) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP27]], i64 [[TMP26]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-no-heap-ptr.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-no-heap-ptr.ll index ab3300ea659b8..eca24adad5258 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-no-heap-ptr.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-no-heap-ptr.ll @@ -26,7 +26,7 @@ define amdgpu_kernel void @k0() sanitize_address #1 { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -129,7 +129,7 @@ define amdgpu_kernel void @k0() sanitize_address #1 { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP78:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP78:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP79:%.*]] = ptrtoint ptr [[TMP78]] to i64 ; CHECK-NEXT: [[TMP80:%.*]] = ptrtoint ptr addrspace(1) [[TMP21]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP80]], i64 [[TMP79]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-asan.ll index c7550dd9576ec..aa54a96d3658c 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-asan.ll @@ -25,7 +25,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP14:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP16:%.*]] = add i32 [[TMP13]], [[TMP14]] ; CHECK-NEXT: [[TMP15:%.*]] = zext i32 [[TMP16]] to i64 -; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP23]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP15]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -128,7 +128,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: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomic-cmpxchg-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomic-cmpxchg-asan.ll index 15b074c2d9c11..fe3e677485c54 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomic-cmpxchg-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomic-cmpxchg-asan.ll @@ -23,7 +23,7 @@ define amdgpu_kernel void @atomic_xchg_kernel(ptr addrspace(1) %out, [8 x i32], ; CHECK-NEXT: [[TMP9:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_ATOMIC_XCHG_KERNEL_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.atomic_xchg_kernel.md, i32 0, i32 1, i32 2), align 4 ; CHECK-NEXT: [[TMP10:%.*]] = add i32 [[TMP8]], [[TMP9]] ; CHECK-NEXT: [[TMP11:%.*]] = zext i32 [[TMP10]] to i64 -; CHECK-NEXT: [[TMP12:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP12:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP13:%.*]] = ptrtoint ptr [[TMP12]] to i64 ; CHECK-NEXT: [[TMP14:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP11]], i64 [[TMP13]]) ; CHECK-NEXT: [[TMP15:%.*]] = inttoptr i64 [[TMP14]] to ptr addrspace(1) @@ -99,7 +99,7 @@ define amdgpu_kernel void @atomic_xchg_kernel(ptr addrspace(1) %out, [8 x i32], ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; CHECK-NEXT: [[TMP43:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP43:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP44:%.*]] = ptrtoint ptr [[TMP43]] to i64 ; CHECK-NEXT: [[TMP45:%.*]] = ptrtoint ptr addrspace(1) [[TMP21]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP45]], i64 [[TMP44]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomicrmw-asan.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomicrmw-asan.ll index 1b3664bf1e4e7..502b1cc9ec5ef 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomicrmw-asan.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test-atomicrmw-asan.ll @@ -25,7 +25,7 @@ define amdgpu_kernel void @atomicrmw_kernel(ptr addrspace(1) %arg0) sanitize_add ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_ATOMICRMW_KERNEL_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.atomicrmw_kernel.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -181,7 +181,7 @@ define amdgpu_kernel void @atomicrmw_kernel(ptr addrspace(1) %arg0) sanitize_add ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label [[FREE:%.*]], label [[END:%.*]] ; CHECK: Free: -; CHECK-NEXT: [[TMP84:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP84:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP85:%.*]] = ptrtoint ptr [[TMP84]] to i64 ; CHECK-NEXT: [[TMP86:%.*]] = ptrtoint ptr addrspace(1) [[TMP21]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP86]], i64 [[TMP85]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test.ll index 066b9429425ac..6c5f32f0a4e3a 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-test.ll @@ -26,7 +26,7 @@ define amdgpu_kernel void @k0() sanitize_address { ; CHECK-NEXT: [[TMP14:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_K0_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.k0.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP16:%.*]] = add i32 [[TMP13]], [[TMP14]] ; CHECK-NEXT: [[TMP15:%.*]] = zext i32 [[TMP16]] to i64 -; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP23:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP23]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP15]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP6:%.*]] = inttoptr i64 [[TMP12]] 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: [[TMP20:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP20:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP21:%.*]] = ptrtoint ptr [[TMP20]] to i64 ; CHECK-NEXT: [[TMP22:%.*]] = ptrtoint ptr addrspace(1) [[TMP19]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP22]], i64 [[TMP21]]) diff --git a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-vector-ptrs.ll b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-vector-ptrs.ll index 34caf91def933..7146965b1c66d 100644 --- a/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-vector-ptrs.ll +++ b/llvm/test/CodeGen/AMDGPU/amdgpu-sw-lower-lds-static-lds-vector-ptrs.ll @@ -25,7 +25,7 @@ define amdgpu_kernel void @example() sanitize_address { ; CHECK-NEXT: [[TMP7:%.*]] = load i32, ptr addrspace(1) getelementptr inbounds ([[LLVM_AMDGCN_SW_LDS_EXAMPLE_MD_TYPE]], ptr addrspace(1) @llvm.amdgcn.sw.lds.example.md, i32 0, i32 2, i32 2), align 4 ; CHECK-NEXT: [[TMP8:%.*]] = add i32 [[TMP6]], [[TMP7]] ; CHECK-NEXT: [[TMP9:%.*]] = zext i32 [[TMP8]] to i64 -; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP10:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP11:%.*]] = ptrtoint ptr [[TMP10]] to i64 ; CHECK-NEXT: [[TMP12:%.*]] = call i64 @__asan_malloc_impl(i64 [[TMP9]], i64 [[TMP11]]) ; CHECK-NEXT: [[TMP13:%.*]] = inttoptr i64 [[TMP12]] to ptr addrspace(1) @@ -62,7 +62,7 @@ define amdgpu_kernel void @example() sanitize_address { ; CHECK-NEXT: call void @llvm.amdgcn.s.barrier() ; CHECK-NEXT: br i1 [[XYZCOND]], label %[[FREE:.*]], label %[[END:.*]] ; CHECK: [[FREE]]: -; CHECK-NEXT: [[TMP33:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP33:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: [[TMP34:%.*]] = ptrtoint ptr [[TMP33]] to i64 ; CHECK-NEXT: [[TMP35:%.*]] = ptrtoint ptr addrspace(1) [[TMP20]] to i64 ; CHECK-NEXT: call void @__asan_free_impl(i64 [[TMP35]], i64 [[TMP34]]) diff --git a/llvm/test/Instrumentation/ThreadSanitizer/atomic-non-integer.ll b/llvm/test/Instrumentation/ThreadSanitizer/atomic-non-integer.ll index 015ee2fe711e1..0ca6465814090 100644 --- a/llvm/test/Instrumentation/ThreadSanitizer/atomic-non-integer.ll +++ b/llvm/test/Instrumentation/ThreadSanitizer/atomic-non-integer.ll @@ -6,7 +6,7 @@ target datalayout = "e-p:64:64:64-i1:8:8-i8:8:8-i16:16:16-i32:32:32-i64:64:64-f3 define float @load_float(ptr %fptr) { ; CHECK-LABEL: define float @load_float( ; CHECK-SAME: ptr [[FPTR:%.*]]) { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = call i32 @__tsan_atomic32_load(ptr [[FPTR]], i32 0) ; CHECK-NEXT: [[TMP3:%.*]] = bitcast i32 [[TMP2]] to float @@ -20,7 +20,7 @@ define float @load_float(ptr %fptr) { define double @load_double(ptr %fptr) { ; CHECK-LABEL: define double @load_double( ; CHECK-SAME: ptr [[FPTR:%.*]]) { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = call i64 @__tsan_atomic64_load(ptr [[FPTR]], i32 0) ; CHECK-NEXT: [[TMP3:%.*]] = bitcast i64 [[TMP2]] to double @@ -34,7 +34,7 @@ define double @load_double(ptr %fptr) { define fp128 @load_fp128(ptr %fptr) { ; CHECK-LABEL: define fp128 @load_fp128( ; CHECK-SAME: ptr [[FPTR:%.*]]) { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = call i128 @__tsan_atomic128_load(ptr [[FPTR]], i32 0) ; CHECK-NEXT: [[TMP3:%.*]] = bitcast i128 [[TMP2]] to fp128 @@ -48,7 +48,7 @@ define fp128 @load_fp128(ptr %fptr) { define void @store_float(ptr %fptr, float %v) { ; CHECK-LABEL: define void @store_float( ; CHECK-SAME: ptr [[FPTR:%.*]], float [[V:%.*]]) { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = bitcast float [[V]] to i32 ; CHECK-NEXT: call void @__tsan_atomic32_store(ptr [[FPTR]], i32 [[TMP2]], i32 0) @@ -62,7 +62,7 @@ define void @store_float(ptr %fptr, float %v) { define void @store_double(ptr %fptr, double %v) { ; CHECK-LABEL: define void @store_double( ; CHECK-SAME: ptr [[FPTR:%.*]], double [[V:%.*]]) { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = bitcast double [[V]] to i64 ; CHECK-NEXT: call void @__tsan_atomic64_store(ptr [[FPTR]], i64 [[TMP2]], i32 0) @@ -76,7 +76,7 @@ define void @store_double(ptr %fptr, double %v) { define void @store_fp128(ptr %fptr, fp128 %v) { ; CHECK-LABEL: define void @store_fp128( ; CHECK-SAME: ptr [[FPTR:%.*]], fp128 [[V:%.*]]) { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = bitcast fp128 [[V]] to i128 ; CHECK-NEXT: call void @__tsan_atomic128_store(ptr [[FPTR]], i128 [[TMP2]], i32 0) diff --git a/llvm/test/Instrumentation/ThreadSanitizer/eh.ll b/llvm/test/Instrumentation/ThreadSanitizer/eh.ll index 9c08b8f077e94..aec0f8b0aab11 100644 --- a/llvm/test/Instrumentation/ThreadSanitizer/eh.ll +++ b/llvm/test/Instrumentation/ThreadSanitizer/eh.ll @@ -10,7 +10,7 @@ declare void @cannot_throw() nounwind define i32 @func1() sanitize_thread { ; CHECK-EXC-LABEL: define i32 @func1 ; CHECK-EXC-SAME: () #[[ATTR1:[0-9]+]] personality ptr @__gcc_personality_v0 { -; CHECK-EXC-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-EXC-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-EXC-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-EXC-NEXT: invoke void @can_throw() ; CHECK-EXC-NEXT: to label [[DOTNOEXC:%.*]] unwind label [[TSAN_CLEANUP:%.*]] @@ -25,7 +25,7 @@ define i32 @func1() sanitize_thread { ; ; CHECK-NOEXC-LABEL: define i32 @func1 ; CHECK-NOEXC-SAME: () #[[ATTR1:[0-9]+]] { -; CHECK-NOEXC-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NOEXC-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NOEXC-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NOEXC-NEXT: call void @can_throw() ; CHECK-NOEXC-NEXT: call void @__tsan_func_exit() @@ -38,7 +38,7 @@ define i32 @func1() sanitize_thread { define i32 @func2() sanitize_thread { ; CHECK-LABEL: define i32 @func2 ; CHECK-SAME: () #[[ATTR1:[0-9]+]] { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: call void @cannot_throw() ; CHECK-NEXT: call void @__tsan_func_exit() @@ -51,7 +51,7 @@ define i32 @func2() sanitize_thread { define i32 @func3(ptr %p) sanitize_thread { ; CHECK-LABEL: define i32 @func3 ; CHECK-SAME: (ptr [[P:%.*]]) #[[ATTR1]] { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: call void @__tsan_read4(ptr [[P]]) ; CHECK-NEXT: [[A:%.*]] = load i32, ptr [[P]], align 4 @@ -65,7 +65,7 @@ define i32 @func3(ptr %p) sanitize_thread { define i32 @func4() sanitize_thread nounwind { ; CHECK-LABEL: define i32 @func4 ; CHECK-SAME: () #[[ATTR2:[0-9]+]] { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr [[TMP1]]) ; CHECK-NEXT: call void @can_throw() ; CHECK-NEXT: call void @__tsan_func_exit() diff --git a/llvm/test/Instrumentation/ThreadSanitizer/no_sanitize_thread.ll b/llvm/test/Instrumentation/ThreadSanitizer/no_sanitize_thread.ll index 965704c24bd8a..9edf0cac3afca 100644 --- a/llvm/test/Instrumentation/ThreadSanitizer/no_sanitize_thread.ll +++ b/llvm/test/Instrumentation/ThreadSanitizer/no_sanitize_thread.ll @@ -25,7 +25,7 @@ entry: ; CHECK: define i32 @read_4_bytes_and_call(ptr %a) { ; CHECK-NEXT: entry: -; CHECK-NEXT: %0 = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: %0 = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr %0) ; CHECK-NEXT: call void @foo() ; CHECK-NEXT: %tmp1 = load i32, ptr %a, align 4 diff --git a/llvm/test/Instrumentation/ThreadSanitizer/sanitize-thread-no-checking.ll b/llvm/test/Instrumentation/ThreadSanitizer/sanitize-thread-no-checking.ll index dbee198736569..254103dfd3689 100644 --- a/llvm/test/Instrumentation/ThreadSanitizer/sanitize-thread-no-checking.ll +++ b/llvm/test/Instrumentation/ThreadSanitizer/sanitize-thread-no-checking.ll @@ -25,7 +25,7 @@ entry: ; CHECK: define i32 @"\01-[WithCalls dealloc]"(ptr %a) ; CHECK-NEXT: entry: -; CHECK-NEXT: %0 = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: %0 = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__tsan_func_entry(ptr %0) ; CHECK-NEXT: call void @__tsan_ignore_thread_begin() ; CHECK-NEXT: %tmp1 = load i32, ptr %a, align 4 diff --git a/llvm/test/Transforms/EntryExitInstrumenter/debug-info.ll b/llvm/test/Transforms/EntryExitInstrumenter/debug-info.ll index 32a60dd6a4a00..b219d1999c8e8 100644 --- a/llvm/test/Transforms/EntryExitInstrumenter/debug-info.ll +++ b/llvm/test/Transforms/EntryExitInstrumenter/debug-info.ll @@ -10,10 +10,10 @@ entry: ret i32 42, !dbg !12 ; CHECK-LABEL: define i32 @f(i32 %x) -; CHECK: call ptr @llvm.returnaddress(i32 0), !dbg ![[ENTRYLOC:[0-9]+]] +; CHECK: call ptr @llvm.returnaddress.p0(i32 0), !dbg ![[ENTRYLOC:[0-9]+]] ; CHECK: call void @__cyg_profile_func_enter{{.*}}, !dbg ![[ENTRYLOC]] -; CHECK: call ptr @llvm.returnaddress(i32 0), !dbg ![[EXITLOC:[0-9]+]] +; CHECK: call ptr @llvm.returnaddress.p0(i32 0), !dbg ![[EXITLOC:[0-9]+]] ; CHECK: call void @__cyg_profile_func_exit{{.*}}, !dbg ![[EXITLOC]] ; CHECK: ret i32 42, !dbg ![[EXITLOC]] } diff --git a/llvm/test/Transforms/EntryExitInstrumenter/mcount-with-frompc.ll b/llvm/test/Transforms/EntryExitInstrumenter/mcount-with-frompc.ll index 0f8cf5c735453..f72c927c85bd4 100644 --- a/llvm/test/Transforms/EntryExitInstrumenter/mcount-with-frompc.ll +++ b/llvm/test/Transforms/EntryExitInstrumenter/mcount-with-frompc.ll @@ -9,7 +9,7 @@ define void @f1() "instrument-function-entry-inlined"="_mcount" { ; CHECK-LABEL: define void @f1() { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @_mcount(ptr [[TMP1]]) ; CHECK-NEXT: ret void ; diff --git a/llvm/test/Transforms/EntryExitInstrumenter/mcount.ll b/llvm/test/Transforms/EntryExitInstrumenter/mcount.ll index 56ccfb9ed2e7e..658bfd654f5a1 100644 --- a/llvm/test/Transforms/EntryExitInstrumenter/mcount.ll +++ b/llvm/test/Transforms/EntryExitInstrumenter/mcount.ll @@ -10,9 +10,9 @@ target triple = "powerpc64le-unknown-linux" define void @leaf_function() #0 { ; CHECK-LABEL: define void @leaf_function() { ; CHECK-NEXT: call void @mcount() -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_enter(ptr @leaf_function, ptr [[TMP1]]) -; CHECK-NEXT: [[TMP2:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP2:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_exit(ptr @leaf_function, ptr [[TMP2]]) ; CHECK-NEXT: ret void ; @@ -23,13 +23,13 @@ define void @leaf_function() #0 { define void @root_function() #0 { ; CHECK-LABEL: define void @root_function() { ; CHECK-NEXT: call void @mcount() -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_enter(ptr @root_function, ptr [[TMP1]]) -; CHECK-NEXT: [[TMP2:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP2:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_enter(ptr @leaf_function, ptr [[TMP2]]) -; CHECK-NEXT: [[TMP3:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP3:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_exit(ptr @leaf_function, ptr [[TMP3]]) -; CHECK-NEXT: [[TMP4:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP4:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_exit(ptr @root_function, ptr [[TMP4]]) ; CHECK-NEXT: ret void ; @@ -100,7 +100,7 @@ define void @f7() #7 { declare ptr @tailcallee() define ptr @tailcaller() #8 { ; CHECK-LABEL: define ptr @tailcaller() { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_exit(ptr @tailcaller, ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = musttail call ptr @tailcallee() ; CHECK-NEXT: ret ptr [[TMP2]] @@ -110,7 +110,7 @@ define ptr @tailcaller() #8 { } define ptr @tailcaller2() #8 { ; CHECK-LABEL: define ptr @tailcaller2() { -; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress(i32 0) +; CHECK-NEXT: [[TMP1:%.*]] = call ptr @llvm.returnaddress.p0(i32 0) ; CHECK-NEXT: call void @__cyg_profile_func_exit(ptr @tailcaller2, ptr [[TMP1]]) ; CHECK-NEXT: [[TMP2:%.*]] = musttail call ptr @tailcallee() ; CHECK-NEXT: ret ptr [[TMP2]] diff --git a/llvm/test/Transforms/EntryExitInstrumenter/pre-inliner-instrumentation.ll b/llvm/test/Transforms/EntryExitInstrumenter/pre-inliner-instrumentation.ll index 89e12a2971290..1ae096c48c4a1 100644 --- a/llvm/test/Transforms/EntryExitInstrumenter/pre-inliner-instrumentation.ll +++ b/llvm/test/Transforms/EntryExitInstrumenter/pre-inliner-instrumentation.ll @@ -15,7 +15,7 @@ define void @leaf_function() #0 { entry: ret void ; INSTRUMENT-LABEL: entry: -; INSTRUMENT-NEXT: %0 ={{.*}} call ptr @llvm.returnaddress(i32 0) +; INSTRUMENT-NEXT: %0 ={{.*}} call ptr @llvm.returnaddress.p0(i32 0) ; INSTRUMENT-NEXT: {{.* call void @__cyg_profile_func_enter\(ptr( nonnull)? @leaf_function, ptr %0\)}} ; NOINSTRUMENT-NOT: {{.*}} call void @__cyg_profile_func_enter ; INSTRUMENT: {{.*}} call void @__cyg_profile_func_exit @@ -31,7 +31,7 @@ entry: call void @leaf_function() ret void ; INSTRUMENT-LABEL: entry: -; INSTRUMENT-NEXT: %0 ={{.*}} call ptr @llvm.returnaddress(i32 0) +; INSTRUMENT-NEXT: %0 ={{.*}} call ptr @llvm.returnaddress.p0(i32 0) ; INSTRUMENT-NEXT: {{.*}} call void @__cyg_profile_func_enter(ptr{{( nonnull)?}} @root_function, ptr %0) ; INSTRUMENT: {{.*}} call void @__cyg_profile_func_enter ; INSTRUMENT: {{.*}} call void @__cyg_profile_func_exit diff --git a/llvm/test/Verifier/LoongArch/intrinsic-immarg.ll b/llvm/test/Verifier/LoongArch/intrinsic-immarg.ll index 488f77ff55ed4..b8f506477ab57 100644 --- a/llvm/test/Verifier/LoongArch/intrinsic-immarg.ll +++ b/llvm/test/Verifier/LoongArch/intrinsic-immarg.ll @@ -14,7 +14,7 @@ define ptr @non_const_depth_frameaddress(i32 %x) nounwind { define ptr @non_const_depth_returnaddress(i32 %x) nounwind { ; CHECK: immarg operand has non-immediate parameter ; CHECK-NEXT: i32 %x - ; CHECK-NEXT: %1 = call ptr @llvm.returnaddress(i32 %x) + ; CHECK-NEXT: %1 = call ptr @llvm.returnaddress.p0(i32 %x) %1 = call ptr @llvm.returnaddress(i32 %x) ret ptr %1 } diff --git a/llvm/test/Verifier/intrinsic-immarg.ll b/llvm/test/Verifier/intrinsic-immarg.ll index 6e68dde62afae..c95c7214fde71 100644 --- a/llvm/test/Verifier/intrinsic-immarg.ll +++ b/llvm/test/Verifier/intrinsic-immarg.ll @@ -4,7 +4,7 @@ declare ptr @llvm.returnaddress(i32) define void @return_address(i32 %var) { ; CHECK: immarg operand has non-immediate parameter ; CHECK-NEXT: i32 %var - ; CHECK-NEXT: %result = call ptr @llvm.returnaddress(i32 %var) + ; CHECK-NEXT: %result = call ptr @llvm.returnaddress.p0(i32 %var) %result = call ptr @llvm.returnaddress(i32 %var) ret void } _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
