| Issue |
91465
|
| Summary |
HIP attempts no sanity checking on hipLaunchKernel declaration and asserts
|
| Labels |
clang
|
| Assignees |
yxsamliu
|
| Reporter |
arsenm
|
Trying to write a minimal reproducer forces you to have a declaration of hipLaunchKernel, but this just asserts if you don't match the signature:
```
// RUN: clang++ -x hip -nogpuinc -nogpulib -S -emit-llvm --offload-arch=gfx90a %s
void hipLaunchKernel();
__attribute__((global)) void test() { }
```
```
1. <eof> parser at end of file
2. /Users/matt/Desktop/hip-assert.hip:5:30: LLVM IR generation of declaration 'test'
3. /Users/matt/Desktop/hip-assert.hip:5:30: Generating code for declaration 'test'
#0 0x000000010723fed4 llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (/usr/local/bin/clang-19+0x102373ed4)
#1 0x000000010723e3d0 llvm::sys::RunSignalHandlers() (/usr/local/bin/clang-19+0x1023723d0)
#2 0x000000010724055c SignalHandler(int) (/usr/local/bin/clang-19+0x10237455c)
#3 0x0000000185c21a24 (/usr/lib/system/libsystem_platform.dylib+0x18046da24)
#4 0x0000000185bf1cc0 (/usr/lib/system/libsystem_pthread.dylib+0x18043dcc0)
#5 0x0000000185afda40 (/usr/lib/system/libsystem_c.dylib+0x180349a40)
#6 0x0000000185afcd30 (/usr/lib/system/libsystem_c.dylib+0x180348d30)
#7 0x000000010ab15cac (anonymous namespace)::CGNVCUDARuntime::emitDeviceStub(clang::CodeGen::CodeGenFunction&, clang::CodeGen::FunctionArgList&) (.cold.35) (/usr/local/bin/clang-19+0x105c49cac)
#8 0x00000001076ecb34 (anonymous namespace)::CGNVCUDARuntime::emitDeviceStub(clang::CodeGen::CodeGenFunction&, clang::CodeGen::FunctionArgList&) (/usr/local/bin/clang-19+0x102820b34)
#9 0x00000001079299ec clang::CodeGen::CodeGenFunction::GenerateCode(clang::GlobalDecl, llvm::Function*, clang::CodeGen::CGFunctionInfo const&) (/usr/local/bin/clang-19+0x102a5d9ec)
#10 0x0000000107949690 clang::CodeGen::CodeGenModule::EmitGlobalFunctionDefinition(clang::GlobalDecl, llvm::GlobalValue*) (/usr/local/bin/clang-19+0x102a7d690)
#11 0x00000001079431a8 clang::CodeGen::CodeGenModule::EmitGlobalDefinition(clang::GlobalDecl, llvm::GlobalValue*) (/usr/local/bin/clang-19+0x102a771a8)
#12 0x0000000107946bbc clang::CodeGen::CodeGenModule::EmitGlobal(clang::GlobalDecl) (/usr/local/bin/clang-19+0x102a7abbc)
```
I'd expect this to just insert the declaration it needs if it doesn't find one already, but it needs to at least error that the signature is not expected
![]()
_______________________________________________
llvm-bugs mailing list
llvm-bugs@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-bugs
