Execution fails when using both plugins

[fwyzard]

There seems to be some initialisation/finalisation issues when using multiple plugins.

If I try to run the same simple operation on both OpenCL and CUDA plugins, in any order, I get

 $ ./simple-sycl-app cpu cuda
Available SYCL platforms:
  - Intel(R) OpenCL, driver version OpenCL 2.1 LINUX
    - Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18

  - Intel(R) OpenCL HD Graphics, driver version OpenCL 2.1 
    - Intel(R) Gen9 HD Graphics NEO, driver version 20.08.15750

  - NVIDIA CUDA, driver version CUDA 10.20
    - Tesla K40c, driver version CUDA 10.20

  - SYCL host platform, driver version 1.2
    - SYCL host device, driver version 1.2

Running on SYCL device Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18
The results are correct!

Running on SYCL device Tesla K40c, driver version CUDA 10.20
Segmentation fault (core dumped)

The stack trace is

#0  0x00007ffff69f1248 in clCreateUserEvent (context=0x2004170, errcode_ret=0x7fffffffccfc) at /data/user/fwyzard/sycl/build-devel-4c3719d637b/tools/sycl/OpenCL/icd/loader/icd_dispatch.c:2216
#1  0x00007ffff63c1e8a in OclpiEventCreate (context=<optimized out>, ret_event=0x2327e10) at /data/user/fwyzard/sycl/llvm/sycl/plugins/opencl/pi_opencl.cpp:561
#2  0x00007ffff7206a93 in cl::sycl::detail::plugin::call_nocheck<(cl::sycl::detail::PiApiKind)44, _pi_context*, _pi_event**> (this=0x63e9d0) at /data/user/fwyzard/sycl/llvm/sycl/include/CL/sycl/detail/pi.def:69
#3  cl::sycl::detail::plugin::call<(cl::sycl::detail::PiApiKind)44, _pi_context*, _pi_event**> (this=0x63e9d0) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/plugin.hpp:63
#4  cl::sycl::detail::Command::prepareEvents (this=0x200bea0, Context=std::shared_ptr<class cl::sycl::detail::context_impl> (use count 9, weak count 0) = {...})
    at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/commands.cpp:125
#5  0x00007ffff720808e in cl::sycl::detail::MemCpyCommand::enqueueImp (this=0x200bea0) at /usr/include/c++/8/bits/shared_ptr.h:129
#6  0x00007ffff7205cfd in cl::sycl::detail::Command::enqueue (this=0x200bea0, EnqueueResult=..., Blocking=<optimized out>) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/commands.cpp:196
#7  0x00007ffff7210bc6 in cl::sycl::detail::Scheduler::GraphProcessor::enqueueCommand (Blocking=cl::sycl::detail::NON_BLOCKING, EnqueueResult=..., Cmd=0xa24a00)
    at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/graph_processor.cpp:65
#8  cl::sycl::detail::Scheduler::GraphProcessor::enqueueCommand (Cmd=Cmd@entry=0xa24a00, EnqueueResult=..., Blocking=Blocking@entry=cl::sycl::detail::NON_BLOCKING)
    at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/graph_processor.cpp:55
#9  0x00007ffff720c056 in cl::sycl::detail::Scheduler::addCG (this=0x7ffff74a4ea0 <cl::sycl::detail::Scheduler::instance>, CommandGroup=std::unique_ptr<class cl::sycl::detail::CG> = {...}, 
    Queue=std::shared_ptr<class cl::sycl::detail::queue_impl> (empty) = {...}) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/scheduler.cpp:72
#10 0x00007ffff7237499 in cl::sycl::handler::finalize (this=this@entry=0x7fffffffd250) at /usr/include/c++/8/bits/move.h:74
#11 0x00007ffff723dedf in cl::sycl::detail::queue_impl::submit_impl(std::function<void (cl::sycl::handler&)> const&, std::shared_ptr<cl::sycl::detail::queue_impl>) (this=0x20095c0, CGF=..., 
    Self=std::shared_ptr<class cl::sycl::detail::queue_impl> (empty) = {...}) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/queue_impl.hpp:355
