[SYCL][UR][OpenCL] Implement partial support for UR_QUEUE_INFO_EMPTY

igchor · igchor · commit 8c508fb328ee · 2025-05-21T20:16:40.000Z
By storing last event in the queue.

This only works for in-order queues and when last enqueued operation
had signalEvent set. Otherwise UR_RESULT_ERROR_UNSUPPORTED_ENUMERATION
is returned.
diff --git a/sycl/source/detail/queue_impl.cpp b/sycl/source/detail/queue_impl.cpp
@@ -500,8 +500,7 @@ event queue_impl::submitMemOpHelper(const std::shared_ptr<queue_impl> &Self,
         }
       }
 
-      if (isInOrder() &&
-          (!isNoEventsMode || MContext->getBackend() == backend::opencl)) {
+      if (isInOrder() && !isNoEventsMode) {
         auto &EventToStoreIn = MGraph.expired() ? MDefaultGraphDeps.LastEventPtr
                                                 : MExtGraphDeps.LastEventPtr;
         EventToStoreIn = EventImpl;
diff --git a/sycl/source/detail/queue_impl.hpp b/sycl/source/detail/queue_impl.hpp
@@ -722,24 +722,14 @@ class queue_impl {
   event finalizeHandlerInOrderNoEventsUnlocked(HandlerType &Handler) {
     assert(isInOrder());
     assert(MGraph.expired());
-    assert(MDefaultGraphDeps.LastEventPtr == nullptr ||
-           MContext->getBackend() == backend::opencl);
+    assert(MDefaultGraphDeps.LastEventPtr == nullptr);
     assert(MNoEventMode);
 
     MEmpty = false;
 
     synchronizeWithExternalEvent(Handler);
 
-    if (MContext->getBackend() == backend::opencl && MGraph.expired()) {
-      // This is needed to support queue_empty() call
-      auto Event = Handler.finalize();
-      if (!getSyclObjImpl(Event)->isDiscarded()) {
-        MDefaultGraphDeps.LastEventPtr = getSyclObjImpl(Event);
-      }
-      return Event;
-    } else {
-      return Handler.finalize();
-    }
+    return Handler.finalize();
   }
 
   template <typename HandlerType = handler>
diff --git a/sycl/test-e2e/InOrderEventsExt/get_last_event.cpp b/sycl/test-e2e/InOrderEventsExt/get_last_event.cpp
@@ -36,17 +36,10 @@ int Check(const sycl::queue &Q, const char *CheckName, const F &CheckFunc) {
               << std::endl;
     return 1;
   }
-  if (Q.get_backend() == sycl::backend::opencl) {
-    if (*E != *LastEvent) {
-      std::cout << "opencl backend should store last event in the queue"
-                << std::endl;
-      return 1;
-    }
-  } else if (LastEvent->get_info<
-                 sycl::info::event::command_execution_status>() ==
-                 sycl::info::event_command_status::complete &&
-             E->get_info<sycl::info::event::command_execution_status>() !=
-                 sycl::info::event_command_status::complete) {
+  if (LastEvent->get_info<sycl::info::event::command_execution_status>() ==
+          sycl::info::event_command_status::complete &&
+      E->get_info<sycl::info::event::command_execution_status>() !=
+          sycl::info::event_command_status::complete) {
     std::cout << "ext_oneapi_get_last_event() returned an event that is "
                  "complete, but the event returned by CheckFunc() is not."
               << std::endl;
diff --git a/sycl/test-e2e/InorderQueue/in_order_ext_oneapi_submit_barrier.cpp b/sycl/test-e2e/InorderQueue/in_order_ext_oneapi_submit_barrier.cpp
@@ -14,7 +14,7 @@ bool checkBarrierEvent(sycl::backend backend, sycl::event LastEvent,
                        sycl::event BarrierEvent, bool noEventMode) {
   // In noEventMode or when using opencl backend,
   // barrier will always return last event
-  if (backend == sycl::backend::opencl || !noEventMode) {
+  if (!noEventMode) {
     return BarrierEvent == LastEvent;
   } else {
     return BarrierEvent != LastEvent;
diff --git a/sycl/unittests/scheduler/InOrderQueueHostTaskDeps.cpp b/sycl/unittests/scheduler/InOrderQueueHostTaskDeps.cpp
@@ -45,11 +45,6 @@ TEST_F(SchedulerTest, InOrderQueueHostTaskDeps) {
       .wait();
 
   size_t expectedCount = 1u;
-
-  // OpenCL needs to store all events so does not need a barrier
-  if (Ctx.get_platform().get_backend() == backend::opencl)
-    expectedCount = 0u;
-
   EXPECT_EQ(GEventsWaitCounter, expectedCount);
 }
 
diff --git a/unified-runtime/source/adapters/opencl/event.hpp b/unified-runtime/source/adapters/opencl/event.hpp
@@ -69,16 +69,20 @@ inline cl_event *ifUrEvent(ur_event_handle_t *ReturnedEvent, cl_event &Event) {
 inline ur_result_t createUREvent(cl_event Event, ur_context_handle_t Context,
                                  ur_queue_handle_t Queue,
                                  ur_event_handle_t *ReturnedEvent) {
+  assert(Queue);
   if (ReturnedEvent) {
     try {
       auto UREvent =
           std::make_unique<ur_event_handle_t_>(Event, Context, Queue);
       *ReturnedEvent = UREvent.release();
+      UR_RETURN_ON_FAILURE(Queue->storeLastEvent(*ReturnedEvent));
     } catch (std::bad_alloc &) {
       return UR_RESULT_ERROR_OUT_OF_RESOURCES;
     } catch (...) {
       return UR_RESULT_ERROR_UNKNOWN;
     }
+  } else {
+    UR_RETURN_ON_FAILURE(Queue->storeLastEvent(nullptr));
   }
   return UR_RESULT_SUCCESS;
 }
diff --git a/unified-runtime/source/adapters/opencl/queue.cpp b/unified-runtime/source/adapters/opencl/queue.cpp
@@ -94,8 +94,8 @@ ur_result_t ur_queue_handle_t_::makeWithNative(native_type NativeQueue,
       UR_RETURN_ON_FAILURE(urDeviceCreateWithNativeHandle(
           hNativeHandle, nullptr, nullptr, &Device));
     }
-    auto URQueue =
-        std::make_unique<ur_queue_handle_t_>(NativeQueue, Context, Device);
+    auto URQueue = std::make_unique<ur_queue_handle_t_>(NativeQueue, Context,
+                                                        Device, false);
     Queue = URQueue.release();
   } catch (std::bad_alloc &) {
     return UR_RESULT_ERROR_OUT_OF_RESOURCES;
@@ -125,14 +125,16 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreate(
 
   cl_int RetErr = CL_INVALID_OPERATION;
 
+  bool InOrder = !(CLProperties & SupportByOpenCL &
+                   CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE);
   if (Version < oclv::V2_0) {
     cl_command_queue Queue =
         clCreateCommandQueue(hContext->CLContext, hDevice->CLDevice,
                              CLProperties & SupportByOpenCL, &RetErr);
     CL_RETURN_ON_FAILURE(RetErr);
     try {
-      auto URQueue =
-          std::make_unique<ur_queue_handle_t_>(Queue, hContext, hDevice);
+      auto URQueue = std::make_unique<ur_queue_handle_t_>(Queue, hContext,
+                                                          hDevice, InOrder);
       *phQueue = URQueue.release();
     } catch (std::bad_alloc &) {
       return UR_RESULT_ERROR_OUT_OF_RESOURCES;
@@ -151,7 +153,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueCreate(
   CL_RETURN_ON_FAILURE(RetErr);
   try {
     auto URQueue =
-        std::make_unique<ur_queue_handle_t_>(Queue, hContext, hDevice);
+        std::make_unique<ur_queue_handle_t_>(Queue, hContext, hDevice, InOrder);
     *phQueue = URQueue.release();
   } catch (std::bad_alloc &) {
     return UR_RESULT_ERROR_OUT_OF_RESOURCES;
@@ -166,12 +168,24 @@ UR_APIEXPORT ur_result_t UR_APICALL urQueueGetInfo(ur_queue_handle_t hQueue,
                                                    size_t propSize,
                                                    void *pPropValue,
                                                    size_t *pPropSizeRet) {
-  if (propName == UR_QUEUE_INFO_EMPTY) {
-    // OpenCL doesn't provide API to check the status of the queue.
-    return UR_RESULT_ERROR_UNSUPPORTED_ENUMERATION;
-  }
   cl_command_queue_info CLCommandQueueInfo = mapURQueueInfoToCL(propName);
   UrReturnHelper ReturnValue(propSize, pPropValue, pPropSizeRet);
+  if (propName == UR_QUEUE_INFO_EMPTY) {
+    if (!hQueue->LastEvent) {
+      // OpenCL doesn't provide API to check the status of the queue.
+      return UR_RESULT_ERROR_UNSUPPORTED_ENUMERATION;
+    } else {
+      ur_event_status_t Status;
+      UR_RETURN_ON_FAILURE(urEventGetInfo(
+          hQueue->LastEvent, UR_EVENT_INFO_COMMAND_EXECUTION_STATUS,
+          sizeof(ur_event_status_t), (void *)&Status, nullptr));
+      if (Status == UR_EVENT_STATUS_COMPLETE) {
+        return ReturnValue(true);
+      } else {
+        return ReturnValue(false);
+      }
+    }
+  }
   switch (propName) {
   case UR_QUEUE_INFO_CONTEXT: {
     return ReturnValue(hQueue->Context);
diff --git a/unified-runtime/source/adapters/opencl/queue.hpp b/unified-runtime/source/adapters/opencl/queue.hpp
@@ -24,10 +24,14 @@ struct ur_queue_handle_t_ : ur::opencl::handle_base {
   std::optional<ur_queue_handle_t> DeviceDefault = std::nullopt;
   std::atomic<uint32_t> RefCount = 0;
   bool IsNativeHandleOwned = true;
+  // Used to implement UR_QUEUE_INFO_EMPTY query
+  bool IsInOrder;
+  ur_event_handle_t LastEvent = nullptr;
 
   ur_queue_handle_t_(native_type Queue, ur_context_handle_t Ctx,
-                     ur_device_handle_t Dev)
-      : handle_base(), CLQueue(Queue), Context(Ctx), Device(Dev) {
+                     ur_device_handle_t Dev, bool InOrder)
+      : handle_base(), CLQueue(Queue), Context(Ctx), Device(Dev),
+        IsInOrder(InOrder) {
     RefCount = 1;
     urDeviceRetain(Device);
     urContextRetain(Context);
@@ -54,4 +58,18 @@ struct ur_queue_handle_t_ : ur::opencl::handle_base {
   uint32_t decrementReferenceCount() noexcept { return --RefCount; }
 
   uint32_t getReferenceCount() const noexcept { return RefCount; }
+
+  ur_result_t storeLastEvent(ur_event_handle_t Event) {
+    if (!IsInOrder) {
+      return UR_RESULT_SUCCESS;
+    }
+    if (LastEvent) {
+      UR_RETURN_ON_FAILURE(urEventRelease(LastEvent));
+    }
+    LastEvent = Event;
+    if (LastEvent) {
+      UR_RETURN_ON_FAILURE(urEventRetain(LastEvent));
+    }
+    return UR_RESULT_SUCCESS;
+  }
 };

Original file line number	Diff line number	Diff line change
`@@ -500,8 +500,7 @@ event queue_impl::submitMemOpHelper(const std::shared_ptr<queue_impl> &Self,`
`500`	`500`	`}`
`501`	`501`	`}`
`502`	`502`
`503`		`- if (isInOrder() &&`
`504`		`- (!isNoEventsMode \|\| MContext->getBackend() == backend::opencl)) {`
	`503`	`+ if (isInOrder() && !isNoEventsMode) {`
`505`	`504`	`auto &EventToStoreIn = MGraph.expired() ? MDefaultGraphDeps.LastEventPtr`
`506`	`505`	`: MExtGraphDeps.LastEventPtr;`
`507`	`506`	`EventToStoreIn = EventImpl;`