[SYCL] error: variable-sized object may not be initialized

[cagnulein]

I'm facing a strange compiler issue. The source code are here https://github.com/cagnulein/sycl-benchmarks/blob/fix_2d_array_with_array/rotate/rotate_usm.cpp

Is it supposed to be ok? If yes, how can i use a dynamic multidimensional array with USM?

$ clang++ -O2 -g -std=c++11 -fsycl rotate_usm.cpp -lOpenCL -o rotate_usm.gpu
rotate_usm.cpp:142:100: error: variable-sized object may not be initialized
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                                                                                                   ^
rotate_usm.cpp:142:99: error: variable-sized object may not be initialized
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                                                                                                  ^
rotate_usm.cpp:142:99: note: in implicit move constructor for '(lambda at rotate_usm.cpp:142:99)' first required here
rotate_usm.cpp:142:99: error: variable-sized object may not be initialized
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                                                                                                  ^
/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:697:50: note: in implicit copy constructor for '(lambda at rotate_usm.cpp:142:99)' first required here
    kernel_parallel_for(KernelFunc);
                                                 ^
rotate_usm.cpp:142:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                ^
error: variable-sized object may not be initialized
rotate_usm.cpp:142:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
error: array initializer must be an initializer list
Stack dump:
0.	Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -emit-llvm-bc -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -faddrsig -o /tmp/rotate_usm-49bd83.o -x c++ rotate_usm.cpp 
1.	 parser at end of file
2.	/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:694:8: instantiating function definition 'cl::sycl::handler::parallel_for'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x55796d2d437a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x55796d2d2114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x55796d2d2252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f24eba49890]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang12InitListExprC2ERKNS_10ASTContextENS_14SourceLocationEN4llvm8ArrayRefIPNS_4ExprEEES4_+0xa8)[0x55796f563888]
/home/sysele/sycl/build/bin/clang-10(+0x367eff8)[0x55796f14bff8]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema21ConstructOpenCLKernelEPNS_12FunctionDeclERNS_13MangleContextE+0x10a2)[0x55796f14d902]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x899)[0x55796f281549]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema29InstantiateFunctionDefinitionENS_14SourceLocationEPNS_12FunctionDeclEbbb+0xbe1)[0x55796f282231]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x350)[0x55796f281000]
/home/sysele/sycl/build/bin/clang-10(+0x3185e7f)[0x55796ec52e7f]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema25ActOnEndOfTranslationUnitEv+0xb6)[0x55796ec53716]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang6Parser17ParseTopLevelDeclERNS_9OpaquePtrINS_12DeclGroupRefEEEb+0x37d)[0x55796eb3d03d]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x208)[0x55796eb30b98]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang13CodeGenAction13ExecuteActionEv+0x32)[0x55796df56eb2]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x55796d9e3949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x55796d9a45d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x55796daa9abb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x55796c42569b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x55796c3c2f82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f24ea6ddb97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x55796c422f4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
rotate_usm.cpp:142:100: error: variable-sized object may not be initialized
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                                                                                                   ^
rotate_usm.cpp:142:99: error: variable-sized object may not be initialized
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                                                                                                  ^
rotate_usm.cpp:142:99: note: in implicit move constructor for '(lambda at rotate_usm.cpp:142:99)' first required here
rotate_usm.cpp:142:99: error: variable-sized object may not be initialized
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                                                                                                  ^
/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:697:50: note: in implicit copy constructor for '(lambda at rotate_usm.cpp:142:99)' first required here
    kernel_parallel_for(KernelFunc);
                                                 ^