#12 0x00007ffff723e92d in cl::sycl::detail::queue_impl::submit(std::function<void (cl::sycl::handler&)> const&, std::shared_ptr<cl::sycl::detail::queue_impl>) (Self=..., CGF=..., this=<optimized out>)
    at /usr/include/c++/8/bits/shared_ptr_base.h:754
#13 cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>) (this=<optimized out>, CGH=...) at /data/user/fwyzard/sycl/llvm/sycl/source/queue.cpp:110
#14 0x0000000000403966 in cl::sycl::queue::submit<main::$_0> (this=0x7fffffffd590, CGF=...) at /data/user/fwyzard/sycl/build-devel-4c3719d637b/lib64/clang/11.0.0/include/CL/sycl/queue.hpp:171
#15 main (argc=<optimized out>, argv=<optimized out>) at simple-sycl-app.cpp:118

And

 $ ./simple-sycl-app cuda cpu
Available SYCL platforms:
  - Intel(R) OpenCL, driver version OpenCL 2.1 LINUX
    - Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18

  - Intel(R) OpenCL HD Graphics, driver version OpenCL 2.1 
    - Intel(R) Gen9 HD Graphics NEO, driver version 20.08.15750

  - NVIDIA CUDA, driver version CUDA 10.20
    - Tesla K40c, driver version CUDA 10.20

  - SYCL host platform, driver version 1.2
    - SYCL host device, driver version 1.2

Running on SYCL device Tesla K40c, driver version CUDA 10.20
The results are correct!

Running on SYCL device Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18
pi_die: cuda_piEventSetCallback not implemented
terminate called without an active exception
Aborted (core dumped)

The stack trace is

#0  0x00007ffff6c2b8df in raise () from /lib64/libc.so.6
#1  0x00007ffff6c15cf5 in abort () from /lib64/libc.so.6
#2  0x00007ffff7acf06b in __gnu_cxx::__verbose_terminate_handler() [clone .cold.1] () from /lib64/libstdc++.so.6
#3  0x00007ffff7ad54cc in __cxxabiv1::__terminate(void (*)()) () from /lib64/libstdc++.so.6
#4  0x00007ffff7ad5527 in std::terminate() () from /lib64/libstdc++.so.6
#5  0x00007ffff718e94c in cl::sycl::detail::pi::die (Message=0x7ffff61b7810 "cuda_piEventSetCallback not implemented") at /data/user/fwyzard/sycl/llvm/sycl/source/detail/pi.cpp:220
#6  0x00007ffff61b2990 in cuda_piEventSetCallback (event=<optimized out>, command_exec_callback_type=<optimized out>, pfn_notify=<optimized out>, user_data=<optimized out>)
    at /data/user/fwyzard/sycl/llvm/sycl/plugins/cuda/pi_cuda.cpp:2381
#7  0x00007ffff7206b25 in cl::sycl::detail::plugin::call_nocheck<(cl::sycl::detail::PiApiKind)48, _pi_event*, int, void (*)(_pi_event*, int, void*), std::shared_ptr<cl::sycl::detail::event_impl>*> (this=0x790bb0)
    at /data/user/fwyzard/sycl/llvm/sycl/include/CL/sycl/detail/pi.def:73
#8  cl::sycl::detail::plugin::call<(cl::sycl::detail::PiApiKind)48, _pi_event*, int, void (*)(_pi_event*, int, void*), std::shared_ptr<cl::sycl::detail::event_impl>*> (this=0x790bb0)
    at /data/user/fwyzard/sycl/llvm/sycl/source/detail/plugin.hpp:63
