Change IPC API (SYCL)

Signed-off-by: Larsen, Steffen <[email protected]>

Author: steffenlarsen

sycl/doc/extensions/experimental/sycl_ext_oneapi_inter_process_communication.asciidoc

@@ -66,9 +66,9 @@ when the kernel is submitted to the queue.
 
 This extension adds the ability for SYCL programs to share device USM memory
 allocations between processes. This is done by the allocating process creating
-a new `ipc_memory` object and transferring the "handle data" to the other
-processes. The other processes can use the handle data to recreate the
-`ipc_memory` object and get a pointer to the corresponding device USM memory.
+a new IPC memory handle through the new free frunctions and transferring the
+returned handle data to the other processes. The other processes can use the
+handle data to retrieve the corresponding device USM memory.
 
 
 == Specification
@@ -83,10 +83,6 @@ the implementation supports this feature, or applications can test the macro's
 value to determine which of the extension's features the implementation
 supports.
 
-_And follow the text with a table like this *unless the extension is
-"experimental"*.  Note that your table may have more than one row if it
-has multiple versions._
-
 [%header,cols="1,5"]
 |===
 |Value
@@ -100,29 +96,22 @@ has multiple versions._
 === Inter-process communicable memory
 
 
-This extension adds the new `ipc_memory` class. This new class adheres to the
-common reference semantics described in
-https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#sec:reference-semantics[Section 4.5.2.]
-in the SYCL 2020 specification.
+This extension adds new free functions under the `ipc_memory` experimental
+namespace.
 
 ```
-namespace sycl::ext::oneapi::experimental {
-using ipc_memory_handle_data_t = span<char, sycl::dynamic_extent>;
+namespace sycl::ext::oneapi::experimental::ipc_memory {
 
-class ipc_memory {
-public:
-  ipc_memory(void *ptr, sycl::context &ctx);
+using handle_data_t = std::vector<char>;
 
-  void put();
+handle_data_t get(void *ptr, const sycl::context &ctx);
 
-  static void *open(ipc_memory_handle_data_t ipc_memory_handle_data,
-                    const sycl::context &ctx, const sycl::device &dev);
-  static void close(void *ptr, const sycl::context &ctx);
+void put(handle_data_t &handle_data, const sycl::context &ctx);
 
-  ipc_memory_handle_data_t get_handle_data() const;
+static void *open(handle_data_t handle_data, const sycl::context &ctx,
+                  const sycl::device &dev);
 
-  void *get_ptr() const;
-};
+static void close(void *ptr, const sycl::context &ctx);
 
 }
 ```
@@ -134,53 +123,45 @@ a|
 a!
 [source]
 ----
-ipc_memory(void *ptr, const sycl::context &ctx)
+get(void *ptr, const sycl::context &ctx)
 ----
 !====
 
-_Effects:_ Constructs an IPC memory object in `ctx` from a pointer `ptr` to
-device USM memory.
-If `ptr` is not pointing to device USM memory, the behaviors of this constructor
-and any resulting objects are undefined.
+_Returns:_ A `handle_data_t` object containing the data of the IPC memory handle
+in `ctx` from a pointer `ptr` to device USM memory.
+Calling this function with a `ptr` that does not point to device USM memory, the
+behaviors is undefined.
 
 !====
 a!
 [source]
 ----
-void put()
+void put(handle_data_t &handle_data, const sycl::context &ctx)
 ----
 !====
 
 _Effects:_ Instructs the underlying IPC memory resources to be returned to
-the backend. This is not required to be called before the `ipc_memory` object
-dies. Freeing the device USM memory used when constructing this instance of
-`ipc_memory` will return the underlying IPC memory resources.
+the backend. Freeing the device USM memory used when the handle data was created
+through a call to `get()` will have the same effect as calling this function,
+so a direct call to this function is not strictly required.
 Calling this function after `sycl::free()` has been called on the device USM
-memory used when constructing this instance of `ipc_memory` will result in
-undefined behavior.
-
-_Throws:_ A `sycl::exception` with `errc::invalid` if `ipc_memory::put()` has
-previously been called on this instance of `ipc_memory`.
+memory used when the handle data was created through a call to `get()` will
+result in undefined behavior.
 
 !====
 a!
 [source]
 ----
