diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index bac765cb8d1e5..b9eb9fdf3c1e5 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -3684,11 +3684,17 @@ ur_result_t ExecCGCommand::enqueueImpQueue() { auto OptWaitValue = SemWait->getWaitValue(); uint64_t WaitValue = OptWaitValue.has_value() ? OptWaitValue.value() : 0; - return Adapter - .call_nocheck( + if (auto Result = Adapter.call_nocheck< + UrApiKind::urBindlessImagesWaitExternalSemaphoreExp>( MQueue->getHandleRef(), SemWait->getExternalSemaphore(), OptWaitValue.has_value(), WaitValue, RawEvents.size(), RawEvents.data(), Event); + Result != UR_RESULT_SUCCESS) + return Result; + + SetEventHandleOrDiscard(); + + return UR_RESULT_SUCCESS; } case CGType::SemaphoreSignal: { assert(MQueue && @@ -3698,11 +3704,17 @@ ur_result_t ExecCGCommand::enqueueImpQueue() { auto OptSignalValue = SemSignal->getSignalValue(); uint64_t SignalValue = OptSignalValue.has_value() ? OptSignalValue.value() : 0; - return Adapter - .call_nocheck( + if (auto Result = Adapter.call_nocheck< + UrApiKind::urBindlessImagesSignalExternalSemaphoreExp>( MQueue->getHandleRef(), SemSignal->getExternalSemaphore(), OptSignalValue.has_value(), SignalValue, RawEvents.size(), RawEvents.data(), Event); + Result != UR_RESULT_SUCCESS) + return Result; + + SetEventHandleOrDiscard(); + + return UR_RESULT_SUCCESS; } case CGType::AsyncAlloc: { // NO-OP. Async alloc calls adapter immediately in order to return a valid diff --git a/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp b/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp index adc9a0186d35b..16f6aa2917920 100644 --- a/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp +++ b/sycl/unittests/Extensions/BindlessImages/Semaphores.cpp @@ -16,6 +16,8 @@ constexpr uint64_t SignalValue = 24; thread_local int urBindlessImagesWaitExternalSemaphoreExp_counter = 0; thread_local bool urBindlessImagesWaitExternalSemaphoreExp_expectHasWaitValue = false; +thread_local ur_event_handle_t + urBindlessImagesWaitExternalSemaphoreExp_lastEvent = nullptr; inline ur_result_t urBindlessImagesWaitExternalSemaphoreExp_replace(void *pParams) { ++urBindlessImagesWaitExternalSemaphoreExp_counter; @@ -30,6 +32,11 @@ urBindlessImagesWaitExternalSemaphoreExp_replace(void *pParams) { EXPECT_EQ(*Params.pnumEventsInWaitList, uint32_t{0}); EXPECT_EQ(*Params.pphEventWaitList, nullptr); EXPECT_NE(*Params.pphEvent, nullptr); + if (*Params.pphEvent) { + urBindlessImagesWaitExternalSemaphoreExp_lastEvent = + mock::createDummyHandle(); + **Params.pphEvent = urBindlessImagesWaitExternalSemaphoreExp_lastEvent; + } return UR_RESULT_SUCCESS; } @@ -38,6 +45,8 @@ thread_local bool urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = false; thread_local uint32_t urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 0; +thread_local ur_event_handle_t + urBindlessImagesSignalExternalSemaphoreExp_lastEvent = nullptr; inline ur_result_t urBindlessImagesSignalExternalSemaphoreExp_replace(void *pParams) { ++urBindlessImagesSignalExternalSemaphoreExp_counter; @@ -57,6 +66,11 @@ urBindlessImagesSignalExternalSemaphoreExp_replace(void *pParams) { EXPECT_EQ(*Params.pphEventWaitList, nullptr); } EXPECT_NE(*Params.pphEvent, nullptr); + if (*Params.pphEvent) { + urBindlessImagesSignalExternalSemaphoreExp_lastEvent = + mock::createDummyHandle(); + **Params.pphEvent = urBindlessImagesSignalExternalSemaphoreExp_lastEvent; + } return UR_RESULT_SUCCESS; } @@ -80,15 +94,19 @@ TEST(BindlessImagesExtensionTests, ExternalSemaphoreWait) { syclexp::external_semaphore_handle_type::opaque_fd; urBindlessImagesWaitExternalSemaphoreExp_expectHasWaitValue = false; - Q.ext_oneapi_wait_external_semaphore(DummySemaphore); + sycl::event E = Q.ext_oneapi_wait_external_semaphore(DummySemaphore); EXPECT_EQ(urBindlessImagesWaitExternalSemaphoreExp_counter, 1); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesWaitExternalSemaphoreExp_lastEvent); DummySemaphore.handle_type = syclexp::external_semaphore_handle_type::timeline_fd; urBindlessImagesWaitExternalSemaphoreExp_expectHasWaitValue = true; - Q.ext_oneapi_wait_external_semaphore(DummySemaphore, WaitValue); + E = Q.ext_oneapi_wait_external_semaphore(DummySemaphore, WaitValue); EXPECT_EQ(urBindlessImagesWaitExternalSemaphoreExp_counter, 2); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesWaitExternalSemaphoreExp_lastEvent); } TEST(BindlessImagesExtensionTests, ExternalSemaphoreSignal) { @@ -126,36 +144,48 @@ TEST(BindlessImagesExtensionTests, ExternalSemaphoreSignal) { urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = false; urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 0; - Q.ext_oneapi_signal_external_semaphore(DummySemaphore); + sycl::event E = Q.ext_oneapi_signal_external_semaphore(DummySemaphore); EXPECT_EQ(urBindlessImagesSignalExternalSemaphoreExp_counter, 1); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesSignalExternalSemaphoreExp_lastEvent); urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = false; urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 1; - Q.ext_oneapi_signal_external_semaphore(DummySemaphore, DummyEvent1); + E = Q.ext_oneapi_signal_external_semaphore(DummySemaphore, DummyEvent1); EXPECT_EQ(urBindlessImagesSignalExternalSemaphoreExp_counter, 2); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesSignalExternalSemaphoreExp_lastEvent); urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = false; urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 2; - Q.ext_oneapi_signal_external_semaphore(DummySemaphore, DummyEventList); + E = Q.ext_oneapi_signal_external_semaphore(DummySemaphore, DummyEventList); EXPECT_EQ(urBindlessImagesSignalExternalSemaphoreExp_counter, 3); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesSignalExternalSemaphoreExp_lastEvent); DummySemaphore.handle_type = syclexp::external_semaphore_handle_type::timeline_fd; urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = true; urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 0; - Q.ext_oneapi_signal_external_semaphore(DummySemaphore, SignalValue); + E = Q.ext_oneapi_signal_external_semaphore(DummySemaphore, SignalValue); EXPECT_EQ(urBindlessImagesSignalExternalSemaphoreExp_counter, 4); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesSignalExternalSemaphoreExp_lastEvent); urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = true; urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 1; - Q.ext_oneapi_signal_external_semaphore(DummySemaphore, SignalValue, - DummyEvent1); + E = Q.ext_oneapi_signal_external_semaphore(DummySemaphore, SignalValue, + DummyEvent1); EXPECT_EQ(urBindlessImagesSignalExternalSemaphoreExp_counter, 5); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesSignalExternalSemaphoreExp_lastEvent); urBindlessImagesSignalExternalSemaphoreExp_expectHasSignalValue = true; urBindlessImagesSignalExternalSemaphoreExp_expectedNumWaitEvents = 2; - Q.ext_oneapi_signal_external_semaphore(DummySemaphore, SignalValue, - DummyEventList); + E = Q.ext_oneapi_signal_external_semaphore(DummySemaphore, SignalValue, + DummyEventList); EXPECT_EQ(urBindlessImagesSignalExternalSemaphoreExp_counter, 6); + EXPECT_EQ(sycl::detail::getSyclObjImpl(E)->getHandle(), + urBindlessImagesSignalExternalSemaphoreExp_lastEvent); }