Skip to content

Commit 9eef844

Browse files
Andrew SavonichevAlexeySachkov
Andrew Savonichev
authored and
AlexeySachkov
committed
Fix incorrect usage of StringRef (#607)
* Fix incorrect usage of StringRef `setName' function invalidates the name returned by `getName', so `InstName' is invalid. Since `takeName' is called later, we can avoid this renaming altogether. Signed-off-by: Andrew Savonichev <[email protected]>
1 parent 2f302ec commit 9eef844

File tree

5 files changed

+12
-16
lines changed

5 files changed

+12
-16
lines changed

‎llvm-spirv/lib/SPIRV/SPIRVUtil.cpp

+3-7
Original file line numberDiff line numberDiff line change
@@ -646,13 +646,9 @@ Instruction *mutateCallInst(
646646
auto Args = getArguments(CI);
647647
Type *RetTy = CI->getType();
648648
auto NewName = ArgMutate(CI, Args, RetTy);
649-
StringRef InstName;
650-
if (CI->hasName()) {
651-
InstName = CI->getName();
652-
CI->setName(InstName + ".old");
653-
}
654-
auto NewCI = addCallInst(M, NewName, RetTy, Args, Attrs, CI, Mangle,
655-
std::string(InstName) + ".tmp", TakeFuncName);
649+
StringRef InstName = CI->getName();
650+
auto NewCI = addCallInst(M, NewName, RetTy, Args, Attrs, CI, Mangle, InstName,
651+
TakeFuncName);
656652
auto NewI = RetMutate(NewCI);
657653
NewI->takeName(CI);
658654
NewI->setDebugLoc(CI->getDebugLoc());

‎llvm-spirv/test/transcoding/AtomicCompareExchangeExplicit_cl20.cl

+3-3
Original file line numberDiff line numberDiff line change
@@ -46,6 +46,6 @@ __kernel void testAtomicCompareExchangeExplicit_cl20(
4646
//CHECK-SPIRV: AtomicCompareExchangeWeak {{[0-9]+}} {{[0-9]+}} {{[0-9]+}} [[WorkgroupScope]] [[AcqRelMemSem]] [[RelaxedMemSem]]
4747

4848
//CHECK-LLVM: call spir_func i1 @_Z39atomic_compare_exchange_strong_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected1.as, i32 %desired, i32 3, i32 0, i32 2)
49-
//CHECK-LLVM: call spir_func i1 @_Z39atomic_compare_exchange_strong_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected2.as, i32 %desired, i32 4, i32 0, i32 1)
50-
//CHECK-LLVM: call spir_func i1 @_Z37atomic_compare_exchange_weak_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected3.as, i32 %desired, i32 3, i32 0, i32 2)
51-
//CHECK-LLVM: call spir_func i1 @_Z37atomic_compare_exchange_weak_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected4.as, i32 %desired, i32 4, i32 0, i32 1)
49+
//CHECK-LLVM: call spir_func i1 @_Z39atomic_compare_exchange_strong_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected3.as, i32 %desired, i32 4, i32 0, i32 1)
50+
//CHECK-LLVM: call spir_func i1 @_Z37atomic_compare_exchange_weak_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected5.as, i32 %desired, i32 3, i32 0, i32 2)
51+
//CHECK-LLVM: call spir_func i1 @_Z37atomic_compare_exchange_weak_explicitPU3AS4VU7_AtomiciPU3AS4ii12memory_orderS4_12memory_scope(i32 addrspace(4)* %0, i32 addrspace(4)* %expected7.as, i32 %desired, i32 4, i32 0, i32 1)

‎llvm-spirv/test/transcoding/GenericCastToPtr.cl

+3-3
Original file line numberDiff line numberDiff line change
@@ -37,7 +37,7 @@ private short2 *testGenericCastToPtrPrivate(generic short2 *a) {
3737
// CHECK-LLVM-LABEL: @testGenericCastToPtrExplicitGlobal
3838
// CHECK-LLVM: %[[VoidPtrCast:[0-9]+]] = bitcast <2 x i16> addrspace(4)* %a to i8 addrspace(4)*
3939
// CHECK-LLVM-NEXT: %[[AddrSpaceCast:[0-9]+]] = bitcast i8 addrspace(4)* %[[VoidPtrCast]] to i8 addrspace(4)*
40-
// CHECK-LLVM-NEXT: %.tmp = call spir_func i8 addrspace(1)* @__to_global(i8 addrspace(4)* %[[AddrSpaceCast]])
40+
// CHECK-LLVM-NEXT: %{{[0-9a-zA-Z.]+}} = call spir_func i8 addrspace(1)* @__to_global(i8 addrspace(4)* %[[AddrSpaceCast]])
4141
// CHECK-LLVM: bitcast i8 addrspace(1)* %{{[0-9]+}} to <2 x i16> addrspace(1)*
4242

4343
global short2 *testGenericCastToPtrExplicitGlobal(generic short2 *a) {
@@ -49,7 +49,7 @@ global short2 *testGenericCastToPtrExplicitGlobal(generic short2 *a) {
4949
// CHECK-LLVM-LABEL: @testGenericCastToPtrExplicitLocal
5050
// CHECK-LLVM: %[[VoidPtrCast:[0-9]+]] = bitcast <2 x i16> addrspace(4)* %a to i8 addrspace(4)*
5151
// CHECK-LLVM-NEXT: %[[AddrSpaceCast:[0-9]+]] = bitcast i8 addrspace(4)* %[[VoidPtrCast]] to i8 addrspace(4)*
52-
// CHECK-LLVM-NEXT: %.tmp = call spir_func i8 addrspace(3)* @__to_local(i8 addrspace(4)* %[[AddrSpaceCast]])
52+
// CHECK-LLVM-NEXT: %{{[0-9a-zA-Z.]+}} = call spir_func i8 addrspace(3)* @__to_local(i8 addrspace(4)* %[[AddrSpaceCast]])
5353
// CHECK-LLVM: bitcast i8 addrspace(3)* %{{[0-9]+}} to <2 x i16> addrspace(3)*
5454

5555
local short2 *testGenericCastToPtrExplicitLocal(generic short2 *a) {
@@ -61,7 +61,7 @@ local short2 *testGenericCastToPtrExplicitLocal(generic short2 *a) {
6161
// CHECK-LLVM-LABEL: @testGenericCastToPtrExplicitPrivate
6262
// CHECK-LLVM: %[[VoidPtrCast:[0-9]+]] = bitcast <2 x i16> addrspace(4)* %a to i8 addrspace(4)*
6363
// CHECK-LLVM-NEXT: %[[AddrSpaceCast:[0-9]+]] = bitcast i8 addrspace(4)* %[[VoidPtrCast]] to i8 addrspace(4)*
64-
// CHECK-LLVM-NEXT: %.tmp = call spir_func i8* @__to_private(i8 addrspace(4)* %[[AddrSpaceCast]])
64+
// CHECK-LLVM-NEXT: %{{[0-9a-zA-Z.]+}} = call spir_func i8* @__to_private(i8 addrspace(4)* %[[AddrSpaceCast]])
6565
// CHECK-LLVM: bitcast i8* %{{[0-9]+}} to <2 x i16>*
6666

6767
private short2 *testGenericCastToPtrExplicitPrivate(generic short2 *a) {

‎llvm-spirv/test/transcoding/atomics.spt

+2-2
Original file line numberDiff line numberDiff line change
@@ -70,8 +70,8 @@
7070
; CHECK-LLVM-12: call spir_func i32 @_Z11atomic_xchgPU3AS1Vii(i32 addrspace(1)* %dst, i32 1) [[attr]]
7171
; CHECK-LLVM-12: call spir_func i32 @_Z10atomic_addPU3AS1Vii(i32 addrspace(1)* %object, i32 0) [[attr]]
7272
; CHECK-LLVM-12: call spir_func i32 @_Z11atomic_xchgPU3AS1Vii(i32 addrspace(1)* %object, i32 %desired) [[attr]]
73-
; CHECK-LLVM-12: %.tmp = call spir_func i32 @_Z11atomic_xchgPU3AS1Vii(i32 addrspace(1)* %object, i32 1) [[attr]]
74-
; CHECK-LLVM-12: trunc i32 %.tmp to i1
73+
; CHECK-LLVM-12: %[[XCHG:[0-9a-zA-Z.]+]] = call spir_func i32 @_Z11atomic_xchgPU3AS1Vii(i32 addrspace(1)* %object, i32 1) [[attr]]
74+
; CHECK-LLVM-12: trunc i32 %[[XCHG]] to i1
7575
; CHECK-LLVM-12: call spir_func i32 @_Z11atomic_xchgPU3AS1Vii(i32 addrspace(1)* %object, i32 0) [[attr]]
7676

7777
; RUN: llvm-spirv -r %t1.spv -o %t2.bc --spirv-target-env="CL2.0"

‎llvm-spirv/test/transcoding/memory_access.ll

+1-1
Original file line numberDiff line numberDiff line change
@@ -28,7 +28,7 @@
2828
; CHECK-LLVM: load volatile i32 addrspace(4)*, i32 addrspace(4)** %ptr, align 8
2929
; CHECK-LLVM: load volatile i32 addrspace(4)*, i32 addrspace(4)** %ptr
3030
; CHECK-LLVM: load volatile i32 addrspace(4)*, i32 addrspace(4)** %ptr, align 8, !nontemporal ![[NTMetadata:[0-9]+]]
31-
; CHECK-LLVM: store i32 %call.old, i32 addrspace(4)* %arrayidx, align 4, !nontemporal ![[NTMetadata:[0-9]+]]
31+
; CHECK-LLVM: store i32 %call, i32 addrspace(4)* %arrayidx, align 4, !nontemporal ![[NTMetadata:[0-9]+]]
3232
; CHECK-LLVM: store i32 addrspace(4)* %5, i32 addrspace(4)** %ptr
3333
; CHECK-LLVM: ![[NTMetadata:[0-9]+]] = !{i32 1}
3434

0 commit comments

Comments
 (0)