Skip to content

[SYCL][BindlessImages] Fix storing result events for semaphores #20080

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 1 commit into from
Sep 15, 2025
Merged

[SYCL][BindlessImages] Fix storing result events for semaphores #20080

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
20 changes: 16 additions & 4 deletions sycl/source/detail/scheduler/commands.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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<UrApiKind::urBindlessImagesWaitExternalSemaphoreExp>(
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 &&
Expand All @@ -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<UrApiKind::urBindlessImagesSignalExternalSemaphoreExp>(
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
Expand Down
50 changes: 40 additions & 10 deletions sycl/unittests/Extensions/BindlessImages/Semaphores.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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;
Expand All @@ -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<ur_event_handle_t>();
**Params.pphEvent = urBindlessImagesWaitExternalSemaphoreExp_lastEvent;
}
return UR_RESULT_SUCCESS;
}

Expand All @@ -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;
Expand All @@ -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<ur_event_handle_t>();
**Params.pphEvent = urBindlessImagesSignalExternalSemaphoreExp_lastEvent;
}
return UR_RESULT_SUCCESS;
}

Expand All @@ -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) {
Expand Down Expand Up @@ -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);
}