Debugging of SYCL kernel doesn't work

SYCL development
1

I am using VS2017 and playing around computecpp sample code of “Reduction”, If I set breakpoint inside “parallel_for”, when debugging, it doesn’t stop at the breakpoint. There is no problem with breakpoint set outside “paralle_for”, What is the reason? have I missed any setting of debugging SYCL kernel? Thanks!

2

Just understand that host_device should be chosen for debugging. I am new to learn SYCL.

3

Hi Henry,

I don’t have Windows (for VS) to replicate the behavior but on my Linux with gdb I can set breakpoints and variable watchers. For standard (C++ - like) debugging you have to be running on the Host with a Debug build.

Can I ask if you are running a Debug build of the computecpp-sdk sample(s) ?
If not, make sure to select Debug Mode in Visual Studio and do a re-build.

Also, if your are not running on Host device make sure to initialize the SYCL queue with host_selector.
Here’s how: cl::sycl::queue myQueue(cl::sycl::host_selector{});

With this setup you should definitely be able to hit the breakpoints you set inside the kernel code.

Here’s my attempt

My CMake command:

cmake ../ -DCMAKE_BUILD_TYPE=Debug -DComputeCpp_DIR=/path/to/computecpp

Running gdb:

gdb ./samples/reduction 
GNU gdb (Ubuntu 8.1-0ubuntu3) 8.1.0.20180409-git
[...]
(gdb) run
Starting program: /home/georgi/projects/computecpp-sdk/build/samples/reduction 
[Thread debugging using libthread_db enabled]
Using host libthread_db library "/lib/x86_64-linux-gnu/libthread_db.so.1".
 SYCL Sample code: 
   Reduction of an STL vector 
[New Thread 0x7ffff60f3700 (LWP 15771)]
[...]
ComputeCpp>  (Selected Platform: Host Mode Platform)
ComputeCpp>  (Selected Device: Host Mode Device)
 Device Name: Host Device
 Platform Name Host Platform
[New Thread 0x7ffff0add700 (LWP 15776)]
[...]
ComputeCpp>  (Runtime cannot wait for a null event )
[Thread 0x7ffff58f2700 (LWP 15772) exited]
[...]
SYCL Reduction result: 593
[Thread 0x7ffedd7fa700 (LWP 15783) exited]
 STL Reduction result: 593
[Thread 0x7ffff30bf700 (LWP 15774) exited]
[Thread 0x7ffff60f3700 (LWP 15771) exited]
[Inferior 1 (process 15767) exited normally]
(gdb) break 106
Breakpoint 1 at 0x55555555949e: file ../samples/reduction.cpp, line 106.
(gdb) run
Starting program: /home/georgi/projects/computecpp-sdk/build/samples/reduction 
[Thread debugging using libthread_db enabled]
Using host libthread_db library "/lib/x86_64-linux-gnu/libthread_db.so.1".
 SYCL Sample code: 
   Reduction of an STL vector 
[New Thread 0x7ffff60f3700 (LWP 15917)]
[...]
ComputeCpp>  (Selected Platform: Host Mode Platform)
ComputeCpp>  (Selected Device: Host Mode Device)
 Device Name: Host Device
 Platform Name Host Platform
[New Thread 0x7ffff0add700 (LWP 15922)]
[...]
[Switching to Thread 0x7ffedcff9700 (LWP 15930)]

Thread 14 "reduction" hit Breakpoint 1, int sycl_reduce<int>(std::vector<int, std::allocator<int> > const&)::{lambda(cl::sycl::handler&)#2}::operator()(cl::sycl::handler&)::{lambda(cl::sycl::nd_item<1>)#1}::operator()(cl::sycl::nd_item) const (__closure=0x7ffed8000b40, id=...) at ../samples/reduction.cpp:106
106	                  scratch[localid] = aI[globalid];
(gdb) _

I set a breakpoint on line 106 in the reduction kernel code where the scratch local accessor gets assigned and as you can see the debugger works as expected - breaking at this specified code location.

Hope this clarifies better how to debug SYCL kernel code on the host device.

Regards,
Georgi