rotate_usm.cpp:142:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                ^
error: variable-sized object may not be initialized
rotate_usm.cpp:142:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
error: array initializer must be an initializer list
Stack dump:
0.	Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -fsyntax-only -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -fsycl-int-header=/tmp/rotate_usm-cbb883.h -faddrsig -o /tmp/rotate_usm-cbb883.h -x c++ rotate_usm.cpp 
1.	 parser at end of file
2.	/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:694:8: instantiating function definition 'cl::sycl::handler::parallel_for'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x5571175f837a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x5571175f6114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x5571175f6252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f9354e0f890]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang12InitListExprC2ERKNS_10ASTContextENS_14SourceLocationEN4llvm8ArrayRefIPNS_4ExprEEES4_+0xa8)[0x557119887888]
/home/sysele/sycl/build/bin/clang-10(+0x367eff8)[0x55711946fff8]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema21ConstructOpenCLKernelEPNS_12FunctionDeclERNS_13MangleContextE+0x10a2)[0x557119471902]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x899)[0x5571195a5549]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema29InstantiateFunctionDefinitionENS_14SourceLocationEPNS_12FunctionDeclEbbb+0xbe1)[0x5571195a6231]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x350)[0x5571195a5000]
/home/sysele/sycl/build/bin/clang-10(+0x3185e7f)[0x557118f76e7f]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema25ActOnEndOfTranslationUnitEv+0xb6)[0x557118f77716]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang6Parser17ParseTopLevelDeclERNS_9OpaquePtrINS_12DeclGroupRefEEEb+0x37d)[0x557118e6103d]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x208)[0x557118e54b98]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x557117d07949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x557117cc85d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x557117dcdabb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x55711674969b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x5571166e6f82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f9353aa3b97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x557116746f4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang version 10.0.0 (https://github.com/intel/llvm 9dd68c53358338e1a167be024406d790f5cb6139)
Target: x86_64-unknown-linux-gnu
Thread model: posix
InstalledDir: /home/sysele/sycl/build/bin
clang-10: note: diagnostic msg: PLEASE submit a bug report to https://bugs.llvm.org/ and include the crash backtrace, preprocessed source, and associated run script.
clang-10: note: diagnostic msg: Error generating preprocessed source(s).

[romanovvlad]

Hi,
it seems that it's impossible to capture variable length array by value.
The new_image variable cannot be captured in a standard C++ lambda like:

auto FF = [=]() {
   (void)new_image;
};

Compiler gives the same error message on such code.

I believe you need to initialize new_image using malloc_shared to use it in the kernel(preudo-code):

unsigned char** new_image = (unsigned char**)malloc_shared(DestBitmapHeight * sizeof(unsigned char*), myQueue.get_device(), myQueue.get_context());

However the crash after diagnostics is a bug.

[cagnulein]

@romanovvlad thanks. I've changed the code following your tip but i have a similar issue.

cagnulein/sycl-benchmarks@42d12d9

$ clang++ -O2 -g -std=c++11 -fsycl rotate_usm.cpp -lOpenCL -o rotate_usm.gpu
error: array initializer must be an initializer list
rotate_usm.cpp:140:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                ^
Stack dump:
0.	Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -emit-llvm-bc -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -faddrsig -o /tmp/rotate_usm-734076.o -x c++ rotate_usm.cpp 
1.	 parser at end of file
2.	/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:694:8: instantiating function definition 'cl::sycl::handler::parallel_for'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x55c017c4f37a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x55c017c4d114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x55c017c4d252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f0e7d2f1890]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang12InitListExprC2ERKNS_10ASTContextENS_14SourceLocationEN4llvm8ArrayRefIPNS_4ExprEEES4_+0xa8)[0x55c019ede888]
/home/sysele/sycl/build/bin/clang-10(+0x367eff8)[0x55c019ac6ff8]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema21ConstructOpenCLKernelEPNS_12FunctionDeclERNS_13MangleContextE+0x10a2)[0x55c019ac8902]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x899)[0x55c019bfc549]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema29InstantiateFunctionDefinitionENS_14SourceLocationEPNS_12FunctionDeclEbbb+0xbe1)[0x55c019bfd231]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x350)[0x55c019bfc000]
/home/sysele/sycl/build/bin/clang-10(+0x3185e7f)[0x55c0195cde7f]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema25ActOnEndOfTranslationUnitEv+0xb6)[0x55c0195ce716]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang6Parser17ParseTopLevelDeclERNS_9OpaquePtrINS_12DeclGroupRefEEEb+0x37d)[0x55c0194b803d]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x208)[0x55c0194abb98]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang13CodeGenAction13ExecuteActionEv+0x32)[0x55c0188d1eb2]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x55c01835e949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x55c01831f5d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x55c018424abb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x55c016da069b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x55c016d3df82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f0e7bf85b97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x55c016d9df4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
error: array initializer must be an initializer list
rotate_usm.cpp:140:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                ^
Stack dump:
0.	Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -fsyntax-only -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -fsycl-int-header=/tmp/rotate_usm-24364b.h -faddrsig -o /tmp/rotate_usm-24364b.h -x c++ rotate_usm.cpp 
1.	 parser at end of file
2.	/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:694:8: instantiating function definition 'cl::sycl::handler::parallel_for'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x557d14f2037a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x557d14f1e114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x557d14f1e252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f3d75fe0890]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang12InitListExprC2ERKNS_10ASTContextENS_14SourceLocationEN4llvm8ArrayRefIPNS_4ExprEEES4_+0xa8)[0x557d171af888]
/home/sysele/sycl/build/bin/clang-10(+0x367eff8)[0x557d16d97ff8]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema21ConstructOpenCLKernelEPNS_12FunctionDeclERNS_13MangleContextE+0x10a2)[0x557d16d99902]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x899)[0x557d16ecd549]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema29InstantiateFunctionDefinitionENS_14SourceLocationEPNS_12FunctionDeclEbbb+0xbe1)[0x557d16ece231]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x350)[0x557d16ecd000]
/home/sysele/sycl/build/bin/clang-10(+0x3185e7f)[0x557d1689ee7f]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema25ActOnEndOfTranslationUnitEv+0xb6)[0x557d1689f716]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang6Parser17ParseTopLevelDeclERNS_9OpaquePtrINS_12DeclGroupRefEEEb+0x37d)[0x557d1678903d]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x208)[0x557d1677cb98]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x557d1562f949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x557d155f05d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x557d156f5abb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x557d1407169b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x557d1400ef82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f3d74c74b97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x557d1406ef4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang version 10.0.0 (https://github.com/intel/llvm 9dd68c53358338e1a167be024406d790f5cb6139)
Target: x86_64-unknown-linux-gnu
Thread model: posix
InstalledDir: /home/sysele/sycl/build/bin
clang-10: note: diagnostic msg: PLEASE submit a bug report to https://bugs.llvm.org/ and include the crash backtrace, preprocessed source, and associated run script.
clang-10: note: diagnostic msg: Error generating preprocessed source(s).

[romanovvlad]

Hi,
old_image should be allocated in the same way. Attaching diff.
usm_allocation.diff.txt

[cagnulein]

@romanovvlad i've tested in small arrays and it works, but with huge arrays (20000*40000 bytes) it gives me a segmentation fault. Looking with gdb this is the backtrace:

Thread 1 "rotate_usm.gpu" received signal SIGSEGV, Segmentation fault.
0x00007ffff5e649cb in ?? () from /usr/local/lib/intel-opencl/libigdrcl.so
(gdb) bt
#0  0x00007ffff5e649cb in ?? () from /usr/local/lib/intel-opencl/libigdrcl.so
#1  0x00007ffff5e65162 in ?? () from /usr/local/lib/intel-opencl/libigdrcl.so
#2  0x00007ffff5e047a6 in ?? () from /usr/local/lib/intel-opencl/libigdrcl.so
#3  0x00007ffff702ccc9 in cl::sycl::detail::usm::USMDispatcher::setKernelArgMemPointer(_pi_kernel*, unsigned int, void const*) ()
   from /home/sysele/sycl/build/lib/libsycl.so
#4  0x00007ffff7019dde in cl::sycl::detail::ExecCGCommand::enqueueImp() () from /home/sysele/sycl/build/lib/libsycl.so
#5  0x00007ffff70117c1 in cl::sycl::detail::Command::enqueue(cl::sycl::detail::EnqueueResultT&, cl::sycl::detail::BlockingT) ()
   from /home/sysele/sycl/build/lib/libsycl.so
#6  0x00007ffff701bc43 in cl::sycl::detail::Scheduler::addCG(std::unique_ptr >, std::shared_ptr) () from /home/sysele/sycl/build/lib/libsycl.so
#7  0x0000000000406f12 in cl::sycl::handler::finalize (this=0x7fffffffd2e8)
    at /home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:422
#8  0x0000000000403aa7 in cl::sycl::detail::queue_impl::submit_impl (this=0x7b5830, cgf=..., self=...)
    at /home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/detail/queue_impl.hpp:240
#9  0x0000000000403853 in cl::sycl::detail::queue_impl::submit (this=0x7b5830, cgf=..., Python Exception  There is no member named _M_use_count.: 
self=)
    at /home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/detail/queue_impl.hpp:114
