llvm-spirv crashes on C++ for OpenCL using printf

[Calandracas606]

I am trying to offline compile an OpenCL program to SPIRV, but the compilation is crashing.

I can reproduce with llvm from two different systems:

• Ubuntu 22.04
  • llvm installed using: https://apt.llvm.org/
  • Ubuntu clang version 18.1.3 (++20240322073153+ef6d1ec07c69-1~exp1~20240322193300.86)
  • llvm-spirv built from: https://github.com/KhronosGroup/SPIRV-LLVM-Translator/tree/llvm_release_180
• llvm version 17.0.6 obtained from the void linux distribution

Both of these can reproduce the issue, but the attached logs are from the binaries in void linux

LOGS logs.tar.gz

Here is a simple reproducer:

clang -c -target spirv64 -o sample.spv kernel.clcpp

kernel.clcpp:

__kernel void my_kernel(const int number) {

  const int col = get_global_id(0);
  const int row = get_global_id(1);

  if (col == 0 && row == 0) {
    // if i remove the printf(...) statement  the compilation succeedss
    printf("number = %i\n", number);
  }

 // do more stuff ...

};

Notes:

• If I remove the printf(...) statement, the compilation succeeds
• Only happens when compiling as C++ for OpenCL
  • renaming kernel.clcpp -> kernel.cl and running clang -c -target spirv64 -o sample.spv kernel.cl succeeds
  • the simple test case doesn't use any C++ for OpenCL features, however the application I'm developing does need them
• I have much more complicated programs which use printf(...) and they compile successfully

PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace.
Stack dump:
0.	Program arguments: /usr/bin/llvm-spirv /tmp/kernel-073d5b.bc -o sample.spv
 #0 0x00007f79f9f6cb5e llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) ./../lib/Support/Unix/Signals.inc:602:22
 #1 0x00007f79f9f6a28b llvm::sys::RunSignalHandlers() ./../lib/Support/Signals.cpp:104:20
 #2 0x00007f79f9f6a28b SignalHandler ./../lib/Support/Unix/Signals.inc:403:31
 #3 0x00007f79f8f768c0 __restore_rt libc_sigaction.c:0:0
 #4 0x00007f7a01c95f4b llvm::CastInfo<llvm::TypedPointerType, llvm::Type*, void>::doCastIfPossible(llvm::Type* const&) /usr/include/llvm/Support/Casting.h:493:5
 #5 0x00007f7a01c95f4b decltype(auto) llvm::dyn_cast<llvm::TypedPointerType, llvm::Type>(llvm::Type*) /usr/include/llvm/Support/Casting.h:663:48
 #6 0x00007f7a01c95f4b SPIRV::BuiltinCallMutator::doConversion() ./../lib/SPIRV/SPIRVBuiltinHelper.cpp:94:35
 #7 0x00007f7a01c66d3f llvm::IRBuilder<llvm::ConstantFolder, llvm::IRBuilderDefaultInserter>::~IRBuilder() /usr/include/llvm/IR/IRBuilder.h:2595:7
 #8 0x00007f7a01c66d3f SPIRV::BuiltinCallMutator::~BuiltinCallMutator() ./../lib/SPIRV/SPIRVBuiltinHelper.h:101:3
 #9 0x00007f7a01c6d272 SPIRV::OCLToSPIRVBase::transBuiltin(llvm::CallInst*, OCLUtil::OCLBuiltinTransInfo&) ./../lib/SPIRV/OCLToSPIRV.cpp:929:1