-static void *open(ipc_memory_handle_data_t ipc_memory_handle_data,
-                  const sycl::context &ctx, const sycl::device &dev)
+static void *open(handle_data_t &handle_data, const sycl::context &ctx,
+                  const sycl::device &dev)
 ----
 !====
 
 _Effects:_ Returns a pointer to the same device USM memory as the device USM
-memory associated with the `ipc_memory` object that the handle data originated
-from.
-The `ipc_memory` object that the handle data originated from is allowed to be
-from another process on the host system.
-If the `ipc_memory` object that the handle data originated from has been
-destroyed, the behaviors of this constructor and any resulting objects are
-undefined.
-If the device USM memory the original `ipc_memory` object was created with was
-not originally allocated on `dev`, the behaviors of this function is undefined.
+memory associated with `handle_data`.
+The handle data is allowed to be from another process on the host system.
+If the handle data has been destroyed, calling this function results in
+undefined behavior.
 
 !====
 a!
@@ -191,35 +172,7 @@ static void close(void *ptr, const sycl::context &ctx)
 !====
 
 _Effects:_ Closes a device USM pointer previously returned by a call to
-`ipc_memory::open()`.
-
-!====
-a!
-[source]
-----
-ipc_memory_handle_data_t get_handle_data() const
-----
-!====
-
-_Returns:_ The handle data of the `ipc_memory` object.
-Utilizing the handle data returned by this API after the `ipc_memory` object has
-been destroyed results in undefined behavior.
-
-_Throws:_ A `sycl::exception` with `errc::invalid` if `ipc_memory::put()` has
-previously been called on this instance of `ipc_memory`.
-
-!====
-a!
-[source]
-----
-void *get_ptr() const
-----
-!====
-
-_Returns:_ A pointer to device USM memory corresponding to the pointer used to
-construct the original `ipc_memory` object.
-Accessing the pointer returned by this API after the `ipc_memory` object has
-been destroyed results in undefined behavior.
+`open()`.
 
 |====
 

sycl/include/sycl/ext/oneapi/experimental/ipc_memory.hpp

@@ -21,45 +21,19 @@ inline namespace _V1 {
 class context;
 class device;
 
-namespace detail {
-class ipc_memory_impl;
-}
+namespace ext::oneapi::experimental::ipc_memory {
 
-namespace ext::oneapi::experimental {
-using ipc_memory_handle_data_t = span<char, sycl::dynamic_extent>;
+using handle_data_t = std::vector<char>;
 
-class __SYCL_EXPORT ipc_memory
-    : public sycl::detail::OwnerLessBase<ipc_memory> {
-public:
-  ipc_memory(void *Ptr, const sycl::context &Ctx);
+__SYCL_EXPORT handle_data_t get(void *Ptr, const sycl::context &Ctx);
 
-  void put();
+__SYCL_EXPORT void put(handle_data_t &HandleData, const sycl::context &Ctx);
 
-  static void *open(ipc_memory_handle_data_t IPCMemoryHandleData,
-                    const sycl::context &Ctx, const sycl::device &Dev);
-  static void close(void *Ptr, const sycl::context &Ctx);
+__SYCL_EXPORT void *open(handle_data_t &HandleData, const sycl::context &Ctx,
+                         const sycl::device &Dev);
 
-  ipc_memory_handle_data_t get_handle_data() const;
+__SYCL_EXPORT void close(void *Ptr, const sycl::context &Ctx);
 
-  void *get_ptr() const;
-
-private:
-  ipc_memory(std::shared_ptr<sycl::detail::ipc_memory_impl> IPCMemImpl)
-      : impl{IPCMemImpl} {}
-
-  std::shared_ptr<sycl::detail::ipc_memory_impl> impl;
-
-  template <class Obj>
-  friend const decltype(Obj::impl) &
-  sycl::detail::getSyclObjImpl(const Obj &SyclObject);
-
-  template <class T>
-  friend T sycl::detail::createSyclObjFromImpl(
-      std::add_rvalue_reference_t<decltype(T::impl)> ImplObj);
-  template <class T>
-  friend T sycl::detail::createSyclObjFromImpl(
-      std::add_lvalue_reference_t<const decltype(T::impl)> ImplObj);
-};
-} // namespace ext::oneapi::experimental
+} // namespace ext::oneapi::experimental::ipc_memory
 } // namespace _V1
 } // namespace sycl