#10 0x000000000040357c in cl::sycl::queue::submit (this=0x7fffffffd9c8, cgf=...)
    at /home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/queue.hpp:87
#11 0x000000000040345e in main (argc=1, argv=0x7fffffffdad8) at rotate_usm.cpp:144

You can find the source code here: cagnulein/sycl-benchmarks@df43b12

[romanovvlad]

Hm sounds like an issue in compute runtime, suggest creating issue there https://github.com/intel/compute-runtime/issues

[keryell]

I have the feeling that you should use just a 2-D array instead of an array of array, since all the dimensions seem to be constexpr if you do not need this complex data structure.
Even on a CPU this double indirection is inefficient anyway.
But in this example, you could just use 2-D cl::sycl::buffer + accessors instead of USM and explicit memory management.
It gives you this 2-D array-like interface through the accessor for free!

[cagnulein]

Hi @keryell , i've already tried the implementation with the cl::sycl::buffer (you can find here https://github.com/cagnulein/sycl-benchmarks/blob/master/rotate/rotate.cpp) and it works; it has only big issue: is really slow copying 800MB of data from the CPU to GPU and back.

With the USM, if i didn't misundarstood them, i will run my alghoritm without copying around my buffer. And when it works, it works very quickly! I talk about 8x times faster than cl::sycl:buffer mode.