#9  cl::sycl::detail::Command::prepareEvents (this=0x146e860, Context=std::shared_ptr<cl::sycl::detail::context_impl> (use count 9, weak count 0) = {...}) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/commands.cpp:129
#10 0x00007ffff720808e in cl::sycl::detail::MemCpyCommand::enqueueImp (this=0x146e860) at /usr/include/c++/8/bits/shared_ptr.h:129
#11 0x00007ffff7205cfd in cl::sycl::detail::Command::enqueue (this=0x146e860, EnqueueResult=..., Blocking=<optimized out>) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/commands.cpp:196
#12 0x00007ffff7210bc6 in cl::sycl::detail::Scheduler::GraphProcessor::enqueueCommand (Blocking=cl::sycl::detail::NON_BLOCKING, EnqueueResult=..., Cmd=0x1494f70)
    at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/graph_processor.cpp:65
#13 cl::sycl::detail::Scheduler::GraphProcessor::enqueueCommand (Cmd=Cmd@entry=0x1494f70, EnqueueResult=..., Blocking=Blocking@entry=cl::sycl::detail::NON_BLOCKING)
    at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/graph_processor.cpp:55
#14 0x00007ffff720c056 in cl::sycl::detail::Scheduler::addCG (this=0x7ffff74a4ea0 <cl::sycl::detail::Scheduler::instance>, CommandGroup=std::unique_ptr<cl::sycl::detail::CG> = {...}, 
    Queue=std::shared_ptr<cl::sycl::detail::queue_impl> (empty) = {...}) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/scheduler/scheduler.cpp:72
#15 0x00007ffff7237499 in cl::sycl::handler::finalize (this=this@entry=0x7fffffffd250) at /usr/include/c++/8/bits/move.h:74
#16 0x00007ffff723dedf in cl::sycl::detail::queue_impl::submit_impl(std::function<void (cl::sycl::handler&)> const&, std::shared_ptr<cl::sycl::detail::queue_impl>) (this=0x9fde00, CGF=..., 
    Self=std::shared_ptr<cl::sycl::detail::queue_impl> (empty) = {...}) at /data/user/fwyzard/sycl/llvm/sycl/source/detail/queue_impl.hpp:355
#17 0x00007ffff723e92d in cl::sycl::detail::queue_impl::submit(std::function<void (cl::sycl::handler&)> const&, std::shared_ptr<cl::sycl::detail::queue_impl>) (Self=..., CGF=..., this=<optimized out>)
    at /usr/include/c++/8/bits/shared_ptr_base.h:754
#18 cl::sycl::queue::submit_impl(std::function<void (cl::sycl::handler&)>) (this=<optimized out>, CGH=...) at /data/user/fwyzard/sycl/llvm/sycl/source/queue.cpp:110
#19 0x0000000000403966 in cl::sycl::queue::submit<main::$_0> (this=0x7fffffffd590, CGF=...) at /data/user/fwyzard/sycl/build-devel-4c3719d637b/lib64/clang/11.0.0/include/CL/sycl/queue.hpp:171
#20 main (argc=<optimized out>, argv=<optimized out>) at simple-sycl-app.cpp:118

[fwyzard]

From the stack trace, the problem seems to be that the wrong plugin is being used:

Running on SYCL device Tesla K40c, driver version CUDA 10.20
The results are correct!

Running on SYCL device Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18
...
#5  0x00007ffff718e94c in cl::sycl::detail::pi::die (Message=0x7ffff61b7810 "cuda_piEventSetCallback not implemented") at /data/user/fwyzard/sycl/llvm/sycl/source/detail/pi.cpp:220
#6  0x00007ffff61b2990 in cuda_piEventSetCallback (event=<optimized out>, command_exec_callback_type=<optimized out>, pfn_notify=<optimized out>, user_data=<optimized out>)
    at /data/user/fwyzard/sycl/llvm/sycl/plugins/cuda/pi_cuda.cpp:2381
...

so the runtime is trying to use the CUDA plugin for the OpenCL CPU device.

And

Running on SYCL device Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18
The results are correct!