sycl/source/detail/ipc_memory_impl.hpp

@@ -8,50 +8,59 @@
 
 #include <detail/adapter_impl.hpp>
 #include <detail/context_impl.hpp>
-#include <detail/ipc_memory_impl.hpp>
 #include <sycl/context.hpp>
 #include <sycl/ext/oneapi/experimental/ipc_memory.hpp>
 
 namespace sycl {
 inline namespace _V1 {
-namespace ext::oneapi::experimental {
+namespace ext::oneapi::experimental::ipc_memory {
 
-ipc_memory::ipc_memory(void *Ptr, const sycl::context &Ctx)
-    : impl(detail::ipc_memory_impl::create(Ptr, Ctx)) {}
+__SYCL_EXPORT handle_data_t get(void *Ptr, const sycl::context &Ctx) {
+  auto CtxImpl = sycl::detail::getSyclObjImpl(Ctx);
+  sycl::detail::adapter_impl &Adapter = CtxImpl->getAdapter();
+
+  size_t HandleSize = 0;
+  Adapter.call<sycl::detail::UrApiKind::urIPCGetMemHandleExp>(
+      CtxImpl->getHandleRef(), Ptr, nullptr, &HandleSize);
 
-void ipc_memory::put() { impl->put(); }
+  handle_data_t Res(HandleSize);
+  Adapter.call<sycl::detail::UrApiKind::urIPCGetMemHandleExp>(
+      CtxImpl->getHandleRef(), Ptr, Res.data(), nullptr);
+  return Res;
+}
 
-void *ipc_memory::open(ipc_memory_handle_data_t IPCMemoryHandleData,
-                       const sycl::context &Ctx, const sycl::device &Dev) {
+__SYCL_EXPORT void put(handle_data_t &HandleData, const sycl::context &Ctx) {
+  auto CtxImpl = sycl::detail::getSyclObjImpl(Ctx);
+  sycl::detail::adapter_impl &Adapter = CtxImpl->getAdapter();
+  Adapter.call<sycl::detail::UrApiKind::urIPCPutMemHandleExp>(
+      CtxImpl->getHandleRef(), HandleData.data());
+}
+
+__SYCL_EXPORT void *open(handle_data_t &HandleData, const sycl::context &Ctx,
+                         const sycl::device &Dev) {
   auto CtxImpl = sycl::detail::getSyclObjImpl(Ctx);
   sycl::detail::adapter_impl &Adapter = CtxImpl->getAdapter();
 
   void *Ptr = nullptr;
   ur_result_t UrRes =
       Adapter.call_nocheck<sycl::detail::UrApiKind::urIPCOpenMemHandleExp>(
           CtxImpl->getHandleRef(), getSyclObjImpl(Dev)->getHandleRef(),
-          IPCMemoryHandleData.data(), IPCMemoryHandleData.size(), &Ptr);
+          HandleData.data(), HandleData.size(), &Ptr);
   if (UrRes == UR_RESULT_ERROR_INVALID_VALUE)
     throw sycl::exception(sycl::make_error_code(errc::invalid),
-                          "IPCMemoryHandleData data size does not correspond "
+                          "HandleData data size does not correspond "
                           "to the target platform's IPC memory handle size.");
   Adapter.checkUrResult(UrRes);
   return Ptr;
 }
 
-void ipc_memory::close(void *Ptr, const sycl::context &Ctx) {
+__SYCL_EXPORT void close(void *Ptr, const sycl::context &Ctx) {
   auto CtxImpl = sycl::detail::getSyclObjImpl(Ctx);
   sycl::detail::adapter_impl &Adapter = CtxImpl->getAdapter();
   Adapter.call<sycl::detail::UrApiKind::urIPCCloseMemHandleExp>(
       CtxImpl->getHandleRef(), Ptr);
 }
 
-ipc_memory_handle_data_t ipc_memory::get_handle_data() const {
-  return impl->get_handle_data();
-}
-
-void *ipc_memory::get_ptr() const { return impl->get_ptr(); }
-
-} // namespace ext::oneapi::experimental
+} // namespace ext::oneapi::experimental::ipc_memory
 } // namespace _V1
 } // namespace sycl