What do you think?

[keryell]

I see. It depends probably on the the GPU architecture. If you have some efficient shared hardware access between CPU and GPU that makes sense. Otherwise, a 2-D buffer with the right access mode in the accessors (such as write_only...) might be more efficient, with a minimal number of huge copy instead a gazillions of inefficient individual accesses,

But my 2-D array suggestion was at the first place that it seems super inefficient to have arrays of arrays like

    unsigned char** new_image = (unsigned char**)malloc_shared(DestBitmapWidth * sizeof(unsigned char*), myQueue.get_device(), myQueue.get_context());
    for(int i=0; i<DestBitmapHeight; i++)
        new_image[i] = (unsigned char*)malloc_shared(DestBitmapWidth * sizeof(unsigned char), myQueue.get_device(), myQueue.get_context());

instead of just a plain old C 2-D array:

auto constexpr DestBitmapWidth = 10;
auto constexpr DestBitmapHeight = 20;
using array_t = unsigned char[DestBitmapWidth][DestBitmapHeight];
auto &new_image = *static_cast<array_t *>(malloc_shared(sizeof(array_t), myQueue.get_device(), myQueue.get_context()));

See https://godbolt.org/z/XkAWqe to get a simpler idea about why it should work (I have not tried myself with oneAPI...)

I hope in the future USM will be extended to allow "just":

auto constexpr DestBitmapWidth = 10;
auto constexpr DestBitmapHeight = 20;
using array_t = unsigned char[DestBitmapWidth][DestBitmapHeight];
auto &new_image = *malloc_shared<array_t>(myQueue);

that would provide also some nice alignment properties as a side effect.

@jbrodman please :-)

[cagnulein]

@keryell thanks. I've tried your implementation here https://github.com/cagnulein/sycl-benchmarks/blob/7bcbe77a5551551e5355ce87adfe91aff99ac96b/rotate/rotate_usm.cpp but i have compilation issue:

