Catch a PI CUDA ERROR in try catch block?

[Aympab]

Hello,

I am running the same SYCL code on different GPUs (NVIDIA and AMD). The two backends have different limits in terms of grid size.
When running with a too large size for the y-dimension of the CUDA grid, I get this plugin interface error:

PI CUDA ERROR:
        Value:           1
        Name:            CUDA_ERROR_INVALID_VALUE
        Description:     invalid argument
        Function:        cuda_piEnqueueKernelLaunch
        Source Location: /opt/sycl/source/llvm/sycl/plugins/cuda/pi_cuda.cpp:3186

I would like to manage this PI exception using a try catch block, but sycl::exception and std::exception don't seem to catch it. Any ideas how to manage this exception in a portable way (i.e. that will compile with another SYCL implementation/backend)?

[AlexeySachkov]

Tagging @intel/llvm-reviewers-cuda here to comment. I don't know much about our CUDA plugin implementation, but this looks rather like a bug to mee. SYCL interface for error reporting is exceptions and it shouldn't matter which exact device/backend we are running on.

[JackAKirk]

There is a way that was agreed in UR to allow these native errors to be passed to sycl exceptions:
oneapi-src/unified-runtime#500 (comment)

I believe UR team is in the process of adapting all cases to correctly return a sycl exception whilst maintaining this degree of native error information for the user. @kbenzie may comment further.

At the moment the native error is printed without using the mechanism that results in a sycl::exception as a short term fix to allow users to see meaningful error information. This prioritizes sensible error messages, at the expense that no sycl::exception is passed along the stack.

[GeorgeWeb]

This seems to be regarding an old build of DPC++ / intel/llvm just by looking at the reported error above. For example, Source Location points to pi_cuda.cpp:3186. All errors are propagated up the stack from the backend adapters and thrown as exceptions by the SYCL runtime. I've tried to reproduce the scenario described above ("When running with a too large size for the y-dimension of the CUDA grid, I get this plugin interface error") using the latest intel/llvm -> sycl branch and I am able to catch the exception just fine.

constexpr const size_t Dim = 3;
try {
  q.submit([&](handler &cgh) {
     cgh.parallel_for<class TestDim>(
         nd_range<Dim>(range<Dim>(1, std::numeric_limits<int32_t>::max(), 1),
                       range<Dim>(1, 1, 1)),
         [=](nd_item<Dim> item) {
           // empty
         });
   }).wait();
} catch (const sycl::exception &e) {
  std::cerr << "Error: " << e.what() << std::endl;
}

Error: Number of work-groups exceed limit for dimension 1 : 65536 > 65535

This should also not be a problem with the latest release from what I can tell looking at the source code that handles this.
Please let me know if I am missing something from the report wrt sycl exceptions and if so I'll look further into it. Thanks!

[npmiller]

Closing this as it should be resolved, please feel free to re-open if the issue persists