[llvm-bugs] [Bug 50967] New: target amdgcn-amd-amdhsa: static_cast after atomic update leads to "Cannot select AtomicLoadAdd..."
via llvm-bugs
llvm-bugs at lists.llvm.org
Fri Jul 2 06:55:31 PDT 2021
https://bugs.llvm.org/show_bug.cgi?id=50967
Bug ID: 50967
Summary: target amdgcn-amd-amdhsa: static_cast after atomic
update leads to "Cannot select AtomicLoadAdd..."
Product: clang
Version: trunk
Hardware: PC
OS: Linux
Status: NEW
Severity: enhancement
Priority: P
Component: OpenMP
Assignee: unassignedclangbugs at nondot.org
Reporter: gcc.j.kelling at hzdr.de
CC: llvm-bugs at lists.llvm.org
The following code:
```c++
#include <iostream>
template<class T>
void f()
{
int b = 0;
#pragma omp target map(tofrom: b)
{
#pragma omp teams distribute
for(int i = 0; i < 1; ++i)
{
T a = 0;
#pragma omp parallel num_threads(64)
{
#pragma omp atomic update
a += 1;
}
b = static_cast<int>(a);
}
}
std::cout << b << std::endl;
}
int main()
{
// OK:
f<int>();
f<unsigned int>();
// ICE:
f<char>();
f<short>();
f<long long>();
f<float>();
f<double>();
}
```
Compiled with:
clang++ -fopenmp -fopenmp=libomp -fopenmp-targets=amdgcn-amd-amdhsa
-Xopenmp-target=amdgcn-amd-amdhsa -march=gfx900 anyToInt.cpp
yields ICE:
```
LLVM ERROR: Cannot select: t43: i16,ch = AtomicLoadAdd<(load store monotonic
(s8) on %ir.0)> t44:1, t45, Constant:i16<1>
t45: i64 = bitcast t44
t44: v2i32,ch = load<(dereferenceable load (s64) from %ir.a.addr.ascast)>
t47, t87, undef:i64
t87: i64 = bitcast t86
t86: v2i32 = BUILD_VECTOR t84, t85
t84: i32 = select t55, FrameIndex:i32<2>, t74
t55: i1 = setcc FrameIndex:i32<2>, Constant:i32<-1>, setne:ch
t26: i32 = FrameIndex<2>
t53: i32 = Constant<-1>
t26: i32 = FrameIndex<2>
t74: i32 = extract_vector_elt t73, Constant:i32<0>
t73: v2i32 = bitcast Constant:i64<0>
t52: i64 = Constant<0>
t71: i32 = Constant<0>
t85: i32 = select t55, t59, t76
t55: i1 = setcc FrameIndex:i32<2>, Constant:i32<-1>, setne:ch
t26: i32 = FrameIndex<2>
t53: i32 = Constant<-1>
t59: i32 = shl t57, TargetConstant:i32<16>
t57: i32 = S_GETREG_B32 TargetConstant:i16<30735>
t56: i16 = TargetConstant<30735>
t58: i32 = TargetConstant<16>
t76: i32 = extract_vector_elt t73, Constant:i32<1>
t73: v2i32 = bitcast Constant:i64<0>
t52: i64 = Constant<0>
t72: i32 = Constant<1>
t29: i64 = undef
t42: i16 = Constant<1>
In function: __omp_outlined__.5
PLEASE submit a bug report to https://bugs.llvm.org/ and include the crash
backtrace.
Stack dump:
0. Program arguments:
/home/kelling/checkout/llvm/llvm-project/install/bin/llc
/tmp/anyToInt-61b537-gfx900-linked-b24860.bc -mtriple=amdgcn-amd-amdhsa
-mcpu=gfx900 -filetype=obj -o /tmp/anyToInt-61b537-gfx900-0255a6.o
1. Running pass 'CallGraph Pass Manager' on module
'/tmp/anyToInt-61b537-gfx900-linked-b24860.bc'.
2. Running pass 'AMDGPU DAG->DAG Pattern Instruction Selection' on
function '@__omp_outlined__.5'
#0 0x00005561bebe8e4f PrintStackTraceSignalHandler(void*) Signals.cpp:0:0
#1 0x00005561bebe669d SignalHandler(int) Signals.cpp:0:0
#2 0x00007efe3bce4980 __restore_rt
(/lib/x86_64-linux-gnu/libpthread.so.0+0x12980)
#3 0x00007efe3a995fb7 raise (/lib/x86_64-linux-gnu/libc.so.6+0x3efb7)
#4 0x00007efe3a997921 abort (/lib/x86_64-linux-gnu/libc.so.6+0x40921)
#5 0x00005561beb43616 llvm::report_fatal_error(llvm::Twine const&, bool)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1d20616)
#6 0x00005561beb4377e
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1d2077e)
#7 0x00005561bea0a6ef llvm::SelectionDAGISel::CannotYetSelect(llvm::SDNode*)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1be76ef)
#8 0x00005561bea0e622 llvm::SelectionDAGISel::SelectCodeCommon(llvm::SDNode*,
unsigned char const*, unsigned int)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1beb622)
#9 0x00005561bd9f8c30 (anonymous
namespace)::AMDGPUDAGToDAGISel::Select(llvm::SDNode*)
AMDGPUISelDAGToDAG.cpp:0:0
#10 0x00005561bea0c0fa llvm::SelectionDAGISel::DoInstructionSelection()
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1be90fa)
#11 0x00005561bea14f2e llvm::SelectionDAGISel::CodeGenAndEmitDAG()
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1bf1f2e)
#12 0x00005561bea18573
llvm::SelectionDAGISel::SelectAllBasicBlocks(llvm::Function const&)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1bf5573)
#13 0x00005561bea1a2ab
llvm::SelectionDAGISel::runOnMachineFunction(llvm::MachineFunction&)
(.part.958) SelectionDAGISel.cpp:0:0
#14 0x00005561bd9ff9ca (anonymous
namespace)::AMDGPUDAGToDAGISel::runOnMachineFunction(llvm::MachineFunction&)
AMDGPUISelDAGToDAG.cpp:0:0
#15 0x00005561bdfc8a26
llvm::MachineFunctionPass::runOnFunction(llvm::Function&)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x11a5a26)
#16 0x00005561be4483f6 llvm::FPPassManager::runOnFunction(llvm::Function&)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x16253f6)
#17 0x00005561bdb87554 (anonymous
namespace)::CGPassManager::runOnModule(llvm::Module&) CallGraphSCCPass.cpp:0:0
#18 0x00005561be449560 llvm::legacy::PassManagerImpl::run(llvm::Module&)
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x1626560)
#19 0x00005561bd4bcda4 compileModule(char**, llvm::LLVMContext&) llc.cpp:0:0
#20 0x00005561bd43b756 main
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x618756)
#21 0x00007efe3a978bf7 __libc_start_main
(/lib/x86_64-linux-gnu/libc.so.6+0x21bf7)
#22 0x00005561bd4b42ba _start
(/home/kelling/checkout/llvm/llvm-project/install/bin/llc+0x6912ba)
clang-13: error: unable to execute command: Aborted (core dumped)
clang-13: error: amdgcn-link command failed due to signal (use -v to see
invocation)
clang version 13.0.0 (https://github.com/llvm/llvm-project.git
51b4ab26ca583b8d313da7663478392196ae6b6a)
Target: x86_64-unknown-linux-gnu
Thread model: posix
```
For any f<Type> where the cast form Type to int is not a bit-copy.
--
You are receiving this mail because:
You are on the CC list for the bug.
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://lists.llvm.org/pipermail/llvm-bugs/attachments/20210702/251cc227/attachment-0001.html>
More information about the llvm-bugs
mailing list