#10 0x00007f7a01c6d9e5 std::_Function_base::~_Function_base() /usr/include/c++/13.2/bits/std_function.h:243:11
#11 0x00007f7a01c6d9e5 std::function<void (SPIRV::BuiltinCallMutator&)>::~function() /usr/include/c++/13.2/bits/std_function.h:334:11
#12 0x00007f7a01c6d9e5 OCLUtil::OCLBuiltinTransInfo::~OCLBuiltinTransInfo() ./../lib/SPIRV/OCLUtil.h:156:8
#13 0x00007f7a01c6d9e5 SPIRV::OCLToSPIRVBase::visitCallBuiltinSimple(llvm::CallInst*, llvm::StringRef, llvm::StringRef) ./../lib/SPIRV/OCLToSPIRV.cpp:1075:1
#14 0x00007f7a01c75588 SPIRV::OCLToSPIRVBase::visitCallInst(llvm::CallInst&) ./../lib/SPIRV/OCLToSPIRV.cpp:420:25
#15 0x00007f7a01c75986 void llvm::InstVisitor<SPIRV::OCLToSPIRVBase, void>::visit<llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::Instruction, false, false, void>, false, false>>(llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::Instruction, false, false, void>, false, false>, llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::Instruction, false, false, void>, false, false>) /usr/include/llvm/IR/InstVisitor.h:88:18
#16 0x00007f7a01c75986 llvm::InstVisitor<SPIRV::OCLToSPIRVBase, void>::visit(llvm::BasicBlock&) /usr/include/llvm/IR/InstVisitor.h:104:10
#17 0x00007f7a01c75986 void llvm::InstVisitor<SPIRV::OCLToSPIRVBase, void>::visit<llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::BasicBlock, false, false, void>, false, false>>(llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::BasicBlock, false, false, void>, false, false>, llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::BasicBlock, false, false, void>, false, false>) /usr/include/llvm/IR/InstVisitor.h:89:42
#18 0x00007f7a01c75986 llvm::InstVisitor<SPIRV::OCLToSPIRVBase, void>::visit(llvm::Function&) /usr/include/llvm/IR/InstVisitor.h:100:10
#19 0x00007f7a01c75986 void llvm::InstVisitor<SPIRV::OCLToSPIRVBase, void>::visit<llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::Function, false, false, void>, false, false>>(llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::Function, false, false, void>, false, false>, llvm::ilist_iterator<llvm::ilist_detail::node_options<llvm::Function, false, false, void>, false, false>) /usr/include/llvm/IR/InstVisitor.h:89:42
#20 0x00007f7a01c75986 llvm::InstVisitor<SPIRV::OCLToSPIRVBase, void>::visit(llvm::Module&) /usr/include/llvm/IR/InstVisitor.h:96:10
#21 0x00007f7a01c75986 SPIRV::OCLToSPIRVBase::runOCLToSPIRV(llvm::Module&) ./../lib/SPIRV/OCLToSPIRV.cpp:177:8
#22 0x00007f7a01c75a82 SPIRV::OCLToSPIRVPass::run(llvm::Module&, llvm::AnalysisManager<llvm::Module>&) ./../lib/SPIRV/OCLToSPIRV.cpp:113:27
#23 0x00007f7a01d352d1 llvm::detail::PassModel<llvm::Module, SPIRV::OCLToSPIRVPass, llvm::PreservedAnalyses, llvm::AnalysisManager<llvm::Module>>::run(llvm::Module&, llvm::AnalysisManager<llvm::Module>&) /usr/include/llvm/IR/PassManagerInternal.h:90:3
#24 0x00007f7a01d69631 llvm::PassManager<llvm::Module, llvm::AnalysisManager<llvm::Module>>::run(llvm::Module&, llvm::AnalysisManager<llvm::Module>&) /usr/include/llvm/IR/PassManager.h:521:20
#25 0x00007f7a01d69631 runSpirvWriterPasses ./../lib/SPIRV/SPIRVWriter.cpp:6000:14
#26 0x000055f5ef291e57 convertLLVMToSPIRV ./../tools/llvm-spirv/llvm-spirv.cpp:366:3
#27 0x000055f5ef291e57 main ./../tools/llvm-spirv/llvm-spirv.cpp:807:30
#28 0x00007f79f8f60c4c __libc_start_call_main ./csu/../sysdeps/nptl/libc_start_call_main.h:74:3
#29 0x00007f79f8f60d05 call_init ./csu/../csu/libc-start.c:128:20
#30 0x00007f79f8f60d05 __libc_start_main@GLIBC_2.2.5 ./csu/../csu/libc-start.c:347:5
#31 0x000055f5ef292141 _start /builddir/glibc-2.38/csu/../sysdeps/x86_64/start.S:117:0
clang: error: unable to execute command: Segmentation fault
clang: error: llvm-spirv command failed due to signal (use -v to see invocation)
clang version 17.0.6
Target: spirv64
Thread model: posix
InstalledDir: /usr/bin
clang: note: diagnostic msg: 
********************

PLEASE ATTACH THE FOLLOWING FILES TO THE BUG REPORT:
Preprocessed source(s) and associated run script(s) are located at:
clang: note: diagnostic msg: /tmp/kernel-027ba7.clcpp
clang: note: diagnostic msg: /tmp/kernel-027ba7.sh
clang: note: diagnostic msg: 

********************

[Calandracas606]

Looks to be a duplicate issue:
#68305
KhronosGroup/SPIRV-LLVM-Translator#2193