Running on SYCL device Tesla K40c, driver version CUDA 10.20
...
#0  0x00007ffff69f1248 in clCreateUserEvent (context=0x2004170, errcode_ret=0x7fffffffccfc) at /data/user/fwyzard/sycl/build-devel-4c3719d637b/tools/sycl/OpenCL/icd/loader/icd_dispatch.c:2216
#1  0x00007ffff63c1e8a in OclpiEventCreate (context=<optimized out>, ret_event=0x2327e10) at /data/user/fwyzard/sycl/llvm/sycl/plugins/opencl/pi_opencl.cpp:561
...

so the runtime is trying to use the OpenCL plugin for the NVIDIA GPU.

[fwyzard]

It does seem to work if I use the host device between the OpenCL and CUDA devices:

$ ./simple-sycl-app cpu host cuda
Available SYCL platforms:
  - Intel(R) OpenCL, driver version OpenCL 2.1 LINUX
    - Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18

  - Intel(R) OpenCL HD Graphics, driver version OpenCL 2.1 
    - Intel(R) Gen9 HD Graphics NEO, driver version 20.08.15750

  - NVIDIA CUDA, driver version CUDA 10.20
    - Tesla K40c, driver version CUDA 10.20

  - SYCL host platform, driver version 1.2
    - SYCL host device, driver version 1.2

Running on SYCL device Intel(R) Core(TM) i9-9900K CPU @ 3.60GHz, driver version 2020.9.1.0.18
The results are correct!

Running on SYCL device SYCL host device, driver version 1.2
The results are correct!

Running on SYCL device Tesla K40c, driver version CUDA 10.20
The results are correct!

[fwyzard]

A minimal test case to reproduce the crashes is

#include <CL/sycl.hpp>
#include <iostream>

int main(int argc, char **argv) {

  // create a buffer of 4 ints to be used inside the kernel code
  cl::sycl::buffer<unsigned int, 1> buffer(4);

  // run over all available devices
  for (const cl::sycl::device &device : cl::sycl::device::get_devices()) {
    std::cout << "Running on SYCL device " << device.get_info<cl::sycl::info::device::name>()
              << ", driver version " << device.get_info<cl::sycl::info::device::driver_version>()
              << std::endl;

    // create a queue
    cl::sycl::queue queue{device};

    // size of the index space for the kernel
    cl::sycl::range<1> NumOfWorkItems{buffer.get_count()};

    // submit a command group(work) to the queue
    queue.submit([&](cl::sycl::handler &cgh) {
      // get write only access to the buffer on a device
      auto accessor = buffer.get_access<cl::sycl::access::mode::write>(cgh);
      // executing the kernel
      cgh.parallel_for<class FillBuffer>(
          NumOfWorkItems, [=](cl::sycl::id<1> WIid) {
            // do nothing
          });
    });

    // get read-only access to the buffer on the host
    // introduce an implicit barrier waiting for queue to complete the work
    const auto host_accessor = buffer.get_access<cl::sycl::access::mode::read>();
  }
}

I noticed that if I move the buffer inside the device loop, instead of reusing it, the crashes disappear.

Is a buffer supposed to be accessible across all devices ?

[Ruyk]

Thanks for the report and the code. Part of this is a known issue (cuda_piEventSetCallback not implemented), for which a patch will come up soon (@steffenlarsen is working on this).
The other problem with the OpenCL event not working we still need to investigate, seems to me its creating the user event in the wrong context, but we'll look into this.

[Ruyk]

Pinged you on #1252, which should fix this problems. There is some more work do to (enable more lit tests and a lifetime issue on the callback) that @steffenlarsen is still looking into, but I would like to see if this fixes the problem in your side.

[fwyzard]

It works (see #1252 (comment)), thank you.

[AlexeySachkov]

@fwyzard, can it be closed or there is some work left to do?

[fwyzard]

I would suggest to close this when #1252 is merged.