From 8be805d42d9740e93c5def64701dc6b86b1a2f47 Mon Sep 17 00:00:00 2001 From: Artur Gainullin Date: Thu, 30 Jul 2026 12:00:43 -0700 Subject: [PATCH 1/2] [SYCL][Doc] Make ipc::event::put() optional in the spec Align the ipc::event put() wording with ipc::memory and ipc::physical_memory: the resources associated with the handle are automatically deallocated when the event object is destroyed, so calling put() is not strictly necessary. Remove the note requiring a put() call for each handle received from get(). Assisted-By: Claude --- ..._ext_oneapi_inter_process_communication.asciidoc | 13 ++++++------- 1 file changed, 6 insertions(+), 7 deletions(-) diff --git a/sycl/doc/extensions/experimental/sycl_ext_oneapi_inter_process_communication.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_oneapi_inter_process_communication.asciidoc index 9a14a3a8c62c1..1774a0ff11179 100644 --- a/sycl/doc/extensions/experimental/sycl_ext_oneapi_inter_process_communication.asciidoc +++ b/sycl/doc/extensions/experimental/sycl_ext_oneapi_inter_process_communication.asciidoc @@ -1282,18 +1282,17 @@ _Preconditions:_ * `ctx` is a context associated with the event object, which was passed to the `get` function call that produced `ipc_handle`. -_Effects:_ Deallocates resources associated with the handle. After the call to -the `put` function, the handle data is invalid and using it in the `put` and -`open` functions will result in undefined behavior. +_Effects:_ Deallocates resources associated with the handle. These resources are +automatically deallocated when the event object is destroyed, so it is not +strictly necessary to call the `put` function. After the resources associated +with the handle have been deallocated, i.e. through a call to the `put` function +or through destruction of the event object, the handle data is invalid and using +it in the `put` and `open` functions will result in undefined behavior. [_Note:_ Any objects retrieved through a call to the `open` function in any process on the system will still be valid after a call to the `put` function. _{endnote}_] -[_Note:_ A call to `put` is required for each handle received from the `get` -function. -_{endnote}_] - [_Note:_ The `put` function can only be called in the same process where the handle was generated using `get`. _{endnote}_] From 417492f9836c3267b2a6dac48106a58a2ae2731f Mon Sep 17 00:00:00 2001 From: Artur Gainullin Date: Thu, 30 Jul 2026 15:34:38 -0700 Subject: [PATCH 2/2] [SYCL] Make ipc::event::put() optional `put()` is already specified as optional for ipc::memory and ipc::physical_memory, so align specification and implementation of `ipc::event` to have same behaviour. Cache the exported IPC event handle on event_impl the first time ipc::event::get() is called, and release it in ~event_impl via urIPCPutEventHandleExp. This makes calling ipc::event::put() optional. ipc::event::put() becomes a no-op. Assisted-By: Claude --- sycl/source/detail/event_impl.cpp | 18 +++++++++++ sycl/source/detail/event_impl.hpp | 10 ++++++ sycl/source/ipc_event.cpp | 17 +++------- .../InterProcessCommunication/Event.cpp | 32 ++++++++++++++++--- 4 files changed, 59 insertions(+), 18 deletions(-) diff --git a/sycl/source/detail/event_impl.cpp b/sycl/source/detail/event_impl.cpp index 4084486139af3..6f9869cc19631 100644 --- a/sycl/source/detail/event_impl.cpp +++ b/sycl/source/detail/event_impl.cpp @@ -41,6 +41,11 @@ void event_impl::initContextIfNeeded() { event_impl::~event_impl() { try { + if (MIPCHandleData) { + getAdapter().call( + MContext->getHandleRef(), MIPCHandleData); + MIPCHandleData = nullptr; + } auto Handle = this->getHandle(); if (Handle) getAdapter().call(Handle); @@ -237,6 +242,19 @@ void event_impl::materializeIPCEvent() { // getHandleReusable. } +std::pair event_impl::getOrCreateIPCHandle() { + std::lock_guard Lock(MMutex); + if (!MIPCHandleData) { + void *HandlePtr = nullptr; + size_t HandleSize = 0; + getAdapter().call( + getHandle(), &HandlePtr, &HandleSize); + MIPCHandleData = HandlePtr; + MIPCHandleDataSize = HandleSize; + } + return {MIPCHandleData, MIPCHandleDataSize}; +} + ur_event_handle_t event_impl::getHandleReusable(queue_impl &Queue) { initContextIfNeeded(); diff --git a/sycl/source/detail/event_impl.hpp b/sycl/source/detail/event_impl.hpp index beaf490f8b00f..9207136e85b5d 100644 --- a/sycl/source/detail/event_impl.hpp +++ b/sycl/source/detail/event_impl.hpp @@ -20,6 +20,7 @@ #include #include #include +#include namespace sycl { inline namespace _V1 { @@ -456,11 +457,20 @@ class event_impl { /// True only for events imported via ipc::event::open(). bool MOpenedFromIpc = false; + /// Exported IPC handle data for a producer IPC event. + void *MIPCHandleData = nullptr; + size_t MIPCHandleDataSize = 0; + public: bool isIPCEnabled() const noexcept { return MIPCEnabled; } bool isOpenedFromIpc() const noexcept { return MOpenedFromIpc; } void setIPCEnabled(bool Value) { MIPCEnabled = Value; } + /// Returns the exported IPC handle data for this producer IPC event, + /// obtaining it from the backend on the first call and caching it on the + /// event. + std::pair getOrCreateIPCHandle(); + protected: /// Indicates that the task associated with this event has been submitted by /// the queue to the device. diff --git a/sycl/source/ipc_event.cpp b/sycl/source/ipc_event.cpp index c772c482ad16d..8d9868b5f3323 100644 --- a/sycl/source/ipc_event.cpp +++ b/sycl/source/ipc_event.cpp @@ -74,23 +74,14 @@ __SYCL_EXPORT handle get(const sycl::event &Evt) { // before the first signal. EvtImpl->materializeIPCEvent(); - sycl::detail::context_impl &CtxImpl = EvtImpl->getContextImpl(); - sycl::detail::adapter_impl &Adapter = CtxImpl.getAdapter(); - - void *HandlePtr = nullptr; - size_t HandleSize = 0; - Adapter.call( - EvtImpl->getHandle(), &HandlePtr, &HandleSize); - + auto [HandlePtr, HandleSize] = EvtImpl->getOrCreateIPCHandle(); return {HandlePtr, HandleSize}; } __SYCL_EXPORT void put(handle &IpcHandle, const sycl::context &Ctx) { - auto CtxImpl = sycl::detail::getSyclObjImpl(Ctx); - sycl::detail::adapter_impl &Adapter = CtxImpl->getAdapter(); - - Adapter.call( - CtxImpl->getHandleRef(), IpcHandle.MData); + // No-op: the exported handle is released when the event is destroyed. + (void)IpcHandle; + (void)Ctx; } } // namespace ext::oneapi::experimental::ipc::event diff --git a/sycl/unittests/Extensions/InterProcessCommunication/Event.cpp b/sycl/unittests/Extensions/InterProcessCommunication/Event.cpp index 08b0e4a42a65f..49e51fd961ced 100644 --- a/sycl/unittests/Extensions/InterProcessCommunication/Event.cpp +++ b/sycl/unittests/Extensions/InterProcessCommunication/Event.cpp @@ -195,10 +195,28 @@ TEST_F(IPCEventTests, GetCallsURAndReturnsHandle) { ASSERT_EQ(Data.size(), DummyHandleDataSize); EXPECT_EQ(memcmp(Data.data(), DummyHandleData, DummyHandleDataSize), 0); } - // get() materializes the UR event; it must be released on destruction. + // The handle is released via put on destruction and the materialized UR + // event is released too. + EXPECT_EQ(urIPCPutEventHandleExp_counter, 1); EXPECT_EQ(urEventRelease_counter, 1); } +// Repeated get() on the same event exports the handle only once (it is cached +// on the event) and returns the same data. +TEST_F(IPCEventTests, RepeatedGetExportsOnce) { + { + sycl::event Evt = syclexp::make_event( + Ctxt, syclexp::properties{syclexp::enable_ipc{true}}); + + sycl::ext::oneapi::experimental::ipc::handle H1 = ipcevt::get(Evt); + sycl::ext::oneapi::experimental::ipc::handle H2 = ipcevt::get(Evt); + + EXPECT_EQ(urIPCGetEventHandleExp_counter, 1); + EXPECT_EQ(H1.data(), H2.data()); + } + EXPECT_EQ(urIPCPutEventHandleExp_counter, 1); +} + // ipc::event::get on a non-IPC event throws errc::invalid. TEST_F(IPCEventTests, GetOnNonIPCEventThrows) { sycl::event Evt; @@ -213,8 +231,9 @@ TEST_F(IPCEventTests, GetOnNonIPCEventThrows) { EXPECT_EQ(urIPCGetEventHandleExp_counter, 0); } -// ipc::event::put calls urIPCPutEventHandleExp with the handle data. -TEST_F(IPCEventTests, PutCallsUR) { +// put() is optional and does not release the handle: the event owns it and +// releases it once on destruction, regardless of whether put() was called. +TEST_F(IPCEventTests, PutIsOptionalNoOp) { { sycl::event Evt = syclexp::make_event( Ctxt, syclexp::properties{syclexp::enable_ipc{true}}); @@ -224,10 +243,13 @@ TEST_F(IPCEventTests, PutCallsUR) { ipcevt::put(H, Ctxt); - EXPECT_EQ(urIPCPutEventHandleExp_counter, 1); + // put() is a no-op: the handle is released with the event, not here. + EXPECT_EQ(urIPCPutEventHandleExp_counter, 0); EXPECT_EQ(urIPCOpenEventHandleExp_counter, 0); } - // The materialized UR event must be released on destruction. + // Exactly one release happens on destruction, whether or not put() was + // called. + EXPECT_EQ(urIPCPutEventHandleExp_counter, 1); EXPECT_EQ(urEventRelease_counter, 1); }