search

llvm / llvm-project

The LLVM Project is a collection of modular and reusable compiler and toolchain technologies.
http://llvm.org
Other
26.77k stars 10.97k forks source link

Crash asking me to report it here #96135

Open Line-fr opened 2 weeks ago

Line-fr commented 2 weeks ago

Hi, I have been trying to port my code to openMP recently and while it seemed to work fine for the first part of my code, for the second, I made it quite unhappy. Instead of getting a proper error allowing me to know what to do about it, it crashed giving me this: (the files do not have the exact same name but from a previous compilation without any modification

compilation command: clang++ main.cpp -fopenmp --offload-arch=gfx90a I use MI250x with rocm which worked well with openmp up to now

errormessage.txt

main-bd92e5.zip

I hope you will be able to solve this crash for the future nice error messages by my side, I am gonna try to solve it blindly Thank you

Line-fr commented 2 weeks ago

Update: even when I remove everything related to openmp in the function that is indicated in the crash report, it still crashs and apparently in the same function

llvmbot commented 2 weeks ago

@llvm/issue-subscribers-clang-codegen

Author: Line (Line-fr)

Hi, I have been trying to port my code to openMP recently and while it seemed to work fine for the first part of my code, for the second, I made it quite unhappy. Instead of getting a proper error allowing me to know what to do about it, it crashed giving me this: (the files do not have the exact same name but from a previous compilation without any modification compilation command: clang++ main.cpp -fopenmp --offload-arch=gfx90a I use MI250x with rocm which worked well with openmp up to now [errormessage.txt](https://github.com/user-attachments/files/15907597/errormessage.txt) [main-bd92e5.zip](https://github.com/user-attachments/files/15907599/main-bd92e5.zip) I hope you will be able to solve this crash for the future nice error messages by my side, I am gonna try to solve it blindly Thank you
EugeneZelenko commented 2 weeks ago

Could you please try 18 or main branch?

Line-fr commented 2 weeks ago

HI Thank you for your reply, I am gonna try recompiling it but I have never done such thing and I am not a sudo user in the server in question. I will keep you in touch of my progress for this compilation and using of another llvm version

Line-fr commented 2 weeks ago

Okay I succeeded and got the error again:

PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace, preprocessed source, and associated run script.
Stack dump:
0.      Program arguments: /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19 -cc1 -triple amdgcn-amd-amdhsa -aux-triple x86_64-unknown-linux-gnu -Werror=atomic-alignment -emit-llvm-bc -emit-llvm-uselists -dumpdir Kate.out- -disable-free -clear-ast-before-backend -disable-llvm-verifier -discard-value-names -main-file-name main.cpp -mrelocation-model pic -pic-level 2 -fhalf-no-semantic-interposition -mframe-pointer=all -ffp-contract=on -fno-rounding-math -mconstructor-aliases -fcuda-is-device -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/ocml.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_daz_opt_off.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_unsafe_math_off.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_finite_only_off.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_correctly_rounded_sqrt_on.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_wavefrontsize64_on.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_isa_version_90a.bc -mlink-builtin-bitcode /opt/rocm/amdgcn/bitcode/oclc_abi_version_400.bc -target-cpu gfx90a -debugger-tuning=gdb -fdebug-compilation-dir=/home/users/u0001492/llvmproject/llvm-project/build/bin -resource-dir /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/lib/clang/19 -internal-isystem /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/lib/clang/19/include/llvm_libc_wrappers -internal-isystem /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/lib/clang/19/include/openmp_wrappers -include __clang_openmp_device_functions.h -internal-isystem /usr/lib/gcc/x86_64-redhat-linux/11/../../../../include/c++/11 -internal-isystem /usr/lib/gcc/x86_64-redhat-linux/11/../../../../include/c++/11/x86_64-redhat-linux -internal-isystem /usr/lib/gcc/x86_64-redhat-linux/11/../../../../include/c++/11/backward -internal-isystem /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/lib/clang/19/include -internal-isystem /usr/local/include -internal-isystem /usr/lib/gcc/x86_64-redhat-linux/11/../../../../x86_64-redhat-linux/include -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/lib/clang/19/include -internal-isystem /usr/local/include -internal-isystem /usr/lib/gcc/x86_64-redhat-linux/11/../../../../x86_64-redhat-linux/include -internal-externc-isystem /include -internal-externc-isystem /usr/include -fdeprecated-macro -ferror-limit 19 -fvisibility=protected -fopenmp -fgnuc-version=4.2.1 -fskip-odr-check-in-gmf -fcxx-exceptions -fexceptions -fcolor-diagnostics -fopenmp-is-target-device -fopenmp-host-ir-file-path /tmp/main-d83634.bc -faddrsig -D__GCC_HAVE_DWARF2_CFI_ASM=1 -o /tmp/main-gfx90a-c538b9.bc -x c++ /home/users/u0001492/Kate/main.cpp
1.      /home/users/u0001492/Kate/BenchCircuits.hpp:7:1: current parser token 'Circuit'
2.      /home/users/u0001492/Kate/simulator.hpp:746:7: LLVM IR generation of declaration 'Simulator'
3.      /home/users/u0001492/Kate/simulator.hpp:946:13: LLVM IR generation of compound statement ('{}')
 #0 0x00000000035414d1 llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x35414d1)
 #1 0x000000000353ec3e SignalHandler(int) Signals.cpp:0:0
 #2 0x00007f0d1403e6f0 __restore_rt (/lib64/libc.so.6+0x3e6f0)
 #3 0x0000000003c10f74 clang::CodeGen::CodeGenFunction::EmitAggregateCopy(clang::CodeGen::LValue, clang::CodeGen::LValue, clang::QualType, clang::CodeGen::AggValueSlot::Overlap_t, bool) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3c10f74)
 #4 0x0000000003d43bca emitReductionListCopy((anonymous namespace)::CopyAction, clang::CodeGen::CodeGenFunction&, clang::QualType, llvm::ArrayRef<clang::Expr const*>, clang::CodeGen::Address, clang::CodeGen::Address, CopyOptionsTy) (.constprop.0.isra.0) CGOpenMPRuntimeGPU.cpp:0:0
 #5 0x0000000003d45ea9 emitShuffleAndReduceFunction(clang::CodeGen::CodeGenModule&, llvm::ArrayRef<clang::Expr const*>, clang::QualType, llvm::Function*, clang::SourceLocation) CGOpenMPRuntimeGPU.cpp:0:0
 #6 0x0000000003d4ef4a clang::CodeGen::CGOpenMPRuntimeGPU::emitReduction(clang::CodeGen::CodeGenFunction&, clang::SourceLocation, llvm::ArrayRef<clang::Expr const*>, llvm::ArrayRef<clang::Expr const*>, llvm::ArrayRef<clang::Expr const*>, llvm::ArrayRef<clang::Expr const*>, clang::CodeGen::CGOpenMPRuntime::ReductionOptionsTy) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3d4ef4a)
 #7 0x00000000038364d0 clang::CodeGen::CodeGenFunction::EmitOMPReductionClauseFinal(clang::OMPExecutableDirective const&, llvm::omp::Directive) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x38364d0)
 #8 0x0000000003866b1f clang::CodeGen::CodeGenFunction::EmitOMPWorksharingLoop(clang::OMPLoopDirective const&, clang::Expr*, llvm::function_ref<std::pair<clang::CodeGen::LValue, clang::CodeGen::LValue> (clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&)> const&, llvm::function_ref<std::pair<llvm::Value*, llvm::Value*> (clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, clang::CodeGen::Address, clang::CodeGen::Address)> const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3866b1f)
 #9 0x00000000038676f6 emitWorksharingDirective(clang::CodeGen::CodeGenFunction&, clang::OMPLoopDirective const&, bool) CGStmtOpenMP.cpp:0:0
#10 0x0000000003cea307 clang::CodeGen::RegionCodeGenTy::operator()(clang::CodeGen::CodeGenFunction&) const (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3cea307)
#11 0x0000000003cea403 (anonymous namespace)::CGOpenMPRegionInfo::EmitBody(clang::CodeGen::CodeGenFunction&, clang::Stmt const*) CGOpenMPRuntime.cpp:0:0
#12 0x00000000038771aa clang::CodeGen::CodeGenFunction::GenerateOpenMPCapturedStmtFunction(clang::CapturedStmt const&, clang::SourceLocation) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x38771aa)
#13 0x0000000003d1db33 emitParallelOrTeamsOutlinedFunction(clang::CodeGen::CodeGenModule&, clang::OMPExecutableDirective const&, clang::CapturedStmt const*, clang::VarDecl const*, llvm::omp::Directive, llvm::StringRef, clang::CodeGen::RegionCodeGenTy const&) CGOpenMPRuntime.cpp:0:0
#14 0x0000000003d1de26 clang::CodeGen::CGOpenMPRuntime::emitParallelOutlinedFunction(clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, clang::VarDecl const*, llvm::omp::Directive, clang::CodeGen::RegionCodeGenTy const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3d1de26)
#15 0x0000000003d52c73 clang::CodeGen::CGOpenMPRuntimeGPU::emitParallelOutlinedFunction(clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, clang::VarDecl const*, llvm::omp::Directive, clang::CodeGen::RegionCodeGenTy const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3d52c73)
#16 0x000000000385c677 emitCommonOMPParallelDirective(clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, llvm::omp::Directive, clang::CodeGen::RegionCodeGenTy const&, llvm::function_ref<void (clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, llvm::SmallVectorImpl<llvm::Value*>&)> const&) (.constprop.0) CGStmtOpenMP.cpp:0:0
#17 0x000000000386f5b7 clang::CodeGen::CodeGenFunction::EmitOMPParallelForDirective(clang::OMPParallelForDirective const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x386f5b7)
#18 0x000000000382e2f9 clang::CodeGen::CodeGenFunction::EmitCompoundStmtWithoutScope(clang::CompoundStmt const&, bool, clang::CodeGen::AggValueSlot) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x382e2f9)
#19 0x000000000382e67b clang::CodeGen::CodeGenFunction::EmitCompoundStmt(clang::CompoundStmt const&, bool, clang::CodeGen::AggValueSlot) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x382e67b)
#20 0x000000000382e92a clang::CodeGen::CodeGenFunction::EmitSimpleStmt(clang::Stmt const*, llvm::ArrayRef<clang::Attr const*>) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x382e92a)
#21 0x0000000003827762 clang::CodeGen::CodeGenFunction::EmitStmt(clang::Stmt const*, llvm::ArrayRef<clang::Attr const*>) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3827762)
#22 0x0000000003861968 emitTargetRegion(clang::CodeGen::CodeGenFunction&, clang::OMPTargetDirective const&, clang::CodeGen::PrePostActionTy&) CGStmtOpenMP.cpp:0:0
#23 0x0000000003cea203 clang::CodeGen::RegionCodeGenTy::operator()(clang::CodeGen::CodeGenFunction&) const (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3cea203)
#24 0x0000000003cea403 (anonymous namespace)::CGOpenMPRegionInfo::EmitBody(clang::CodeGen::CodeGenFunction&, clang::Stmt const*) CGOpenMPRuntime.cpp:0:0
#25 0x00000000038771aa clang::CodeGen::CodeGenFunction::GenerateOpenMPCapturedStmtFunction(clang::CapturedStmt const&, clang::SourceLocation) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x38771aa)
#26 0x0000000003d1d90e std::_Function_handler<llvm::Function* (llvm::StringRef), clang::CodeGen::CGOpenMPRuntime::emitTargetOutlinedFunctionHelper(clang::OMPExecutableDirective const&, llvm::StringRef, llvm::Function*&, llvm::Constant*&, bool, clang::CodeGen::RegionCodeGenTy const&)::'lambda'(llvm::StringRef)>::_M_invoke(std::_Any_data const&, llvm::StringRef&&) CGOpenMPRuntime.cpp:0:0
#27 0x0000000006c0fd86 llvm::OpenMPIRBuilder::emitTargetRegionFunction(llvm::TargetRegionEntryInfo&, std::function<llvm::Function* (llvm::StringRef)>&, bool, llvm::Function*&, llvm::Constant*&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x6c0fd86)
#28 0x0000000003cdf3e2 clang::CodeGen::CGOpenMPRuntime::emitTargetOutlinedFunctionHelper(clang::OMPExecutableDirective const&, llvm::StringRef, llvm::Function*&, llvm::Constant*&, bool, clang::CodeGen::RegionCodeGenTy const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3cdf3e2)
#29 0x0000000003d40326 clang::CodeGen::CGOpenMPRuntimeGPU::emitNonSPMDKernel(clang::OMPExecutableDirective const&, llvm::StringRef, llvm::Function*&, llvm::Constant*&, bool, clang::CodeGen::RegionCodeGenTy const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3d40326)
#30 0x0000000003839efc clang::CodeGen::CodeGenFunction::EmitOMPTargetDeviceFunction(clang::CodeGen::CodeGenModule&, llvm::StringRef, clang::OMPTargetDirective const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3839efc)
#31 0x0000000003ceefb4 clang::CodeGen::CGOpenMPRuntime::scanForTargetRegionsFunctions(clang::Stmt const*, llvm::StringRef) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3ceefb4)
#32 0x0000000003ceed2f clang::CodeGen::CGOpenMPRuntime::scanForTargetRegionsFunctions(clang::Stmt const*, llvm::StringRef) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3ceed2f)
#33 0x0000000003ceed2f clang::CodeGen::CGOpenMPRuntime::scanForTargetRegionsFunctions(clang::Stmt const*, llvm::StringRef) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3ceed2f)
#34 0x0000000003ceed2f clang::CodeGen::CGOpenMPRuntime::scanForTargetRegionsFunctions(clang::Stmt const*, llvm::StringRef) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3ceed2f)
#35 0x0000000003cef17a clang::CodeGen::CGOpenMPRuntime::emitTargetFunctions(clang::GlobalDecl) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3cef17a)
#36 0x00000000038d9e6b clang::CodeGen::CodeGenModule::EmitGlobal(clang::GlobalDecl) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x38d9e6b)
#37 0x00000000038e254f clang::CodeGen::CodeGenModule::EmitTopLevelDecl(clang::Decl*) (.part.0) CodeGenModule.cpp:0:0
#38 0x0000000003d699d1 (anonymous namespace)::CodeGeneratorImpl::HandleTopLevelDecl(clang::DeclGroupRef) ModuleBuilder.cpp:0:0
#39 0x0000000003d5b7e5 clang::BackendConsumer::HandleTopLevelDecl(clang::DeclGroupRef) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3d5b7e5)
#40 0x00000000055a6194 clang::ParseAST(clang::Sema&, bool, bool) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x55a6194)
#41 0x000000000400ca19 clang::FrontendAction::Execute() (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x400ca19)
#42 0x0000000003f8902b clang::CompilerInstance::ExecuteAction(clang::FrontendAction&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x3f8902b)
#43 0x00000000040cf853 clang::ExecuteCompilerInvocation(clang::CompilerInstance*) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0x40cf853)
#44 0x0000000000b665e1 cc1_main(llvm::ArrayRef<char const*>, char const*, void*) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0xb665e1)
#45 0x0000000000b60242 ExecuteCC1Tool(llvm::SmallVectorImpl<char const*>&, llvm::ToolContext const&) driver.cpp:0:0
#46 0x0000000000b621ee clang_main(int, char**, llvm::ToolContext const&) (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0xb621ee)
#47 0x0000000000a4b0f4 main (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0xa4b0f4)
#48 0x00007f0d14029590 __libc_start_call_main (/lib64/libc.so.6+0x29590)
#49 0x00007f0d14029640 __libc_start_main@GLIBC_2.2.5 (/lib64/libc.so.6+0x29640)
#50 0x0000000000b5fcd5 _start (/hs/work0/home/users/u0001492/llvmproject/llvm-project/build/bin/clang-19+0xb5fcd5)
clang++: error: unable to execute command: Segmentation fault (core dumped)
clang++: error: clang frontend command failed due to signal (use -v to see invocation)
clang version 19.0.0git (https://github.com/llvm/llvm-project.git 480a788e4946ac7b313291f26bca19aa65d649f1)
Target: x86_64-unknown-linux-gnu
Thread model: posix
InstalledDir: /hs/work0/home/users/u0001492/llvmproject/llvm-project/build/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/main-gfx90a-5d74f6.cpp
clang++: note: diagnostic msg: /tmp/main-8ceb21.cpp
clang++: note: diagnostic msg: /tmp/main-gfx90a-5d74f6.sh
clang++: note: diagnostic msg:

********************
Line-fr commented 2 weeks ago

I found why in my code! It was not in the function printed by the error. I don't know if it is supposed to make such a crash but

pragma omp parallel for reduction(+:allresults[0: *(2localqbits)**]) breaks it

pragma omp parallel for reduction(+:allresults[0:64]) works

This new functionnality of openMP 4.5 allowing to reduce arrays seems to break on non compiling time constant size array I wanted to test this and was expecting a clear error if it was not possible. In the end, it seems like it is not possible but the error was.. much worse than expected.

I hope this issue was not useless to you, sorry for the time

Endilll commented 2 weeks ago

I hope this issue was not useless to you, sorry for the time

It's absolutely not. We shouldn't be crashing on both good inputs and bad inputs, so there's something to fix here.

llvmbot commented 2 weeks ago

@llvm/issue-subscribers-openmp

Author: Line (Line-fr)

Hi, I have been trying to port my code to openMP recently and while it seemed to work fine for the first part of my code, for the second, I made it quite unhappy. Instead of getting a proper error allowing me to know what to do about it, it crashed giving me this: (the files do not have the exact same name but from a previous compilation without any modification compilation command: clang++ main.cpp -fopenmp --offload-arch=gfx90a I use MI250x with rocm which worked well with openmp up to now [errormessage.txt](https://github.com/user-attachments/files/15907597/errormessage.txt) [main-bd92e5.zip](https://github.com/user-attachments/files/15907599/main-bd92e5.zip) I hope you will be able to solve this crash for the future nice error messages by my side, I am gonna try to solve it blindly Thank you
jhuber6 commented 2 weeks ago

I found why in my code! It was not in the function printed by the error. I don't know if it is supposed to make such a crash but

pragma omp parallel for reduction(+:allresults[0: *(2localqbits)**]) breaks it

pragma omp parallel for reduction(+:allresults[0:64]) works

This new functionnality of openMP 4.5 allowing to reduce arrays seems to break on non compiling time constant size array I wanted to test this and was expecting a clear error if it was not possible. In the end, it seems like it is not possible but the error was.. much worse than expected.

I hope this issue was not useless to you, sorry for the time

The compiler should never crash like this. It would be very helpful if you could create a small reproducer and either post it here or through https://godbolt.org/.

Line-fr commented 2 weeks ago

Hi, it was not so easy to isolate exactly at first but by copying my code and removing progressively, I ended up reducing it drastically!

clang++ crashtry.cpp -fopenmp --offload-arch=gfx90a

#include<omp.h>
#include<iostream>

using namespace std;

int main(){

    int localqbits = 7;

    double allresults[64]; //allocating on gpu or not do not change anything, allresults just needs to be a pointer
    #pragma omp target device(0) is_device_ptr(allresults)
    {
        #pragma omp parallel for reduction(+:allresults[0:2*localqbits])
        for (int m = 0; m < 1; m++){
        }
    }
    return 0;
}

Note that if I add "const int localqbits" it stops crashing.

the bug is on the default clang of my system. Here is how I compiled main branch to test again:

cmake -S llvm -B build -G Ninja -DLLVM_ENABLE_PROJECTS="clang;lld" -DCMAKE_BUILD_TYPE=release -DLLVM_ENABLE_RUNTIMES="openmp" -DRUNTIMES_amdgcn-amd-amdhsa_LLVM_ENABLE_RUNTIMES=libc -DLLVM_RUNTIME_TARGETS="default;amdgcn-amd-amdhsa"

cmake --build build --target check-all

result of compilation checks

Testing Time: 1444.15s

Total Discovered Tests: 101228
  Skipped          :    23 (0.02%)
  Unsupported      :  2429 (2.40%)
  Passed           : 98592 (97.40%)
  Expectedly Failed:   184 (0.18%)

1 warning(s) in tests

It crashes on main branch too

I hope that it helped you and that I did a good enough job at reducing it!

Endilll commented 1 week ago

@Line-fr Thank you, you did a good job! I see the crash on Compiler Explorer (https://godbolt.org/z/64ffdvEeq) using both trunk and 18.1.0 with assertions enabled, but the crashes are different. Clang 19.0 crash:

Referring to an instruction in another function!
  %8 = getelementptr double, ptr %7, i64 %10, !dbg !320
fatal error: error in backend: Broken module found, compilation aborted!
clang++: error: clang frontend command failed with exit code 70 (use -v to see invocation)
Compiler returned: 70

Clang 18.1.0 crash:

clang-18: /root/llvm-project/clang/lib/CodeGen/CodeGenFunction.cpp:2244:
clang::CodeGen::CodeGenFunction::VlaSizePair clang::CodeGen::CodeGenFunction::getVLASize(const clang::VariableArrayType*):
Assertion `vlaSize && "no size for VLA!"' failed.

PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace, preprocessed source, and associated run script.
Stack dump:
0.  Program arguments: /opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18 -cc1 -triple amdgcn-amd-amdhsa -aux-triple x86_64-unknown-linux-gnu -emit-llvm-bc -emit-llvm-uselists -disable-free -clear-ast-before-backend -main-file-name example.cpp -mrelocation-model pic -pic-level 2 -fhalf-no-semantic-interposition -mframe-pointer=all -ffp-contract=on -fno-rounding-math -mconstructor-aliases -target-cpu gfx90a -fcuda-is-device -target-cpu gfx90a -debug-info-kind=constructor -dwarf-version=4 -debugger-tuning=gdb -fdebug-compilation-dir=/app -resource-dir /opt/compiler-explorer/clang-assertions-18.1.0/lib/clang/18 -internal-isystem /opt/compiler-explorer/clang-assertions-18.1.0/lib/clang/18/include/llvm_libc_wrappers -internal-isystem /opt/compiler-explorer/clang-assertions-18.1.0/lib/clang/18/include/openmp_wrappers -include __clang_openmp_device_functions.h -internal-isystem /opt/compiler-explorer/gcc-13.2.0/lib/gcc/x86_64-linux-gnu/13.2.0/../../../../include/c++/13.2.0 -internal-isystem /opt/compiler-explorer/gcc-13.2.0/lib/gcc/x86_64-linux-gnu/13.2.0/../../../../include/c++/13.2.0/x86_64-linux-gnu -internal-isystem /opt/compiler-explorer/gcc-13.2.0/lib/gcc/x86_64-linux-gnu/13.2.0/../../../../include/c++/13.2.0/backward -internal-isystem /opt/compiler-explorer/clang-assertions-18.1.0/lib/clang/18/include -internal-isystem /usr/local/include -internal-isystem /opt/compiler-explorer/gcc-13.2.0/lib/gcc/x86_64-linux-gnu/13.2.0/../../../../x86_64-linux-gnu/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /opt/compiler-explorer/clang-assertions-18.1.0/lib/clang/18/include -internal-isystem /usr/local/include -internal-isystem /opt/compiler-explorer/gcc-13.2.0/lib/gcc/x86_64-linux-gnu/13.2.0/../../../../x86_64-linux-gnu/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -fdeprecated-macro -ferror-limit 19 -fvisibility=protected -fopenmp -nogpulib -fgnuc-version=4.2.1 -fskip-odr-check-in-gmf -fcxx-exceptions -fexceptions -fcolor-diagnostics -mllvm --x86-asm-syntax=intel -fopenmp-is-target-device -fopenmp-host-ir-file-path /tmp/example-df8714.bc -faddrsig -D__GCC_HAVE_DWARF2_CFI_ASM=1 -o /tmp/example-gfx90a-246bf2.bc -x c++ <source>
1.  <eof> parser at end of file
2.  <source>:5:5: LLVM IR generation of declaration 'main'
3.  <source>:11:5: LLVM IR generation of compound statement ('{}')
 #0 0x00000000038d53d8 llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x38d53d8)
 #1 0x00000000038d2b2c SignalHandler(int) Signals.cpp:0:0
 #2 0x00007f2806e42520 (/lib/x86_64-linux-gnu/libc.so.6+0x42520)
 #3 0x00007f2806e969fc pthread_kill (/lib/x86_64-linux-gnu/libc.so.6+0x969fc)
 #4 0x00007f2806e42476 gsignal (/lib/x86_64-linux-gnu/libc.so.6+0x42476)
 #5 0x00007f2806e287f3 abort (/lib/x86_64-linux-gnu/libc.so.6+0x287f3)
 #6 0x00007f2806e2871b (/lib/x86_64-linux-gnu/libc.so.6+0x2871b)
 #7 0x00007f2806e39e96 (/lib/x86_64-linux-gnu/libc.so.6+0x39e96)
 #8 0x0000000003c5e818 clang::CodeGen::CodeGenFunction::getVLASize(clang::VariableArrayType const*) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c5e818)
 #9 0x0000000003c68d71 clang::CodeGen::CodeGenFunction::emitArrayLength(clang::ArrayType const*, clang::QualType&, clang::CodeGen::Address&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c68d71)
#10 0x0000000003ffd55d clang::CodeGen::CodeGenFunction::EmitAggregateCopy(clang::CodeGen::LValue, clang::CodeGen::LValue, clang::QualType, clang::CodeGen::AggValueSlot::Overlap_t, bool) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3ffd55d)
#11 0x0000000004135750 emitReductionListCopy((anonymous namespace)::CopyAction, clang::CodeGen::CodeGenFunction&, clang::QualType, llvm::ArrayRef<clang::Expr const*>, clang::CodeGen::Address, clang::CodeGen::Address, CopyOptionsTy) (.isra.0) CGOpenMPRuntimeGPU.cpp:0:0
#12 0x000000000413a9ec emitShuffleAndReduceFunction(clang::CodeGen::CodeGenModule&, llvm::ArrayRef<clang::Expr const*>, clang::QualType, llvm::Function*, clang::SourceLocation) CGOpenMPRuntimeGPU.cpp:0:0
#13 0x000000000414519d clang::CodeGen::CGOpenMPRuntimeGPU::emitReduction(clang::CodeGen::CodeGenFunction&, clang::SourceLocation, llvm::ArrayRef<clang::Expr const*>, llvm::ArrayRef<clang::Expr const*>, llvm::ArrayRef<clang::Expr const*>, llvm::ArrayRef<clang::Expr const*>, clang::CodeGen::CGOpenMPRuntime::ReductionOptionsTy) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x414519d)
#14 0x0000000003c0bbaa clang::CodeGen::CodeGenFunction::EmitOMPReductionClauseFinal(clang::OMPExecutableDirective const&, llvm::omp::Directive) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c0bbaa)
#15 0x0000000003c3774c clang::CodeGen::CodeGenFunction::EmitOMPWorksharingLoop(clang::OMPLoopDirective const&, clang::Expr*, llvm::function_ref<std::pair<clang::CodeGen::LValue, clang::CodeGen::LValue> (clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&)> const&, llvm::function_ref<std::pair<llvm::Value*, llvm::Value*> (clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, clang::CodeGen::Address, clang::CodeGen::Address)> const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c3774c)
#16 0x0000000003c38714 emitWorksharingDirective(clang::CodeGen::CodeGenFunction&, clang::OMPLoopDirective const&, bool) CGStmtOpenMP.cpp:0:0
#17 0x00000000040d6a1d clang::CodeGen::RegionCodeGenTy::operator()(clang::CodeGen::CodeGenFunction&) const (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x40d6a1d)
#18 0x00000000040d6d24 (anonymous namespace)::CGOpenMPRegionInfo::EmitBody(clang::CodeGen::CodeGenFunction&, clang::Stmt const*) CGOpenMPRuntime.cpp:0:0
#19 0x0000000003c4a623 clang::CodeGen::CodeGenFunction::GenerateOpenMPCapturedStmtFunction(clang::CapturedStmt const&, clang::SourceLocation) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c4a623)
#20 0x000000000410cbe8 emitParallelOrTeamsOutlinedFunction(clang::CodeGen::CodeGenModule&, clang::OMPExecutableDirective const&, clang::CapturedStmt const*, clang::VarDecl const*, llvm::omp::Directive, llvm::StringRef, clang::CodeGen::RegionCodeGenTy const&) CGOpenMPRuntime.cpp:0:0
#21 0x000000000410cfaf clang::CodeGen::CGOpenMPRuntime::emitParallelOutlinedFunction(clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, clang::VarDecl const*, llvm::omp::Directive, clang::CodeGen::RegionCodeGenTy const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x410cfaf)
#22 0x0000000004148333 clang::CodeGen::CGOpenMPRuntimeGPU::emitParallelOutlinedFunction(clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, clang::VarDecl const*, llvm::omp::Directive, clang::CodeGen::RegionCodeGenTy const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x4148333)
#23 0x0000000003c2b350 emitCommonOMPParallelDirective(clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, llvm::omp::Directive, clang::CodeGen::RegionCodeGenTy const&, llvm::function_ref<void (clang::CodeGen::CodeGenFunction&, clang::OMPExecutableDirective const&, llvm::SmallVectorImpl<llvm::Value*>&)> const&) CGStmtOpenMP.cpp:0:0
#24 0x0000000003c3e448 clang::CodeGen::CodeGenFunction::EmitOMPParallelForDirective(clang::OMPParallelForDirective const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c3e448)
#25 0x0000000003bf73ab clang::CodeGen::CodeGenFunction::EmitStmt(clang::Stmt const*, llvm::ArrayRef<clang::Attr const*>) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3bf73ab)
#26 0x0000000003bfda3c clang::CodeGen::CodeGenFunction::EmitCompoundStmtWithoutScope(clang::CompoundStmt const&, bool, clang::CodeGen::AggValueSlot) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3bfda3c)
#27 0x0000000003bfde24 clang::CodeGen::CodeGenFunction::EmitCompoundStmt(clang::CompoundStmt const&, bool, clang::CodeGen::AggValueSlot) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3bfde24)
#28 0x0000000003bfe0fc clang::CodeGen::CodeGenFunction::EmitSimpleStmt(clang::Stmt const*, llvm::ArrayRef<clang::Attr const*>) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3bfe0fc)
#29 0x0000000003bf6ce5 clang::CodeGen::CodeGenFunction::EmitStmt(clang::Stmt const*, llvm::ArrayRef<clang::Attr const*>) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3bf6ce5)
#30 0x0000000003c302db emitTargetRegion(clang::CodeGen::CodeGenFunction&, clang::OMPTargetDirective const&, clang::CodeGen::PrePostActionTy&) CGStmtOpenMP.cpp:0:0
#31 0x00000000040d69ea clang::CodeGen::RegionCodeGenTy::operator()(clang::CodeGen::CodeGenFunction&) const (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x40d69ea)
#32 0x00000000040d6d24 (anonymous namespace)::CGOpenMPRegionInfo::EmitBody(clang::CodeGen::CodeGenFunction&, clang::Stmt const*) CGOpenMPRuntime.cpp:0:0
#33 0x0000000003c4a623 clang::CodeGen::CodeGenFunction::GenerateOpenMPCapturedStmtFunction(clang::CapturedStmt const&, clang::SourceLocation) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c4a623)
#34 0x000000000410db02 std::_Function_handler<llvm::Function* (llvm::StringRef), clang::CodeGen::CGOpenMPRuntime::emitTargetOutlinedFunctionHelper(clang::OMPExecutableDirective const&, llvm::StringRef, llvm::Function*&, llvm::Constant*&, bool, clang::CodeGen::RegionCodeGenTy const&)::'lambda'(llvm::StringRef)>::_M_invoke(std::_Any_data const&, llvm::StringRef&&) CGOpenMPRuntime.cpp:0:0
#35 0x0000000007817fe9 llvm::OpenMPIRBuilder::emitTargetRegionFunction(llvm::TargetRegionEntryInfo&, std::function<llvm::Function* (llvm::StringRef)>&, bool, llvm::Function*&, llvm::Constant*&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x7817fe9)
#36 0x00000000040d2310 clang::CodeGen::CGOpenMPRuntime::emitTargetOutlinedFunctionHelper(clang::OMPExecutableDirective const&, llvm::StringRef, llvm::Function*&, llvm::Constant*&, bool, clang::CodeGen::RegionCodeGenTy const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x40d2310)
#37 0x000000000413e3d1 clang::CodeGen::CGOpenMPRuntimeGPU::emitTargetOutlinedFunction(clang::OMPExecutableDirective const&, llvm::StringRef, llvm::Function*&, llvm::Constant*&, bool, clang::CodeGen::RegionCodeGenTy const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x413e3d1)
#38 0x0000000003c07b35 clang::CodeGen::CodeGenFunction::EmitOMPTargetDeviceFunction(clang::CodeGen::CodeGenModule&, llvm::StringRef, clang::OMPTargetDirective const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3c07b35)
#39 0x00000000040ec77d clang::CodeGen::CGOpenMPRuntime::scanForTargetRegionsFunctions(clang::Stmt const*, llvm::StringRef) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x40ec77d)
#40 0x00000000040ec46e clang::CodeGen::CGOpenMPRuntime::scanForTargetRegionsFunctions(clang::Stmt const*, llvm::StringRef) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x40ec46e)
#41 0x00000000040ecb75 clang::CodeGen::CGOpenMPRuntime::emitTargetFunctions(clang::GlobalDecl) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x40ecb75)
#42 0x0000000003cce099 clang::CodeGen::CodeGenModule::EmitGlobal(clang::GlobalDecl) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x3cce099)
#43 0x0000000003cd8173 clang::CodeGen::CodeGenModule::EmitTopLevelDecl(clang::Decl*) (.part.0) CodeGenModule.cpp:0:0
#44 0x000000000415fd16 (anonymous namespace)::CodeGeneratorImpl::HandleTopLevelDecl(clang::DeclGroupRef) ModuleBuilder.cpp:0:0
#45 0x0000000004150de8 clang::BackendConsumer::HandleTopLevelDecl(clang::DeclGroupRef) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x4150de8)
#46 0x00000000060f02d4 clang::ParseAST(clang::Sema&, bool, bool) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x60f02d4)
#47 0x000000000415d928 clang::CodeGenAction::ExecuteAction() (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x415d928)
#48 0x00000000043cbfe9 clang::FrontendAction::Execute() (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x43cbfe9)
#49 0x000000000434ba9e clang::CompilerInstance::ExecuteAction(clang::FrontendAction&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x434ba9e)
#50 0x00000000044ab01e clang::ExecuteCompilerInvocation(clang::CompilerInstance*) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0x44ab01e)
#51 0x0000000000c10cd6 cc1_main(llvm::ArrayRef<char const*>, char const*, void*) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0xc10cd6)
#52 0x0000000000c084fa ExecuteCC1Tool(llvm::SmallVectorImpl<char const*>&, llvm::ToolContext const&) driver.cpp:0:0
#53 0x0000000000c0d844 clang_main(int, char**, llvm::ToolContext const&) (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0xc0d844)
#54 0x0000000000b05134 main (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0xb05134)
#55 0x00007f2806e29d90 (/lib/x86_64-linux-gnu/libc.so.6+0x29d90)
#56 0x00007f2806e29e40 __libc_start_main (/lib/x86_64-linux-gnu/libc.so.6+0x29e40)
#57 0x0000000000c07fde _start (/opt/compiler-explorer/clang-assertions-18.1.0/bin/clang-18+0xc07fde)
clang++: error: unable to execute command: Aborted (core dumped)
clang++: error: clang frontend command failed due to signal (use -v to see invocation)
Compiler returned: 254
jhuber6 commented 1 week ago

If you compile it with an NVPTX architecture you get a warning bout VLAs.

<source>:12:46: error: cannot generate code for reduction on array section, which requires a variable length array
   12 |         #pragma omp parallel for reduction(+:allresults[0:2*localqbits])
      |                                              ^
<source>:12:46: note: variable length arrays are not supported for the current target

I wonder why you don't need this in the AMDGPU version, considering we don't fully support VLAs there either.

jhuber6 commented 1 week ago

FWIW, this works if you make the localqbits constexpr, I'm going to assume this is a bug in VLA handling somewhere and it should probably just be rejected in sema if we don't support it.