$ clang++ -O2 -g -std=c++11 -fsycl rotate_usm.cpp -lOpenCL -o rotate_usm.gpu
error: array initializer must be an initializer list
rotate_usm.cpp:147:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                ^
Stack dump:
0.	Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -emit-llvm-bc -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -faddrsig -o /tmp/rotate_usm-c12fb6.o -x c++ rotate_usm.cpp 
1.	 parser at end of file
2.	/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:694:8: instantiating function definition 'cl::sycl::handler::parallel_for'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x556a1148d37a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x556a1148b114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x556a1148b252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f31a9521890]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang12InitListExprC2ERKNS_10ASTContextENS_14SourceLocationEN4llvm8ArrayRefIPNS_4ExprEEES4_+0xa8)[0x556a1371c888]
/home/sysele/sycl/build/bin/clang-10(+0x367eff8)[0x556a13304ff8]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema21ConstructOpenCLKernelEPNS_12FunctionDeclERNS_13MangleContextE+0x10a2)[0x556a13306902]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x899)[0x556a1343a549]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema29InstantiateFunctionDefinitionENS_14SourceLocationEPNS_12FunctionDeclEbbb+0xbe1)[0x556a1343b231]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x350)[0x556a1343a000]
/home/sysele/sycl/build/bin/clang-10(+0x3185e7f)[0x556a12e0be7f]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema25ActOnEndOfTranslationUnitEv+0xb6)[0x556a12e0c716]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang6Parser17ParseTopLevelDeclERNS_9OpaquePtrINS_12DeclGroupRefEEEb+0x37d)[0x556a12cf603d]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x208)[0x556a12ce9b98]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang13CodeGenAction13ExecuteActionEv+0x32)[0x556a1210feb2]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x556a11b9c949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x556a11b5d5d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x556a11c62abb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x556a105de69b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x556a1057bf82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f31a81b5b97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x556a105dbf4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
error: array initializer must be an initializer list
rotate_usm.cpp:147:17: note: in instantiation of function template specialization 'cl::sycl::handler::parallel_for' requested here
            cgh.parallel_for(sycl::range<2>(DestBitmapWidth,DestBitmapHeight), [=](sycl::id<2> idx) {
                ^
Stack dump:
0.	Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -fsyntax-only -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -fsycl-int-header=/tmp/rotate_usm-5b07ab.h -faddrsig -o /tmp/rotate_usm-5b07ab.h -x c++ rotate_usm.cpp 
1.	 parser at end of file
2.	/home/sysele/sycl/build/lib/clang/10.0.0/include/CL/sycl/handler.hpp:694:8: instantiating function definition 'cl::sycl::handler::parallel_for'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x55a7da30537a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x55a7da303114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x55a7da303252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f522dc4e890]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang12InitListExprC2ERKNS_10ASTContextENS_14SourceLocationEN4llvm8ArrayRefIPNS_4ExprEEES4_+0xa8)[0x55a7dc594888]
/home/sysele/sycl/build/bin/clang-10(+0x367eff8)[0x55a7dc17cff8]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema21ConstructOpenCLKernelEPNS_12FunctionDeclERNS_13MangleContextE+0x10a2)[0x55a7dc17e902]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x899)[0x55a7dc2b2549]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema29InstantiateFunctionDefinitionENS_14SourceLocationEPNS_12FunctionDeclEbbb+0xbe1)[0x55a7dc2b3231]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema28PerformPendingInstantiationsEb+0x350)[0x55a7dc2b2000]
/home/sysele/sycl/build/bin/clang-10(+0x3185e7f)[0x55a7dbc83e7f]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang4Sema25ActOnEndOfTranslationUnitEv+0xb6)[0x55a7dbc84716]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang6Parser17ParseTopLevelDeclERNS_9OpaquePtrINS_12DeclGroupRefEEEb+0x37d)[0x55a7dbb6e03d]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x208)[0x55a7dbb61b98]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x55a7daa14949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x55a7da9d55d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x55a7daadaabb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x55a7d945669b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x55a7d93f3f82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f522c8e2b97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x55a7d9453f4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang version 10.0.0 (https://github.com/intel/llvm 9dd68c53358338e1a167be024406d790f5cb6139)
Target: x86_64-unknown-linux-gnu
Thread model: posix
InstalledDir: /home/sysele/sycl/build/bin
clang-10: note: diagnostic msg: PLEASE submit a bug report to https://bugs.llvm.org/ and include the crash backtrace, preprocessed source, and associated run script.
clang-10: note: diagnostic msg: Error generating preprocessed source(s).

[keryell]

A limitation of the current compiler that might not handle references?
Can you try removing the reference by using instead for example

auto old_image = static_cast<array_t *>(malloc_shared(sizeof(array_t), myQueue.get_device(), myQueue.get_context()));

and replacing the old_image by (*old_image) where it is used?

[cagnulein]

@keryell i've tried but this is the result:

$ clang++ -O2 -g -std=c++11 -fsycl rotate_usm.cpp -lOpenCL -o rotate_usm.gpu
Stack dump:
0.      Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple spir64-unknown-unknown-sycldevice -fsycl-is-device -aux-triple x86_64-unknown-linux-gnu -std=c++11 -disable-llvm-passes -sycl-std=1.2.1 -emit-llvm-bc -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=all -fmath-errno -mconstructor-aliases -fuse-init-array -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -faddrsig -o /tmp/rotate_usm-6a600c.o -x c++ rotate_usm.cpp
1.       parser at end of file
2.      Per-module optimization passes
3.      Running pass 'Bitcode Writer' on module 'rotate_usm.cpp'.
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x561fd1cf537a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x561fd1cf3114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x561fd1cf3252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7fa71d4e8890]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm15ValueEnumeratorC2ERKNS_6ModuleEb+0x7d1)[0x561fd2af3f21]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm13BitcodeWriter11writeModuleERKNS_6ModuleEbPKNS_18ModuleSummaryIndexEbPSt5arrayIjLm5EE+0xaf)[0x561fd2ae2f9f]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm18WriteBitcodeToFileERKNS_6ModuleERNS_11raw_ostreamEbPKNS_18ModuleSummaryIndexEbPSt5arrayIjLm5EE+0xdf)[0x561fd2ae6e7f]
/home/sysele/sycl/build/bin/clang-10(+0x25f931c)[0x561fd2ae731c]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm6legacy15PassManagerImpl3runERNS_6ModuleE+0x3b1)[0x561fd175a791]
/home/sysele/sycl/build/bin/clang-10(+0x1a2e4b1)[0x561fd1f1c4b1]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang17EmitBackendOutputERNS_17DiagnosticsEngineERKNS_19HeaderSearchOptionsERKNS_14CodeGenOptionsERKNS_13TargetOptionsERKNS_11LangOptionsERKN4llvm10DataLayoutEPNSE_6ModuleENS_13BackendActionESt10unique_ptrINSE_17raw_pwrite_streamESt14default_deleteISM_EE+0x3cf)[0x561fd1f1df5f]
/home/sysele/sycl/build/bin/clang-10(+0x248c499)[0x561fd297a499]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x489)[0x561fd3551e19]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang13CodeGenAction13ExecuteActionEv+0x32)[0x561fd2977eb2]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x561fd2404949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x561fd23c55d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x561fd24caabb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x561fd0e4669b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x561fd0de3f82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7fa71c17cb97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x561fd0e43f4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
Stack dump:
0.      Program arguments: /home/sysele/sycl/build/bin/clang-10 -cc1 -triple x86_64-unknown-linux-gnu -sycl-std=1.2.1 -emit-obj -disable-free -disable-llvm-verifier -discard-value-names -main-file-name rotate_usm.cpp -mrelocation-model static -mthread-model posix -mframe-pointer=none -fmath-errno -masm-verbose -mconstructor-aliases -munwind-tables -fuse-init-array -target-cpu x86-64 -dwarf-column-info -debug-info-kind=limited -dwarf-version=4 -debugger-tuning=gdb -resource-dir /home/sysele/sycl/build/lib/clang/10.0.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/x86_64-linux-gnu/c++/7.4.0 -internal-isystem /usr/lib/gcc/x86_64-linux-gnu/7.4.0/../../../../include/c++/7.4.0/backward -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -internal-isystem /usr/local/include -internal-isystem /home/sysele/sycl/build/lib/clang/10.0.0/include -internal-externc-isystem /usr/include/x86_64-linux-gnu -internal-externc-isystem /include -internal-externc-isystem /usr/include -O2 -std=c++11 -fdeprecated-macro -fdebug-compilation-dir /home/sysele/work/sycl/rotate -ferror-limit 19 -fmessage-length 0 -fgnuc-version=4.2.1 -fobjc-runtime=gcc -fcxx-exceptions -fexceptions -fdiagnostics-show-option -fcolor-diagnostics -vectorize-loops -vectorize-slp -include /tmp/rotate_usm-a06107.h -dependency-filter /tmp/rotate_usm-a06107.h -fsycl-is-host -faddrsig -o /tmp/rotate_usm-d79ae0.o -x c++ rotate_usm.cpp
1.       parser at end of file
2.      Per-function optimization
3.      Running pass 'Early CSE' on function '@"_ZZZ4mainENK3$_0clERN2cl4sycl7handlerEENKUlNS1_2idILi2EEEE_clES5_"'
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys15PrintStackTraceERNS_11raw_ostreamE+0x2a)[0x5611951ea37a]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm3sys17RunSignalHandlersEv+0x34)[0x5611951e8114]
/home/sysele/sycl/build/bin/clang-10(+0x1805252)[0x5611951e8252]
/lib/x86_64-linux-gnu/libpthread.so.0(+0x12890)[0x7f1de49b0890]
/home/sysele/sycl/build/bin/clang-10(+0xc32d68)[0x561194615d68]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm19SimplifyInstructionEPNS_11InstructionERKNS_13SimplifyQueryEPNS_25OptimizationRemarkEmitterE+0x33f)[0x56119461ae4f]
/home/sysele/sycl/build/bin/clang-10(+0x15cbc16)[0x561194faec16]
/home/sysele/sycl/build/bin/clang-10(+0x15cf518)[0x561194fb2518]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm13FPPassManager13runOnFunctionERNS_8FunctionE+0x4d8)[0x561194c50438]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm6legacy23FunctionPassManagerImpl3runERNS_8FunctionE+0x3c)[0x561194c504cc]
/home/sysele/sycl/build/bin/clang-10(_ZN4llvm6legacy19FunctionPassManager3runERNS_8FunctionE+0x242)[0x561194c507b2]
/home/sysele/sycl/build/bin/clang-10(+0x1a2e7db)[0x5611954117db]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang17EmitBackendOutputERNS_17DiagnosticsEngineERKNS_19HeaderSearchOptionsERKNS_14CodeGenOptionsERKNS_13TargetOptionsERKNS_11LangOptionsERKN4llvm10DataLayoutEPNSE_6ModuleENS_13BackendActionESt10unique_ptrINSE_17raw_pwrite_streamESt14default_deleteISM_EE+0x3cf)[0x561195412f5f]
/home/sysele/sycl/build/bin/clang-10(+0x248c499)[0x561195e6f499]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang8ParseASTERNS_4SemaEbb+0x489)[0x561196a46e19]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang13CodeGenAction13ExecuteActionEv+0x32)[0x561195e6ceb2]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang14FrontendAction7ExecuteEv+0xb9)[0x5611958f9949]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang16CompilerInstance13ExecuteActionERNS_14FrontendActionE+0x1f5)[0x5611958ba5d5]
/home/sysele/sycl/build/bin/clang-10(_ZN5clang25ExecuteCompilerInvocationEPNS_16CompilerInstanceE+0xc9b)[0x5611959bfabb]
/home/sysele/sycl/build/bin/clang-10(_Z8cc1_mainN4llvm8ArrayRefIPKcEES2_Pv+0x6eb)[0x56119433b69b]
/home/sysele/sycl/build/bin/clang-10(main+0x22a2)[0x5611942d8f82]
/lib/x86_64-linux-gnu/libc.so.6(__libc_start_main+0xe7)[0x7f1de3644b97]
/home/sysele/sycl/build/bin/clang-10(_start+0x2a)[0x561194338f4a]
clang-10: error: unable to execute command: Segmentation fault (core dumped)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang-10: error: clang frontend command failed due to signal (use -v to see invocation)
clang version 10.0.0 (https://github.com/intel/llvm 9dd68c53358338e1a167be024406d790f5cb6139)
Target: x86_64-unknown-linux-gnu
Thread model: posix
InstalledDir: /home/sysele/sycl/build/bin
clang-10: note: diagnostic msg: PLEASE submit a bug report to https://bugs.llvm.org/ and include the crash backtrace, preprocessed source, and associated run script.
clang-10: note: diagnostic msg: Error generating preprocessed source(s).