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 0489ecf0d8969..b962725934c22 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 @@ -57,6 +57,10 @@ there to be a per-platform default context even if the core SYCL specification does not provide a convenient way to get it. _{endnote}_] +This extension also depends on the following other SYCL extensions: + +* link:sycl_ext_oneapi_virtual_mem.asciidoc[sycl_ext_oneapi_virtual_mem] +* link:sycl_ext_oneapi_properties.asciidoc[sycl_ext_oneapi_properties] == Status @@ -70,11 +74,11 @@ specification.* == Overview -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 handle through the new free functions and transferring the +This extension adds the ability for SYCL programs to share various objects +or USM memory allocations between processes. This is done by creating a new IPC +handle in one process through the new free functions 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. +handle data to retrieve the corresponding object or allocation. == Specification @@ -99,7 +103,71 @@ implementation supports. feature-test macro always has this value. |=== -=== Extension to `enum class aspect` +=== Generic types + +This extension adds the following types which can be used to share SYCL objects +with another process. + +``` +namespace sycl::ext::oneapi::experimental::ipc { + +using handle_data_t = std::vector; + +// Requires C++20 +using handle_data_view_t = std::span; + +struct handle { + handle_data_t data() const; + + // Requires C++20 + handle_data_view_t data_view() const; +}; + +} +``` + +|==== +a| +[frame=all,grid=none] +!==== +a! +[source] +---- +handle_data_t data() const; +---- +!==== + +_Preconditions:_ The resources associated with this handle have not been +released by the relevant IPC API or by the destruction of the original object. + +_Returns:_ The handle data associated with the IPC handle object. This data can +be transferred to other processes, but cannot be used to recreate a `handle` +object. + +!==== +a! +[source] +---- +handle_data_view_t data_view() const; +---- +!==== + +_Preconditions:_ The resources associated with this handle have not been +released by the relevant IPC API or by the destruction of the original object. + +_Returns:_ A `std::span` with a view into the handle data associated with the +IPC handle object. This data can be transferred to other processes, but cannot +be used to recreate a `handle` object. + +|==== + +=== Inter-process communicable memory + +This extension allows for the exchange of USM device memory pointers between +processes. The following aspect and set of free functions are specific to USM +memory sharing feature. + +==== Extension to `enum class aspect` for USM memory [source] ---- @@ -114,15 +182,301 @@ enum class aspect { If a SYCL device has this aspect, that device supports the `get` and `open` functions specified in the following section. +==== New functions and types associated with USM memory -=== Inter-process communicable memory +This extension adds new free functions under the `ipc::memory` experimental +namespace. + +``` +namespace sycl::ext::oneapi::experimental::ipc::memory { + +handle get(void *ptr, const sycl::context &ctx); + +handle get(void *ptr); + +void put(handle &ipc_handle, const sycl::context &ctx); + +void put(handle &ipc_handle); + +void *open(const handle_data_t &handle_data, + const sycl::context &ctx, const sycl::device &dev); + +void *open(const handle_data_t &handle_data, const sycl::device &dev); + +void *open(const handle_data_t &handle_data); + +// Requires C++20 +void *open(const handle_data_view_t &handle_data, const sycl::context &ctx, + const sycl::device &dev); + +// Requires C++20 +void *open(const handle_data_view_t &handle_data, const sycl::device &dev); + +// Requires C++20 +void *open(const handle_data_view_t &handle_data); + +void close(void *ptr, const sycl::context &ctx); + +void close(void *ptr); + +} +``` + +|==== +a| +[frame=all,grid=none] +!==== +a! +[source] +---- +handle get(void *ptr, const sycl::context &ctx) +---- +!==== + +_Preconditions:_ `ptr` is a pointer to USM device memory on some device _D_, and +`ctx` is the same context as `ptr` was allocated against, using the USM device +memory allocation routines. + +_Returns:_ An IPC "handle" to this USM memory allocation. The bytes of this +handle can be transferred to another process on the same system, and the other +process can use the handle to get a pointer to the same USM allocation through a +call to the `open` function. + +_Throws:_ An exception with the `errc::feature_not_supported` error code if +device _D_ does not have `aspect::ext_oneapi_ipc_memory`. + +!==== +a! +[source] +---- +handle get(void *ptr) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +return ipc::memory::get(ptr, ctxt); +---- + +!==== +a! +[source] +---- +void put(handle &ipc_handle, const sycl::context &ctx) +---- +!==== + +_Preconditions:_ `ipc_handle` is the IPC "handle" to USM device memory that was +returned from a call to `get`. The `put` function has not been previously called on the +handle. + +_Effects:_ Deallocates resources associated with the handle. These resources are +automatically deallocated when the USM device memory is freed, 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 freeing the USM device memory, the handle data is invalid and using +it in the `put` and `open` functions will result in undefined behavior. + +[_Note:_ Any pointers 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 and +must still be freed through calls to the `close` function. +_{endnote}_] + +!==== +a! +[source] +---- +void put(handle &ipc_handle) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +ipc::memory::put(ipc_handle, ctxt); +---- + +!==== +a! +[source] +---- +void *open(const handle_data_t &handle_data, const sycl::context &ctx, + const sycl::device &dev) +---- +!==== + +_Preconditions:_ `handle_data` is the IPC "handle" to USM device memory that was +returned from a call to the `get` function either in this process or in some +other process on the same system. That USM device memory is accessible on device +`dev`. + +_Returns:_ A pointer to the same USM device memory represented by `handle_data`. +The returned pointer is associated with context `ctx`. It can be used wherever a +USM device pointer for device `dev` and context `ctx` are expected, except it +cannot be passed to `sycl::free`. Instead, use the `close` function to free this +memory pointer. + +[_Note:_ The `open` function can be called multiple times on the same handle +within the same process. Each call to the `open` function may return a unique +pointer value even for the same handle, therefore each call to the `open` +function must have a matching call to the `close` function. +_{endnote}_] + +[_Note:_ The pointer returned from a call to the `open` function is no longer +valid if the associated USM device memory is freed through a call to the +`sycl::free` function. +_{endnote}_] + +[_Note:_ For sub-page allocations, the implementation may need to share the +entire page with the other process. As a result, the other process may have +access to adjacent USM memory that happens to share the same page. Applications +should be aware of the potential security implications if the USM memory +contains sensitive information, and the other application isn't trusted. +_{endnote}_] + +_Throws:_ + + * An exception with the `errc::feature_not_supported` error code if device + `dev` does not have `aspect::ext_oneapi_ipc_memory`. + * An exception with the `errc::invalid` error code if the handle data + `handle_data` has an unexpected number of bytes. + +!==== +a! +[source] +---- +void *open(const handle_data_t &handle_data, const sycl::device &dev) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::context ctxt = dev.get_platform().khr_get_default_context(); +return ipc::memory::open(handle_data, ctxt, dev); +---- + +!==== +a! +[source] +---- +void *open(const handle_data_t &handle_data) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +return ipc::memory::open(handle_data, ctxt, d); +---- + +!==== +a! +[source] +---- +void *open(const handle_data_view_t &handle_data_view, + const sycl::context &ctx, const sycl::device &dev) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +handle_data_t handle_data{handle_data_view.begin(), handle_data_view.end()}; +return ipc::memory::open(handle_data, ctx, dev); +---- + +!==== +a! +[source] +---- +void *open(const handle_data_view_t &handle_data_view, const sycl::device &dev) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::context ctxt = dev.get_platform().khr_get_default_context(); +return ipc::memory::open(handle_data_view, ctxt, dev); +---- + +!==== +a! +[source] +---- +void *open(const handle_data_view_t &handle_data_view) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +return ipc::memory::open(handle_data_view, ctxt, d); +---- + +!==== +a! +[source] +---- +void close(void *ptr, const sycl::context &ctx) +---- +!==== + +_Precondition:_ `ptr` was previously returned from a call to the `open` function +in this same process, where `ctx` was passed as the context. This `ptr` value +has not yet been closed by calling the `close` function. + +_Effects:_ Closes a device USM pointer previously returned by a call to +the `open` function. + +!==== +a! +[source] +---- +void close(void *ptr) +---- +!==== + +_Effects_: Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +ipc::memory::close(ptr, ctxt); +---- + +|==== + +==== New functions and types associated with USM memory (deprecated) This extension adds a new type aliases `handle_data_t` and `handle_data_view_t` as well as a new IPC memory handle type `handle` under the `ipc_memory` experimental namespace. +Note, that the `ipc_memory` namespace is deprecated and applications should use +the `ipc::memory` namespace instead. + ``` -namespace sycl::ext::oneapi::experimental::ipc_memory { +namespace sycl::ext::oneapi::experimental { +namespace [[deprecated]] ipc_memory { using handle_data_t = std::vector; @@ -136,6 +490,7 @@ struct handle { handle_data_view_t data_view() const; }; +} } ``` @@ -179,7 +534,8 @@ Additionally, this extension adds new free functions under the `ipc_memory` experimental namespace. ``` -namespace sycl::ext::oneapi::experimental::ipc_memory { +namespace sycl::ext::oneapi::experimental { +namespace [[deprecated]] ipc_memory { handle get(void *ptr, const sycl::context &ctx); @@ -210,6 +566,7 @@ void close(void *ptr, const sycl::context &ctx); void close(void *ptr); +} } ``` @@ -291,7 +648,7 @@ _Effects_: Equivalent to: ---- sycl::device d; sycl::context ctxt = d.get_platform().khr_get_default_context(); -ipc_memory::put(handle_data, ctxt); +ipc_memory::put(ipc_handle, ctxt); ---- !==== @@ -352,15 +709,14 @@ _Effects_: Equivalent to: [source,c++,indent=2] ---- sycl::context ctxt = dev.get_platform().khr_get_default_context(); -return ipc_memory::put(handle_data, ctxt, dev); +return ipc_memory::open(handle_data, ctxt, dev); ---- !==== a! [source] ---- -void *open(const handle_data_t &handle_data, const sycl::context &ctx, - const sycl::device &dev) +void *open(const handle_data_t &handle_data) ---- !==== @@ -387,7 +743,7 @@ _Effects_: Equivalent to: [source,c++,indent=2] ---- handle_data_t handle_data{handle_data_view.begin(), handle_data_view.end()}; -return ipc_memory::put(handle_data, ctx, dev); +return ipc_memory::open(handle_data, ctx, dev); ---- !==== @@ -403,15 +759,14 @@ _Effects_: Equivalent to: [source,c++,indent=2] ---- sycl::context ctxt = dev.get_platform().khr_get_default_context(); -return ipc_memory::put(handle_data_view, ctxt, dev); +return ipc_memory::open(handle_data_view, ctxt, dev); ---- !==== a! [source] ---- -void *open(const handle_data_view_t &handle_data_view, const sycl::context &ctx, - const sycl::device &dev) +void *open(const handle_data_view_t &handle_data_view) ---- !==== @@ -458,6 +813,258 @@ ipc_memory::close(ptr, ctxt); |==== +=== Inter-process exchange of physical memory + +This extension allows for the exchange of a physical memory handle between +processes. A given physical memory handle can be mapped into a virtual address +space in multiple processes, so the processes can access the same device +memory. + +The following aspect and set of free functions are specific to +physical memory handle sharing feature. + +The physical memory is represented by the +`sycl::ext::oneapi::experimental::physical_mem` object instance +(from the link:sycl_ext_oneapi_virtual_mem.asciidoc[ +sycl_ext_oneapi_virtual_mem] extension). + +==== Extension to `enum class aspect` for physical memory + +[source] +---- +namespace sycl { +enum class aspect { + ... + ext_oneapi_ipc_physical_memory +} +} +---- + +If a SYCL device has this aspect, that device supports the functions in the +`ipc::physical_memory` namespace specified in the following section. + +==== New property for the physical_mem constructor + +This extension adds the following property, which can be passed to the +physical_mem object constructor. + +[source,c++] +---- +namespace sycl::ext::oneapi::experimental { + +struct enable_ipc { + enable_ipc(bool enable = true); +}; +using enable_ipc_key = enable_ipc; + +} // namespace sycl::ext::oneapi::experimental +---- + +This property controls whether the physical memory object can be shared across +processes. + +_Effects:_ Creates a new `enable_ipc` property with a boolean value +indicating whether the physical memory object can be shared across processes. + +_Remarks:_ `physical_mem` constructor throws an exception with the +`errc::feature_not_supported` error code if inter-process sharing is enabled +via `enable_ipc` and the device associated with `physical_mem` does not have +the aspect `aspect::ext_oneapi_ipc_physical_memory`. + +==== New physical_mem member function + +This extension adds a new physical_mem type member function, which can be used +to check if a given physical memory allocation can be passed between processes. + +[source,c++] +---- +namespace sycl::ext::oneapi::experimental { + +class physical_mem { + // ... + bool ext_oneapi_ipc_enabled() const; +}; + +} // namespace sycl::ext::oneapi::experimental +---- + +_Returns:_ True if the physical_mem object was created with inter-process +sharing enabled. + +==== New functions associated with physical memory + +This extension adds new free functions under the `ipc::physical_memory` experimental +namespace. + +``` +namespace sycl::ext::oneapi::experimental::ipc::physical_memory { + +handle get(const physical_mem &phys_mem); + +physical_mem open(const handle_data_t &handle_data, const sycl::context &ctx, + const sycl::device &dev); + +physical_mem open(const handle_data_t &handle_data, const sycl::device &dev); + +physical_mem open(const handle_data_t &handle_data); + +// Requires C++20 +physical_mem open(const handle_data_view_t &handle_data_view, + const sycl::context &ctx, const sycl::device &dev); + +// Requires C++20 +physical_mem open(const handle_data_view_t &handle_data_view, + const sycl::device &dev); + +// Requires C++20 +physical_mem open(const handle_data_view_t &handle_data_view); + +} +``` + +|==== +a| +[frame=all,grid=none] +!==== +a! +[source] +---- +handle get(const physical_mem &phys_mem) +---- +!==== + +_Returns:_ An IPC "handle" to this physical memory object. The bytes of this +handle can be transferred to another process on the same system, and the other +process can use the handle to get a physical memory object representing the +same memory allocation through a call to the `open` function. + +_Throws:_ An exception with the `errc::invalid` error code if `phys_mem` was +not created with inter-process sharing enabled via the `enable_ipc` property. + +!==== +a! +[source] +---- +physical_mem open(const handle_data_t &handle_data, const sycl::context &ctx, + const sycl::device &dev) +---- +!==== + +_Preconditions:_ + + * `handle_data` is the IPC "handle" to a physical memory object + that was returned from a call to the `get` function either in this process + or in some other process on the same system. + * `dev` is the same device that was associated with the `physical_mem` object + passed to the `get` function call that produced `handle_data`. + +_Returns:_ A physical memory object represented by `handle_data`. The returned +object is associated with context `ctx` and device `dev`. + +[_Note:_ The `open` function can be called multiple times on the same handle +within the same process. Each call to the `open` function may return a unique +physical memory object even for the same handle. +_{endnote}_] + +[_Note:_ The physical memory object returned from a call to the `open` function +is no longer valid if the original object associated with `handle_data` (passed +to a `get` function call) is destroyed. +_{endnote}_] + +_Throws:_ + + * An exception with the `errc::feature_not_supported` error code if device + `dev` does not have `aspect::ext_oneapi_ipc_physical_memory`. + * An exception with the `errc::invalid` error code if the handle data + `handle_data` has an unexpected number of bytes. + * An exception with the `errc::invalid` error code if `ctx` does not contain + `dev`. + +!==== +a! +[source] +---- +physical_mem open(const handle_data_t &handle_data, const sycl::device &dev) +---- +!==== + +_Effects:_ Equivalent to: + +[source,c++,indent=2] +---- +sycl::context ctxt = dev.get_platform().khr_get_default_context(); +return ipc::physical_memory::open(handle_data, ctxt, dev); +---- + +!==== +a! +[source] +---- +physical_mem open(const handle_data_t &handle_data) +---- +!==== + +_Effects:_ Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +return ipc::physical_memory::open(handle_data, ctxt, d); +---- + +!==== +a! +[source] +---- +physical_mem open(const handle_data_view_t &handle_data_view, + const sycl::context &ctx, const sycl::device &dev) +---- +!==== + +_Effects:_ Equivalent to: + +[source,c++,indent=2] +---- +handle_data_t handle_data{handle_data_view.begin(), handle_data_view.end()}; +return ipc::physical_memory::open(handle_data, ctx, dev); +---- + +!==== +a! +[source] +---- +physical_mem open(const handle_data_view_t &handle_data_view, + const sycl::device &dev) +---- +!==== + +_Effects:_ Equivalent to: + +[source,c++,indent=2] +---- +sycl::context ctxt = dev.get_platform().khr_get_default_context(); +return ipc::physical_memory::open(handle_data_view, ctxt, dev); +---- + +!==== +a! +[source] +---- +physical_mem open(const handle_data_view_t &handle_data_view) +---- +!==== + +_Effects:_ Equivalent to: + +[source,c++,indent=2] +---- +sycl::device d; +sycl::context ctxt = d.get_platform().khr_get_default_context(); +return ipc::physical_memory::open(handle_data_view, ctxt, d); +---- + +|==== == Issues diff --git a/sycl/doc/extensions/experimental/sycl_ext_oneapi_virtual_mem.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_oneapi_virtual_mem.asciidoc index 0432f01f71681..e86df0a6c2218 100644 --- a/sycl/doc/extensions/experimental/sycl_ext_oneapi_virtual_mem.asciidoc +++ b/sycl/doc/extensions/experimental/sycl_ext_oneapi_virtual_mem.asciidoc @@ -41,6 +41,9 @@ This extension is written against the SYCL 2020 revision 8 specification. All references below to the "core SYCL specification" or to section numbers in the SYCL specification refer to that revision. +This extension also depends on the following other SYCL extensions: + +* link:sycl_ext_oneapi_properties.asciidoc[sycl_ext_oneapi_properties] == Status @@ -242,8 +245,11 @@ enum class address_access_mode : /*unspecified*/ { class physical_mem { public: - physical_mem(const device &syclDevice, const context &syclContext, size_t numBytes); - physical_mem(const queue &syclQueue, size_t numBytes); + template + physical_mem(const device &syclDevice, const context &syclContext, size_t numBytes, + PropertyListT props = {}); + template + physical_mem(const queue &syclQueue, size_t numBytes, PropertyListT props = {}); /* -- common interface members -- */ @@ -265,7 +271,9 @@ public: |============================ |Member function |Description -|`physical_mem(const device &syclDevice, const context &syclContext, size_t numBytes)` | +|`template + physical_mem(const device &syclDevice, const context &syclContext, size_t numBytes, + PropertyListT props = {})` | Constructs a `physical_mem` instance using the `syclDevice` provided. This device must either be contained by `syclContext` or it must be a descendent device of some device that is contained by that context, otherwise this function @@ -281,8 +289,12 @@ an exception with `errc::feature_not_supported`. If the constructor is unable to allocate the required memory on `syclDevice`, the call throws an exception with `errc::memory_allocation`. -|`physical_mem(const queue &syclQueue, size_t numBytes)` | -Same as `physical_mem(syclQueue.get_device(), syclQueue.get_context(), numBytes)`. +`props` can optionally be used to provide the physical memory properties. + +|`template + physical_mem(const queue &syclQueue, size_t numBytes, PropertyListT props = {})` | +Same as `physical_mem(syclQueue.get_device(), syclQueue.get_context(), numBytes, +props)`. |`void *map(uintptr_t ptr, size_t numBytes, address_access_mode mode, size_t offset = 0)` | Maps a virtual memory range, specified by `ptr` and `numBytes`, to the physical