[SYCL] Minor host task corrections (#2145)

1. Fixed compilation error which appeared in case of usage
get_native_mem method with accessor with host target.
2. Exceptions which raised in host task reported as async exceptions.
3. Fixed stable deadlock which appeared when reporting an async
exception raised in a host task.

Signed-off-by: Ivan Karachun <[email protected]>

Author: Ivan Karachun

sycl/include/CL/sycl/backend/cuda.hpp

@@ -50,5 +50,12 @@ struct interop<backend::cuda, accessor<DataT, Dimensions, AccessMode,
   using type = CUdeviceptr;
 };
 
+template <typename DataT, int Dimensions, access::mode AccessMode>
+struct interop<backend::cuda, accessor<DataT, Dimensions, AccessMode,
+                                       access::target::constant_buffer,
+                                       access::placeholder::false_t>> {
+  using type = CUdeviceptr;
+};
+
 } // namespace sycl
 } // namespace cl

sycl/include/CL/sycl/backend/level_zero.hpp

@@ -37,6 +37,13 @@ struct interop<backend::level0, accessor<DataT, Dimensions, AccessMode,
   using type = char *;
 };
 
+template <typename DataT, int Dimensions, access::mode AccessMode>
+struct interop<backend::level0, accessor<DataT, Dimensions, AccessMode,
+                                         access::target::constant_buffer,
+                                         access::placeholder::false_t>> {
+  using type = char *;
+};
+
 namespace level0 {
 
 // Implementation of various "make" functions resides in libsycl.so

sycl/include/CL/sycl/backend/opencl.hpp

@@ -43,6 +43,13 @@ struct interop<backend::opencl, accessor<DataT, Dimensions, AccessMode,
   using type = cl_mem;
 };
 
+template <typename DataT, int Dimensions, access::mode AccessMode>
+struct interop<backend::opencl, accessor<DataT, Dimensions, AccessMode,
+                                         access::target::constant_buffer,
+                                         access::placeholder::false_t>> {
+  using type = cl_mem;
+};
+
 namespace opencl {
 
 // Implementation of various "make" functions resides in SYCL RT because

sycl/include/CL/sycl/interop_handle.hpp

@@ -45,7 +45,8 @@ class interop_handle {
   template <backend BackendName = backend::opencl, typename DataT, int Dims,
             access::mode Mode, access::target Target, access::placeholder IsPlh>
   typename std::enable_if<
-      Target != access::target::host_buffer,
+      Target == access::target::global_buffer ||
+          Target == access::target::constant_buffer,
       typename interop<BackendName,
                        accessor<DataT, Dims, Mode, Target, IsPlh>>::type>::type
   get_native_mem(const accessor<DataT, Dims, Mode, Target, IsPlh> &Acc) const {
@@ -63,12 +64,14 @@ class interop_handle {
   template <backend BackendName = backend::opencl, typename DataT, int Dims,
             access::mode Mode, access::target Target, access::placeholder IsPlh>
   typename std::enable_if<
-      Target == access::target::host_buffer,
-      typename interop<BackendName,
-                       accessor<DataT, Dims, Mode, Target, IsPlh>>::type>::type
+      !(Target == access::target::global_buffer ||
+        Target == access::target::constant_buffer),
+      typename interop<BackendName, accessor<DataT, Dims, Mode,
+                                             access::target::global_buffer,
+                                             IsPlh>>::type>::type
   get_native_mem(const accessor<DataT, Dims, Mode, Target, IsPlh> &) const {
-    throw invalid_object_error("Getting memory object out of host accessor is "
-                               "not allowed",
+    throw invalid_object_error("Getting memory object out of accessor for "
+                               "specified target is not allowed",
                                PI_INVALID_MEM_OBJECT);
   }
 

sycl/source/detail/queue_impl.cpp

@@ -192,11 +192,18 @@ void queue_impl::wait(const detail::code_location &CodeLoc) {
   TelemetryEvent = instrumentationProlog(CodeLoc, Name, StreamID, IId);
 #endif
 
-  std::lock_guard<mutex_class> Guard(MMutex);
-  for (std::weak_ptr<event_impl> &EventImplWeakPtr : MEvents) {
-    if (std::shared_ptr<event_impl> EventImplPtr = EventImplWeakPtr.lock())
-      EventImplPtr->wait(EventImplPtr);
+  std::vector<std::shared_ptr<event_impl>> Events;
+  {
+    std::lock_guard<mutex_class> Guard(MMutex);
+    for (std::weak_ptr<event_impl> &EventImplWeakPtr : MEvents)
+      if (std::shared_ptr<event_impl> EventImplPtr = EventImplWeakPtr.lock())
+        Events.push_back(EventImplPtr);
+  }
+
+  for (std::shared_ptr<event_impl> &Event : Events) {
+    Event->wait(Event);
   }
+
   for (event &Event : MUSMEvents) {
     Event.wait();
   }

sycl/source/detail/scheduler/commands.cpp

@@ -201,15 +201,19 @@ class DispatchHostTask {
 
     CGHostTask &HostTask = static_cast<CGHostTask &>(MThisCmd->getCG());
 
-    // we're ready to call the user-defined lambda now
-    if (HostTask.MHostTask->isInteropTask()) {
-      interop_handle IH{MReqToMem, HostTask.MQueue,
-                        getSyclObjImpl(HostTask.MQueue->get_device()),
-                        HostTask.MQueue->getContextImplPtr()};
-
-      HostTask.MHostTask->call(IH);
-    } else
-      HostTask.MHostTask->call();
+    try {
+      // we're ready to call the user-defined lambda now
+      if (HostTask.MHostTask->isInteropTask()) {
+        interop_handle IH{MReqToMem, HostTask.MQueue,
+                          getSyclObjImpl(HostTask.MQueue->get_device()),
+                          HostTask.MQueue->getContextImplPtr()};
+
+        HostTask.MHostTask->call(IH);
+      } else
+        HostTask.MHostTask->call();
+    } catch (...) {
+      HostTask.MQueue->reportAsyncException(std::current_exception());
+    }
 
     HostTask.MHostTask.reset();
 

sycl/test/host-interop-task/interop-task.cpp

@@ -219,11 +219,41 @@ void test5() {
   }
 }
 
+// The test checks that an exception which is thrown from host_task body
+// is reported as asynchronous.
+void test6() {
+  queue Queue([](sycl::exception_list ExceptionList) {
+    if (ExceptionList.size() != 1) {
+      std::cerr << "Should be one exception in exception list" << std::endl;
+      std::abort();
+    }
+    std::rethrow_exception(*ExceptionList.begin());
+  });
+
+  try {
+    size_t size = 1;
+    buffer<int, 1> Buf{size};
+    Queue.submit([&](sycl::handler &CGH) {
+      auto acc = Buf.get_access<mode::write, target::host_buffer>(CGH);
+      CGH.codeplay_host_task(
+          [=](interop_handle IH) { (void)IH.get_native_mem(acc); });
+    });
+    Queue.wait_and_throw();
+    assert(!"Expected exception was not caught");
+  } catch (sycl::exception &ExpectedException) {
+    assert(std::string(ExpectedException.what())
+                   .find("memory object out of accessor for specified target "
+                         "is not allowed") != std::string::npos &&
+           "Unexpected error was caught!");
+  }
+}
+
 int main() {
   test1();
   test2();
   test3();
   test4();
   test5();
+  test6();
   return 0;
 }