[SYCL][CUDA] Link error will happen with '-g' compiling option

[XunZhangIntel]

@steffenlarsen After patch '#1773', '-g' option do now work correctly with cuda backend. This is a small reproducer sycl codes as below:

#include <CL/sycl.hpp>

int main(int argc, char* argv[]) {
cl::sycl::platform platform = cl::sycl::platform::get_platforms()[0];
cl::sycl::device device = platform.get_devices()[0];
cl::sycl::queue queue = cl::sycl::queue(device);
cl::sycl::buffer<uint64_t, 1> t1(10);
cl::sycl::range<1> global_range = {10};
queue.submit([&](cl::sycl::handler& cgh) {
auto table = t1.get_accesscl::sycl::access::mode::write(cgh);
cgh.parallel_for(global_range, [=](cl::sycl::id<1> global_id) {
int glob_tid = global_id[0];
table[glob_tid] = glob_tid;
});
});
queue.wait();
}

clang++ -g -fsycl -fsycl-targets=nvptx64-nvidia-cuda-sycldevice test.cpp -o test.run -lOpenCL

DISubprogram attached to more than one function !2569 = distinct !DISubprogram(name: "_ZTSZZ4mainENK3$_0clERN2cl4sycl7handlerEE11init_kernel", scope: !2570, file: !2570, line: 20, type: !2571, flags: DIFlagArtificial | DIFlagPrototyped, spFlags: DISPFlagDefinition | DISPFlagOptimized, unit: !0, retainedNodes: !2574) void (i64 addrspace(1)*, %"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"*, %"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"*, %"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"*, [3 x i32]*)* @"_ZTSZZ4mainENK3$_0clERN2cl4sycl7handlerEE11init_kernel_with_offset" DISubprogram attached to more than one function !2569 = distinct !DISubprogram(name: "_ZTSZZ4mainENK3$_0clERN2cl4sycl7handlerEE11init_kernel", scope: !2570, file: !2570, line: 20, type: !2571, flags: DIFlagArtificial | DIFlagPrototyped, spFlags: DISPFlagDefinition | DISPFlagOptimized, unit: !0, retainedNodes: !2574) void (i64 addrspace(1)*, %"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"*, %"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"*, %"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"*, [3 x i32]*)* @"_ZTSZZ4mainENK3$_0clERN2cl4sycl7handlerEE11init_kernel_with_offset" fatal error: error in backend: Broken module found, compilation aborted! clang-11: error: clang frontend command failed with exit code 70 (use -v to see invocation)

Please help check this issue.

[bader]

@XunZhangIntel, #2109 reverts #1773, which should resolve this regression. Right?

[XunZhangIntel]

@bader I need to verify it.

[XunZhangIntel]

@bader The issue still exist.

[bader]

Thanks. Most likely the problem was introduced before #1773.

[steffenlarsen]

Thank you for reporting this, @XunZhangIntel!

I am unable to reproduce this issue. Do you have a small reproducer? Looks to be init_kernel causing the problem if that helps isolating a reproducer.

[XunZhangIntel]

@steffenlarsen My small reproducer is provided in my comments above. And I used the compiler with the HEAD commit is 363ad5f. Can you try with it?

[steffenlarsen]

@steffenlarsen My small reproducer is provided in my comments above. And I used the compiler with the HEAD commit is 363ad5f. Can you try with it?

I completely missed that, sorry. Thank you, I will try to reproduce it locally.

[steffenlarsen]

I was able to reproduce the error and it should be fixed in #2129.