From 65361bf678c70e614bf755e41e6c52c23305b055 Mon Sep 17 00:00:00 2001 From: Zhan Xue Date: Tue, 30 Jun 2026 18:24:46 +0800 Subject: [PATCH 1/3] Add Python bindings for SYCL IPC memory via dpctl Add inter-process communication (IPC) support for SYCL USM memory, enabling zero-copy GPU memory sharing across processes. It wraps sycl::ext::oneapi::experimental::ipc::memory (get/open/close/put). C API (libsyclinterface): - dpctl_sycl_ipc_memory_interface.h: declares DPCTLIPCMem_GetHandle, DPCTLIPCMem_OpenHandle, DPCTLIPCMem_CloseHandle, DPCTLIPCMem_FreeHandleData - dpctl_sycl_ipc_memory_interface.cpp: implements the C API by calling the SYCL experimental ipc::memory functions; auto-discovered by CMake. Cython declarations: - _backend.pxd: extern block for the 4 new C functions Python subpackage (dpctl.ipc): - IPCMemoryHandle: Cython extension class wrapping IPC memory export/import __init__(usm_memory): calls get() + put(), stores handle as bytes to_bytes(): serializable payload for cross-process transport open(handle_bytes, device, nbytes): returns MemoryUSMDevice close_mapping(usm_memory): explicitly closes an IPC mapping Signed-off-by: Zhan Xue --- dpctl/CMakeLists.txt | 1 + dpctl/_backend.pxd | 16 ++ dpctl/ipc/CMakeLists.txt | 10 + dpctl/ipc/__init__.pxd | 18 ++ dpctl/ipc/__init__.py | 82 ++++++ dpctl/ipc/_ipc_memory.pxd | 32 +++ dpctl/ipc/_ipc_memory.pyx | 253 ++++++++++++++++++ libsyclinterface/CMakeLists.txt | 38 +++ .../syclinterface/Config/dpctl_config.h.in | 2 + .../dpctl_sycl_ipc_memory_interface.h | 105 ++++++++ .../dpctl_sycl_ipc_memory_interface.cpp | 172 ++++++++++++ setup.py | 2 + 12 files changed, 731 insertions(+) create mode 100644 dpctl/ipc/CMakeLists.txt create mode 100644 dpctl/ipc/__init__.pxd create mode 100644 dpctl/ipc/__init__.py create mode 100644 dpctl/ipc/_ipc_memory.pxd create mode 100644 dpctl/ipc/_ipc_memory.pyx create mode 100644 libsyclinterface/include/syclinterface/dpctl_sycl_ipc_memory_interface.h create mode 100644 libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp diff --git a/dpctl/CMakeLists.txt b/dpctl/CMakeLists.txt index a24c7443f9..508e484b36 100644 --- a/dpctl/CMakeLists.txt +++ b/dpctl/CMakeLists.txt @@ -206,3 +206,4 @@ target_link_libraries(DpctlCAPI INTERFACE ${_trgt}_headers) add_subdirectory(program) add_subdirectory(memory) add_subdirectory(utils) +add_subdirectory(ipc) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 2f8911fdc5..25a9c67eb4 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -606,3 +606,19 @@ cdef extern from "syclinterface/dpctl_sycl_extension_interface.h": DPCTLSyclRawKernelArgRef Ref) cdef bint DPCTLRawKernelArg_Available() + +cdef extern from "syclinterface/dpctl_sycl_ipc_memory_interface.h": + cdef int DPCTLIPCMem_GetHandle( + DPCTLSyclUSMRef Ptr, + DPCTLSyclContextRef CRef, + char **DataOut, + size_t *SizeOut) + cdef DPCTLSyclUSMRef DPCTLIPCMem_OpenHandle( + const char *HandleData, + size_t HandleDataSize, + DPCTLSyclContextRef CRef, + DPCTLSyclDeviceRef DRef) + cdef void DPCTLIPCMem_CloseHandle( + DPCTLSyclUSMRef MappedPtr, + DPCTLSyclContextRef CRef) + cdef void DPCTLIPCMem_FreeHandleData(char *Data) diff --git a/dpctl/ipc/CMakeLists.txt b/dpctl/ipc/CMakeLists.txt new file mode 100644 index 0000000000..d50b329cfd --- /dev/null +++ b/dpctl/ipc/CMakeLists.txt @@ -0,0 +1,10 @@ +if(DPCTL_HAS_IPC_MEMORY) + set(_cy_file ${CMAKE_CURRENT_SOURCE_DIR}/_ipc_memory.pyx) + get_filename_component(_trgt ${_cy_file} NAME_WLE) + build_dpctl_ext(${_trgt} ${_cy_file} "dpctl/ipc" RELATIVE_PATH "..") + target_link_libraries(DpctlCAPI INTERFACE ${_trgt}_headers) +endif() + +# Always install __init__.py (handles graceful fallback when IPC not supported) +install(FILES ${CMAKE_CURRENT_SOURCE_DIR}/__init__.py + DESTINATION ${CMAKE_INSTALL_PREFIX}/dpctl/ipc) diff --git a/dpctl/ipc/__init__.pxd b/dpctl/ipc/__init__.pxd new file mode 100644 index 0000000000..2231abca1e --- /dev/null +++ b/dpctl/ipc/__init__.pxd @@ -0,0 +1,18 @@ +# Data Parallel Control (dpctl) +# +# Copyright 2020-2026 Intel Corporation +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +# distutils: language = c++ +# cython: language_level=3 diff --git a/dpctl/ipc/__init__.py b/dpctl/ipc/__init__.py new file mode 100644 index 0000000000..81c21ed2b7 --- /dev/null +++ b/dpctl/ipc/__init__.py @@ -0,0 +1,82 @@ +# Data Parallel Control (dpctl) +# +# Copyright 2020-2026 Intel Corporation +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +""" +**Data Parallel Control IPC** provides Python objects for inter-process +communication of SYCL USM memory. + +- :class:`IPCMemoryHandle` wraps ``sycl::ext::oneapi::experimental::ipc::memory`` + to export/import USM device pointers across processes. + +Requires oneAPI DPC++ compiler >= 2026.1 (with SYCL IPC memory runtime support). +""" + +_IPC_MEMORY_AVAILABLE = False +_IPC_MEMORY_ERROR = "" + +try: + from ._ipc_memory import IPCMemoryHandle + _IPC_MEMORY_AVAILABLE = True +except ImportError as _e: + _IPC_MEMORY_ERROR = ( + "dpctl.ipc.IPCMemoryHandle is not available. " + "SYCL IPC memory support was not detected at build time. " + "This requires oneAPI DPC++ compiler >= 2026.0 with " + "sycl::ext::oneapi::experimental::ipc::memory runtime support. " + f"(Import error: {_e})" + ) + + class IPCMemoryHandle: + """Placeholder for unavailable IPC memory support.""" + + def __init__(self, *args, **kwargs): + raise RuntimeError(_IPC_MEMORY_ERROR) + + @staticmethod + def open(*args, **kwargs): + raise RuntimeError(_IPC_MEMORY_ERROR) + + @staticmethod + def close_mapping(*args, **kwargs): + raise RuntimeError(_IPC_MEMORY_ERROR) + + +def is_ipc_memory_supported(): + """Return True if IPC memory is supported by the current SYCL runtime. + + Returns: + bool: True if IPCMemoryHandle can be used, False otherwise. + """ + return _IPC_MEMORY_AVAILABLE + + +def check_ipc_memory_support(): + """Raise RuntimeError if IPC memory is not supported. + + Use this for early failure at application startup. + + Raises: + RuntimeError: if the SYCL IPC memory API is not available. + """ + if not _IPC_MEMORY_AVAILABLE: + raise RuntimeError(_IPC_MEMORY_ERROR) + + +__all__ = [ + "IPCMemoryHandle", + "is_ipc_memory_supported", + "check_ipc_memory_support", +] diff --git a/dpctl/ipc/_ipc_memory.pxd b/dpctl/ipc/_ipc_memory.pxd new file mode 100644 index 0000000000..5d01cf3b24 --- /dev/null +++ b/dpctl/ipc/_ipc_memory.pxd @@ -0,0 +1,32 @@ +# Data Parallel Control (dpctl) +# +# Copyright 2020-2026 Intel Corporation +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +# distutils: language = c++ +# cython: language_level=3 + +"""Declarations for the IPCMemoryHandle Cython extension type.""" + +from .._backend cimport DPCTLSyclContextRef, DPCTLSyclDeviceRef, DPCTLSyclUSMRef +from .._sycl_context cimport SyclContext +from .._sycl_device cimport SyclDevice +from .._sycl_queue cimport SyclQueue +from ..memory._memory cimport _Memory, MemoryUSMDevice + + +cdef class IPCMemoryHandle: + cdef bytes _handle_bytes + cdef SyclContext _ctx + cdef bint _closed diff --git a/dpctl/ipc/_ipc_memory.pyx b/dpctl/ipc/_ipc_memory.pyx new file mode 100644 index 0000000000..bb07d65d52 --- /dev/null +++ b/dpctl/ipc/_ipc_memory.pyx @@ -0,0 +1,253 @@ +# Data Parallel Control (dpctl) +# +# Copyright 2020-2026 Intel Corporation +# +# Licensed under the Apache License, Version 2.0 (the "License"); +# you may not use this file except in compliance with the License. +# You may obtain a copy of the License at +# +# http://www.apache.org/licenses/LICENSE-2.0 +# +# Unless required by applicable law or agreed to in writing, software +# distributed under the License is distributed on an "AS IS" BASIS, +# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +# See the License for the specific language governing permissions and +# limitations under the License. + +# distutils: language = c++ +# cython: language_level=3 +# cython: linetrace=True + +"""Python bindings for SYCL IPC memory +(``sycl::ext::oneapi::experimental::ipc::memory``). + +Typical usage for exporting a USM pointer:: + + import dpctl + from dpctl.memory import MemoryUSMDevice + from dpctl.ipc import IPCMemoryHandle + + # Sender process + mem = MemoryUSMDevice(1024) + handle = IPCMemoryHandle(mem) + raw = handle.to_bytes() # send *raw* to another process + + # Receiver process (same system, different PID) + remote_mem = IPCMemoryHandle.open(raw, device) +""" + +from cpython.bytes cimport PyBytes_AS_STRING, PyBytes_FromStringAndSize +from libc.stdlib cimport free as libc_free + +from dpctl._backend cimport ( + DPCTLSyclContextRef, + DPCTLSyclDeviceRef, + DPCTLSyclUSMRef, + DPCTLIPCMem_GetHandle, + DPCTLIPCMem_OpenHandle, + DPCTLIPCMem_CloseHandle, + DPCTLIPCMem_FreeHandleData, + DPCTLQueue_GetContext, + DPCTLQueue_Delete, + DPCTLContext_Delete, + DPCTLDevice_Copy, +) + +from .._sycl_context cimport SyclContext +from .._sycl_device cimport SyclDevice +from .._sycl_queue cimport SyclQueue +from ..memory._memory cimport _Memory, MemoryUSMDevice + +import dpctl + + +__all__ = ["IPCMemoryHandle"] + + +cdef class IPCMemoryHandle: + """Wrapper around a SYCL IPC memory handle. + + Instances are created by passing a :class:`dpctl.memory.MemoryUSMDevice` + (or any ``_Memory`` subclass backed by a device USM pointer) to the + constructor. The resulting object exposes :meth:`to_bytes` which + returns an opaque ``bytes`` payload suitable for inter-process + transport (e.g. via pickle, ZMQ, shared-memory). + + On the receiving side, call :meth:`IPCMemoryHandle.open` with the + payload and a target device to obtain a + :class:`dpctl.memory.MemoryUSMDevice` backed by the IPC-mapped + memory. + + Parameters + ---------- + usm_memory : dpctl.memory._Memory + USM memory object whose device pointer to export. + context : dpctl.SyclContext, optional + SYCL context to use. Defaults to the context of *usm_memory*'s + queue. + + Raises + ------ + TypeError + If *usm_memory* is not a ``_Memory`` instance. + RuntimeError + If the SYCL runtime fails to produce an IPC handle. + """ + + def __cinit__(self): + self._handle_bytes = None + self._ctx = None + self._closed = False + + def __init__(self, _Memory usm_memory not None, SyclContext context=None): + cdef DPCTLSyclUSMRef ptr = usm_memory.get_data_ptr() + if ptr is NULL: + raise ValueError("USM memory object has a null pointer") + + cdef SyclQueue q = usm_memory.queue + if context is None: + context = q.sycl_context + + cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() + cdef char *data_out = NULL + cdef size_t size_out = 0 + + cdef int rc = DPCTLIPCMem_GetHandle(ptr, ctx_ref, &data_out, &size_out) + if rc != 0: + raise RuntimeError( + "DPCTLIPCMem_GetHandle failed — device may not support " + "aspect::ext_oneapi_ipc_memory" + ) + + try: + self._handle_bytes = PyBytes_FromStringAndSize(data_out, + size_out) + finally: + DPCTLIPCMem_FreeHandleData(data_out) + + self._ctx = context + self._closed = False + + def to_bytes(self): + """Return the raw IPC handle data as ``bytes``. + + The returned object can be pickled, sent over a socket, or + written to shared memory for another process to consume via + :meth:`open`. + """ + if self._closed: + raise RuntimeError("IPC handle has already been closed") + return self._handle_bytes + + @staticmethod + def open(bytes handle_bytes not None, + SyclDevice device not None, + SyclContext context=None, + Py_ssize_t nbytes=0): + """Open an IPC handle in this process. + + Parameters + ---------- + handle_bytes : bytes + Opaque payload from :meth:`to_bytes` (possibly from another + process). + device : dpctl.SyclDevice + Device to map the memory on. + context : dpctl.SyclContext, optional + SYCL context to use. Defaults to the default context for + *device*'s platform. + nbytes : int, optional + Byte size of the original allocation. If 0, the size is + determined by the driver (if supported). + + Returns + ------- + dpctl.memory.MemoryUSMDevice + A USM device memory object backed by the IPC-mapped pointer. + The mapping is closed when the returned object is garbage + collected. + + Raises + ------ + RuntimeError + If the handle cannot be opened. + """ + cdef const char *raw = PyBytes_AS_STRING(handle_bytes) + cdef size_t raw_size = len(handle_bytes) + + if context is None: + context = device.sycl_platform.default_context + + cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() + cdef DPCTLSyclDeviceRef dev_ref = device.get_device_ref() + + cdef DPCTLSyclUSMRef mapped_ptr = DPCTLIPCMem_OpenHandle( + raw, raw_size, ctx_ref, dev_ref) + if mapped_ptr is NULL: + raise RuntimeError("DPCTLIPCMem_OpenHandle failed") + + # Build a MemoryUSMDevice around the mapped pointer. + # Use the device-cached queue so dpctl tracks the allocation. + cdef SyclQueue q + try: + q = dpctl.SyclQueue(context, device) + except Exception: + DPCTLIPCMem_CloseHandle(mapped_ptr, ctx_ref) + raise + + # Wrap as MemoryUSMDevice — nbytes must be known by the caller + # or the driver. We require nbytes > 0 for safety. + if nbytes <= 0: + DPCTLIPCMem_CloseHandle(mapped_ptr, ctx_ref) + raise ValueError("nbytes must be > 0 for IPC open") + + cdef object mem = MemoryUSMDevice.create_from_usm_pointer_size_qref( + mapped_ptr, nbytes, q.get_queue_ref()) + return mem + + @staticmethod + def close_mapping(_Memory usm_memory not None, + SyclContext context=None): + """Explicitly close an IPC mapping. + + After calling this, *usm_memory* is invalidated and must not be + used again. Its destructor will not attempt to free the pointer. + + Parameters + ---------- + usm_memory : dpctl.memory._Memory + The memory object returned by :meth:`open`. + context : dpctl.SyclContext, optional + Context used when opening. Defaults to the memory's queue + context. + """ + cdef DPCTLSyclUSMRef ptr = usm_memory.get_data_ptr() + if ptr is NULL: + return + + if context is None: + context = usm_memory.queue.sycl_context + + cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() + DPCTLIPCMem_CloseHandle(ptr, ctx_ref) + + # Prevent the _Memory destructor from calling sycl::free on the + # now-unmapped pointer. Setting _opaque_ptr to NULL makes + # __dealloc__ skip OpaqueSmartPtr_Delete. + usm_memory._opaque_ptr = NULL + usm_memory._memory_ptr = NULL + usm_memory.nbytes = 0 + + def close(self): + """Mark this handle as closed (driver resources already released + during construction via ``put``).""" + self._closed = True + + def __dealloc__(self): + self._closed = True + + def __repr__(self): + return ( + f"IPCMemoryHandle(size={len(self._handle_bytes)}, " + f"closed={self._closed})" + ) diff --git a/libsyclinterface/CMakeLists.txt b/libsyclinterface/CMakeLists.txt index b4331bffbb..8225ad7c13 100644 --- a/libsyclinterface/CMakeLists.txt +++ b/libsyclinterface/CMakeLists.txt @@ -111,6 +111,35 @@ endif() message(STATUS "LIB_ZE: ${LIBZE_LOADER_FILENAME}") message(STATUS "LIB_CL: ${LIBCL_LOADER_FILENAME}") + +# --- IPC Memory support detection --- +option(DPCTL_ENABLE_IPC_MEMORY + "Enable IPC memory support (requires SYCL IPC runtime)" + ON +) +if(DPCTL_ENABLE_IPC_MEMORY) + include(CheckCXXSourceCompiles) + set(CMAKE_REQUIRED_FLAGS "${SYCL_FLAGS}") + set(CMAKE_REQUIRED_INCLUDES "${SYCL_INCLUDE_DIR}") + check_cxx_source_compiles(" + #include + int main() { return 0; } + " DPCTL_IPC_MEMORY_HEADER_FOUND) + if(DPCTL_IPC_MEMORY_HEADER_FOUND) + # Check if libsycl.so has the openIPCMemHandle symbol + set(DPCTL_HAS_IPC_MEMORY 1) + set(DPCTL_HAS_IPC_MEMORY 1 PARENT_SCOPE) + message(STATUS "SYCL IPC memory support: ENABLED (ipc_memory.hpp found)") + else() + message(STATUS "SYCL IPC memory support: DISABLED (ipc_memory.hpp not found)") + endif() + unset(CMAKE_REQUIRED_FLAGS) + unset(CMAKE_REQUIRED_INCLUDES) + unset(CMAKE_REQUIRED_LIBRARIES) +else() + message(STATUS "SYCL IPC memory support: DISABLED (DPCTL_ENABLE_IPC_MEMORY=OFF)") +endif() + configure_file( ${CMAKE_CURRENT_SOURCE_DIR}/include/syclinterface/Config/dpctl_config.h.in ${CMAKE_CURRENT_SOURCE_DIR}/include/syclinterface/Config/dpctl_config.h @@ -222,8 +251,17 @@ file(GLOB_RECURSE sources list(REMOVE_ITEM sources "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_vector_templ.cpp" + ) +# Conditionally exclude IPC memory source if not supported +if(NOT DPCTL_HAS_IPC_MEMORY) + list(REMOVE_ITEM sources + "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_sycl_ipc_memory_interface.cpp" + ) + message(STATUS "Excluding dpctl_sycl_ipc_memory_interface.cpp from build") +endif() + file(GLOB_RECURSE helper_sources ${CMAKE_CURRENT_SOURCE_DIR}/helper/source/*.cpp ) diff --git a/libsyclinterface/include/syclinterface/Config/dpctl_config.h.in b/libsyclinterface/include/syclinterface/Config/dpctl_config.h.in index 0568f38507..97e28d1219 100644 --- a/libsyclinterface/include/syclinterface/Config/dpctl_config.h.in +++ b/libsyclinterface/include/syclinterface/Config/dpctl_config.h.in @@ -35,3 +35,5 @@ #define DPCTL_LIBZE_LOADER_FILENAME "@LIBZE_LOADER_FILENAME@" #define DPCTL_LIBCL_LOADER_FILENAME "@LIBCL_LOADER_FILENAME@" +/* Defined when the SYCL runtime supports IPC memory (ipc::memory API). */ +#cmakedefine DPCTL_HAS_IPC_MEMORY 1 diff --git a/libsyclinterface/include/syclinterface/dpctl_sycl_ipc_memory_interface.h b/libsyclinterface/include/syclinterface/dpctl_sycl_ipc_memory_interface.h new file mode 100644 index 0000000000..a62265af1c --- /dev/null +++ b/libsyclinterface/include/syclinterface/dpctl_sycl_ipc_memory_interface.h @@ -0,0 +1,105 @@ +//===- dpctl_sycl_ipc_memory_interface.h - C API for SYCL IPC mem -*-C++-*-===// +// +// Data Parallel Control (dpctl) +// +// Copyright 2020-2026 Intel Corporation +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. +// +//===----------------------------------------------------------------------===// +/// +/// \file +/// This header declares a C interface to +/// sycl::ext::oneapi::experimental::ipc::memory functions. +/// +//===----------------------------------------------------------------------===// + +#pragma once + +#include "Support/DllExport.h" +#include "Support/ExternC.h" +#include "Support/MemOwnershipAttrs.h" +#include "dpctl_data_types.h" +#include "dpctl_sycl_types.h" + +DPCTL_C_EXTERN_C_BEGIN + +/** + * @defgroup IPCMemoryInterface IPC Memory Interface + */ + +/*! + * @brief Get an IPC memory handle for a USM device pointer. + * + * Wraps ``sycl::ext::oneapi::experimental::ipc::memory::get()``. + * The returned handle bytes are copied out and the driver-side handle + * resource is released (via ``put``) before returning. + * + * @param Ptr USM device pointer to export. + * @param CRef Sycl context associated with the pointer. + * @param DataOut [out] Pointer to receive a malloc'd byte buffer + * containing the IPC handle data. Caller must free + * with DPCTLIPCMem_FreeHandleData(). + * @param SizeOut [out] Pointer to receive the byte count of DataOut. + * @return 0 on success, non-zero on failure. + * @ingroup IPCMemoryInterface + */ +DPCTL_API +int DPCTLIPCMem_GetHandle(__dpctl_keep DPCTLSyclUSMRef Ptr, + __dpctl_keep const DPCTLSyclContextRef CRef, + char **DataOut, + size_t *SizeOut); + +/*! + * @brief Open an IPC memory handle in the receiving process. + * + * Wraps ``sycl::ext::oneapi::experimental::ipc::memory::open()``. + * + * @param HandleData Byte buffer from DPCTLIPCMem_GetHandle. + * @param HandleDataSize Size of HandleData in bytes. + * @param CRef Sycl context for the receiving side. + * @param DRef Sycl device to map the memory on. + * @return A USM pointer to the IPC-mapped memory, or nullptr on failure. + * @ingroup IPCMemoryInterface + */ +DPCTL_API +__dpctl_give DPCTLSyclUSMRef +DPCTLIPCMem_OpenHandle(const char *HandleData, + size_t HandleDataSize, + __dpctl_keep const DPCTLSyclContextRef CRef, + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Close an IPC memory mapping opened by DPCTLIPCMem_OpenHandle. + * + * Wraps ``sycl::ext::oneapi::experimental::ipc::memory::close()``. + * + * @param MappedPtr The USM pointer returned by DPCTLIPCMem_OpenHandle. + * @param CRef Sycl context used when opening the handle. + * @ingroup IPCMemoryInterface + */ +DPCTL_API +void DPCTLIPCMem_CloseHandle(__dpctl_keep DPCTLSyclUSMRef MappedPtr, + __dpctl_keep const DPCTLSyclContextRef CRef); + +/*! + * @brief Free a handle data buffer returned by DPCTLIPCMem_GetHandle. + * + * @param Data Pointer previously returned via the DataOut parameter + * of DPCTLIPCMem_GetHandle. + * @ingroup IPCMemoryInterface + */ +DPCTL_API +void DPCTLIPCMem_FreeHandleData(char *Data); + +DPCTL_C_EXTERN_C_END diff --git a/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp b/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp new file mode 100644 index 0000000000..babf00f002 --- /dev/null +++ b/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp @@ -0,0 +1,172 @@ +//===- dpctl_sycl_ipc_memory_interface.cpp - C API for SYCL IPC memory ----===// +// +// Data Parallel Control (dpctl) +// +// Copyright 2020-2026 Intel Corporation +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. +// +//===----------------------------------------------------------------------===// +/// +/// \file +/// This file implements the functions declared in +/// dpctl_sycl_ipc_memory_interface.h. +/// +//===----------------------------------------------------------------------===// + +#include "dpctl_sycl_ipc_memory_interface.h" +#include "Config/dpctl_config.h" +#include "dpctl_error_handlers.h" +#include "dpctl_sycl_type_casters.hpp" +#include +#include +#include +#include +#include + +using namespace sycl; + +// Support both the new namespace (ipc::memory, oneAPI >= 2026.1) and the +// deprecated namespace (ipc_memory, oneAPI 2026.0). +#if __has_include() +// New layout: ipc::memory namespace with separate ipc::handle/handle_data_t +namespace ipc = sycl::ext::oneapi::experimental::ipc::memory; +using ipc_handle_data_t = sycl::ext::oneapi::experimental::ipc::handle_data_t; +#else +// Old layout: everything in ipc_memory namespace +namespace ipc = sycl::ext::oneapi::experimental::ipc_memory; +using ipc_handle_data_t = ipc::handle_data_t; +#endif + +namespace +{ +static_assert(__SYCL_COMPILER_VERSION >= __SYCL_COMPILER_VERSION_REQUIRED, + "The compiler does not meet minimum version requirement"); + +using namespace dpctl::syclinterface; +} // end of anonymous namespace + +int DPCTLIPCMem_GetHandle(__dpctl_keep DPCTLSyclUSMRef Ptr, + __dpctl_keep const DPCTLSyclContextRef CRef, + char **DataOut, + size_t *SizeOut) +{ + if (!Ptr) { + error_handler("Input Ptr is nullptr.", __FILE__, __func__, __LINE__); + return 1; + } + if (!CRef) { + error_handler("Input CRef is nullptr.", __FILE__, __func__, __LINE__); + return 1; + } + if (!DataOut || !SizeOut) { + error_handler("Output pointers are nullptr.", __FILE__, __func__, + __LINE__); + return 1; + } + + try { + auto *RawPtr = unwrap(Ptr); + auto *Ctx = unwrap(CRef); + + // Obtain the IPC handle from the SYCL runtime. + auto Handle = ipc::get(RawPtr, *Ctx); + + // Copy handle data into a malloc'd buffer for the caller. + auto HandleData = Handle.data(); // std::vector + + // Release driver-side resources immediately; the bytes are sufficient + // for the receiver to call open(). + ipc::put(Handle, *Ctx); + + size_t Size = HandleData.size(); + char *Buf = static_cast(std::malloc(Size)); + if (!Buf) { + error_handler("Failed to allocate handle data buffer.", __FILE__, + __func__, __LINE__); + return 1; + } + std::memcpy(Buf, HandleData.data(), Size); + + *DataOut = Buf; + *SizeOut = Size; + return 0; + } catch (std::exception const &e) { + error_handler(e, __FILE__, __func__, __LINE__); + return 1; + } +} + +__dpctl_give DPCTLSyclUSMRef +DPCTLIPCMem_OpenHandle(const char *HandleData, + size_t HandleDataSize, + __dpctl_keep const DPCTLSyclContextRef CRef, + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + if (!HandleData) { + error_handler("Input HandleData is nullptr.", __FILE__, __func__, + __LINE__); + return nullptr; + } + if (!CRef) { + error_handler("Input CRef is nullptr.", __FILE__, __func__, __LINE__); + return nullptr; + } + if (!DRef) { + error_handler("Input DRef is nullptr.", __FILE__, __func__, __LINE__); + return nullptr; + } + + try { + auto *Ctx = unwrap(CRef); + auto *Dev = unwrap(DRef); + + // Rebuild handle_data_t (vector) from the raw byte buffer. + ipc_handle_data_t HData( + reinterpret_cast(HandleData), + reinterpret_cast(HandleData) + HandleDataSize); + + void *MappedPtr = ipc::open(HData, *Ctx, *Dev); + return wrap(MappedPtr); + } catch (std::exception const &e) { + error_handler(e, __FILE__, __func__, __LINE__); + return nullptr; + } +} + +void DPCTLIPCMem_CloseHandle(__dpctl_keep DPCTLSyclUSMRef MappedPtr, + __dpctl_keep const DPCTLSyclContextRef CRef) +{ + if (!MappedPtr) { + error_handler("Input MappedPtr is nullptr.", __FILE__, __func__, + __LINE__); + return; + } + if (!CRef) { + error_handler("Input CRef is nullptr.", __FILE__, __func__, __LINE__); + return; + } + + try { + auto *RawPtr = unwrap(MappedPtr); + auto *Ctx = unwrap(CRef); + ipc::close(RawPtr, *Ctx); + } catch (std::exception const &e) { + error_handler(e, __FILE__, __func__, __LINE__); + } +} + +void DPCTLIPCMem_FreeHandleData(char *Data) +{ + std::free(Data); +} diff --git a/setup.py b/setup.py index f7c5d3ef9c..d4adc3d94c 100644 --- a/setup.py +++ b/setup.py @@ -28,6 +28,7 @@ "dpctl.memory", "dpctl.program", "dpctl.utils", + "dpctl.ipc", ], package_data={ "dpctl": [ @@ -47,6 +48,7 @@ "*.pxd", "memory/*.pxd", "program/*.pxd", + "ipc/*.pxd", ] }, include_package_data=False, From 3dea37996de9b16771225297ce114d9a952f9e36 Mon Sep 17 00:00:00 2001 From: Zhan Xue Date: Thu, 2 Jul 2026 04:26:11 +0800 Subject: [PATCH 2/3] Add aspect::ext_oneapi_ipc_memory for SyclDevice add has_aspect_ext_oneapi_ipc_memory checks in __init__ and open() Signed-off-by: Zhan Xue --- dpctl/_backend.pxd | 1 + dpctl/_sycl_device.pyx | 12 ++++++++++++ dpctl/ipc/_ipc_memory.pyx | 11 +++++++++++ .../helper/source/dpctl_utils_helper.cpp | 10 ++++++++++ .../include/syclinterface/dpctl_sycl_enum_types.h | 3 ++- 5 files changed, 36 insertions(+), 1 deletion(-) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 25a9c67eb4..fe624cb189 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -101,6 +101,7 @@ cdef extern from "syclinterface/dpctl_sycl_enum_types.h": _emulated "emulated", _is_component "is_component", _is_composite "is_composite", + _ext_oneapi_ipc_memory "ext_oneapi_ipc_memory", ctypedef enum _partition_affinity_domain_type \ "DPCTLPartitionAffinityDomainType": diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index deda5a6ff8..506664ab18 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -885,6 +885,18 @@ cdef class SyclDevice(_SyclDevice): cdef _aspect_type AT = _aspect_type._is_composite return DPCTLDevice_HasAspect(self._device_ref, AT) + @property + def has_aspect_ext_oneapi_ipc_memory(self): + """ Returns ``True`` if this device supports inter-process + communication (IPC) memory handles, ``False`` otherwise. + + Returns: + bool: + Indicates if device supports IPC memory. + """ + cdef _aspect_type AT = _aspect_type._ext_oneapi_ipc_memory + return DPCTLDevice_HasAspect(self._device_ref, AT) + @property def image_2d_max_width(self): """ Returns the maximum width of a 2D image or 1D image in pixels. diff --git a/dpctl/ipc/_ipc_memory.pyx b/dpctl/ipc/_ipc_memory.pyx index bb07d65d52..1c6d6ed9d3 100644 --- a/dpctl/ipc/_ipc_memory.pyx +++ b/dpctl/ipc/_ipc_memory.pyx @@ -105,6 +105,12 @@ cdef class IPCMemoryHandle: raise ValueError("USM memory object has a null pointer") cdef SyclQueue q = usm_memory.queue + cdef SyclDevice dev = q.sycl_device + if not dev.has_aspect_ext_oneapi_ipc_memory: + raise RuntimeError( + "Device does not support aspect::ext_oneapi_ipc_memory" + ) + if context is None: context = q.sycl_context @@ -175,6 +181,11 @@ cdef class IPCMemoryHandle: cdef const char *raw = PyBytes_AS_STRING(handle_bytes) cdef size_t raw_size = len(handle_bytes) + if not device.has_aspect_ext_oneapi_ipc_memory: + raise RuntimeError( + "Device does not support aspect::ext_oneapi_ipc_memory" + ) + if context is None: context = device.sycl_platform.default_context diff --git a/libsyclinterface/helper/source/dpctl_utils_helper.cpp b/libsyclinterface/helper/source/dpctl_utils_helper.cpp index 3b7a1eb3c8..f060598b31 100644 --- a/libsyclinterface/helper/source/dpctl_utils_helper.cpp +++ b/libsyclinterface/helper/source/dpctl_utils_helper.cpp @@ -224,6 +224,9 @@ std::string DPCTL_AspectToStr(aspect aspectTy) case aspect::ext_oneapi_is_composite: ss << "is_composite"; break; + case aspect::ext_oneapi_ipc_memory: + ss << "ext_oneapi_ipc_memory"; + break; default: throw std::runtime_error("Unsupported aspect type"); } @@ -299,6 +302,9 @@ aspect DPCTL_StrToAspectType(const std::string &aspectTyStr) else if (aspectTyStr == "is_composite") { aspectTy = aspect::ext_oneapi_is_composite; } + else if (aspectTyStr == "ext_oneapi_ipc_memory") { + aspectTy = aspect::ext_oneapi_ipc_memory; + } else { // \todo handle the error throw std::runtime_error("Unsupported aspect type"); @@ -351,6 +357,8 @@ aspect DPCTL_DPCTLAspectTypeToSyclAspect(DPCTLSyclAspectType AspectTy) return aspect::ext_oneapi_is_component; case DPCTLSyclAspectType::is_composite: return aspect::ext_oneapi_is_composite; + case DPCTLSyclAspectType::ext_oneapi_ipc_memory: + return aspect::ext_oneapi_ipc_memory; default: throw std::runtime_error("Unsupported aspect type"); } @@ -401,6 +409,8 @@ DPCTLSyclAspectType DPCTL_SyclAspectToDPCTLAspectType(aspect Aspect) return DPCTLSyclAspectType::is_composite; case aspect::ext_oneapi_is_component: return DPCTLSyclAspectType::is_component; + case aspect::ext_oneapi_ipc_memory: + return DPCTLSyclAspectType::ext_oneapi_ipc_memory; default: throw std::runtime_error("Unsupported aspect type"); } diff --git a/libsyclinterface/include/syclinterface/dpctl_sycl_enum_types.h b/libsyclinterface/include/syclinterface/dpctl_sycl_enum_types.h index e8c4cba7e9..e6ee311980 100644 --- a/libsyclinterface/include/syclinterface/dpctl_sycl_enum_types.h +++ b/libsyclinterface/include/syclinterface/dpctl_sycl_enum_types.h @@ -133,7 +133,8 @@ typedef enum host_debuggable, emulated, is_component, - is_composite + is_composite, + ext_oneapi_ipc_memory } DPCTLSyclAspectType; /*! From 83bb9b4478a2428b9181c4baf534cdb3b4e670cf Mon Sep 17 00:00:00 2001 From: Zhan Xue Date: Thu, 2 Jul 2026 18:07:48 +0800 Subject: [PATCH 3/3] Move IPC memory API from dpctl.ipc to dpctl.memory Migrate the SYCL IPC memory functionality from the standalone dpctl.ipc.IPCMemoryHandle class into methods on dpctl.memory._Memory: - _Memory.get_ipc_handle() - _Memory.open_ipc_handle(...) - _Memory.close_ipc_mapping() Signed-off-by: Zhan Xue --- dpctl/CMakeLists.txt | 1 - dpctl/_backend.pxd | 1 + dpctl/ipc/CMakeLists.txt | 10 - dpctl/ipc/__init__.pxd | 18 -- dpctl/ipc/__init__.py | 82 ------ dpctl/ipc/_ipc_memory.pxd | 32 --- dpctl/ipc/_ipc_memory.pyx | 264 ------------------ dpctl/memory/_memory.pyx | 152 ++++++++++ libsyclinterface/CMakeLists.txt | 15 +- .../helper/source/dpctl_utils_helper.cpp | 8 + .../dpctl_sycl_extension_interface.h | 3 + .../source/dpctl_sycl_extension_interface.cpp | 10 + .../dpctl_sycl_ipc_memory_interface.cpp | 14 +- .../source/dpctl_sycl_ipc_memory_stubs.cpp | 60 ++++ setup.py | 2 - 15 files changed, 249 insertions(+), 423 deletions(-) delete mode 100644 dpctl/ipc/CMakeLists.txt delete mode 100644 dpctl/ipc/__init__.pxd delete mode 100644 dpctl/ipc/__init__.py delete mode 100644 dpctl/ipc/_ipc_memory.pxd delete mode 100644 dpctl/ipc/_ipc_memory.pyx create mode 100644 libsyclinterface/source/dpctl_sycl_ipc_memory_stubs.cpp diff --git a/dpctl/CMakeLists.txt b/dpctl/CMakeLists.txt index 508e484b36..a24c7443f9 100644 --- a/dpctl/CMakeLists.txt +++ b/dpctl/CMakeLists.txt @@ -206,4 +206,3 @@ target_link_libraries(DpctlCAPI INTERFACE ${_trgt}_headers) add_subdirectory(program) add_subdirectory(memory) add_subdirectory(utils) -add_subdirectory(ipc) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index fe624cb189..91629e2947 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -596,6 +596,7 @@ cdef extern from "syclinterface/dpctl_sycl_extension_interface.h": DPCTLSyclWorkGroupMemoryRef Ref) cdef bint DPCTLWorkGroupMemory_Available() + cdef bint DPCTLIPCMem_Available() cdef struct DPCTLOpaqueRawKernelArg ctypedef DPCTLOpaqueRawKernelArg *DPCTLSyclRawKernelArgRef diff --git a/dpctl/ipc/CMakeLists.txt b/dpctl/ipc/CMakeLists.txt deleted file mode 100644 index d50b329cfd..0000000000 --- a/dpctl/ipc/CMakeLists.txt +++ /dev/null @@ -1,10 +0,0 @@ -if(DPCTL_HAS_IPC_MEMORY) - set(_cy_file ${CMAKE_CURRENT_SOURCE_DIR}/_ipc_memory.pyx) - get_filename_component(_trgt ${_cy_file} NAME_WLE) - build_dpctl_ext(${_trgt} ${_cy_file} "dpctl/ipc" RELATIVE_PATH "..") - target_link_libraries(DpctlCAPI INTERFACE ${_trgt}_headers) -endif() - -# Always install __init__.py (handles graceful fallback when IPC not supported) -install(FILES ${CMAKE_CURRENT_SOURCE_DIR}/__init__.py - DESTINATION ${CMAKE_INSTALL_PREFIX}/dpctl/ipc) diff --git a/dpctl/ipc/__init__.pxd b/dpctl/ipc/__init__.pxd deleted file mode 100644 index 2231abca1e..0000000000 --- a/dpctl/ipc/__init__.pxd +++ /dev/null @@ -1,18 +0,0 @@ -# Data Parallel Control (dpctl) -# -# Copyright 2020-2026 Intel Corporation -# -# Licensed under the Apache License, Version 2.0 (the "License"); -# you may not use this file except in compliance with the License. -# You may obtain a copy of the License at -# -# http://www.apache.org/licenses/LICENSE-2.0 -# -# Unless required by applicable law or agreed to in writing, software -# distributed under the License is distributed on an "AS IS" BASIS, -# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -# See the License for the specific language governing permissions and -# limitations under the License. - -# distutils: language = c++ -# cython: language_level=3 diff --git a/dpctl/ipc/__init__.py b/dpctl/ipc/__init__.py deleted file mode 100644 index 81c21ed2b7..0000000000 --- a/dpctl/ipc/__init__.py +++ /dev/null @@ -1,82 +0,0 @@ -# Data Parallel Control (dpctl) -# -# Copyright 2020-2026 Intel Corporation -# -# Licensed under the Apache License, Version 2.0 (the "License"); -# you may not use this file except in compliance with the License. -# You may obtain a copy of the License at -# -# http://www.apache.org/licenses/LICENSE-2.0 -# -# Unless required by applicable law or agreed to in writing, software -# distributed under the License is distributed on an "AS IS" BASIS, -# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -# See the License for the specific language governing permissions and -# limitations under the License. - -""" -**Data Parallel Control IPC** provides Python objects for inter-process -communication of SYCL USM memory. - -- :class:`IPCMemoryHandle` wraps ``sycl::ext::oneapi::experimental::ipc::memory`` - to export/import USM device pointers across processes. - -Requires oneAPI DPC++ compiler >= 2026.1 (with SYCL IPC memory runtime support). -""" - -_IPC_MEMORY_AVAILABLE = False -_IPC_MEMORY_ERROR = "" - -try: - from ._ipc_memory import IPCMemoryHandle - _IPC_MEMORY_AVAILABLE = True -except ImportError as _e: - _IPC_MEMORY_ERROR = ( - "dpctl.ipc.IPCMemoryHandle is not available. " - "SYCL IPC memory support was not detected at build time. " - "This requires oneAPI DPC++ compiler >= 2026.0 with " - "sycl::ext::oneapi::experimental::ipc::memory runtime support. " - f"(Import error: {_e})" - ) - - class IPCMemoryHandle: - """Placeholder for unavailable IPC memory support.""" - - def __init__(self, *args, **kwargs): - raise RuntimeError(_IPC_MEMORY_ERROR) - - @staticmethod - def open(*args, **kwargs): - raise RuntimeError(_IPC_MEMORY_ERROR) - - @staticmethod - def close_mapping(*args, **kwargs): - raise RuntimeError(_IPC_MEMORY_ERROR) - - -def is_ipc_memory_supported(): - """Return True if IPC memory is supported by the current SYCL runtime. - - Returns: - bool: True if IPCMemoryHandle can be used, False otherwise. - """ - return _IPC_MEMORY_AVAILABLE - - -def check_ipc_memory_support(): - """Raise RuntimeError if IPC memory is not supported. - - Use this for early failure at application startup. - - Raises: - RuntimeError: if the SYCL IPC memory API is not available. - """ - if not _IPC_MEMORY_AVAILABLE: - raise RuntimeError(_IPC_MEMORY_ERROR) - - -__all__ = [ - "IPCMemoryHandle", - "is_ipc_memory_supported", - "check_ipc_memory_support", -] diff --git a/dpctl/ipc/_ipc_memory.pxd b/dpctl/ipc/_ipc_memory.pxd deleted file mode 100644 index 5d01cf3b24..0000000000 --- a/dpctl/ipc/_ipc_memory.pxd +++ /dev/null @@ -1,32 +0,0 @@ -# Data Parallel Control (dpctl) -# -# Copyright 2020-2026 Intel Corporation -# -# Licensed under the Apache License, Version 2.0 (the "License"); -# you may not use this file except in compliance with the License. -# You may obtain a copy of the License at -# -# http://www.apache.org/licenses/LICENSE-2.0 -# -# Unless required by applicable law or agreed to in writing, software -# distributed under the License is distributed on an "AS IS" BASIS, -# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -# See the License for the specific language governing permissions and -# limitations under the License. - -# distutils: language = c++ -# cython: language_level=3 - -"""Declarations for the IPCMemoryHandle Cython extension type.""" - -from .._backend cimport DPCTLSyclContextRef, DPCTLSyclDeviceRef, DPCTLSyclUSMRef -from .._sycl_context cimport SyclContext -from .._sycl_device cimport SyclDevice -from .._sycl_queue cimport SyclQueue -from ..memory._memory cimport _Memory, MemoryUSMDevice - - -cdef class IPCMemoryHandle: - cdef bytes _handle_bytes - cdef SyclContext _ctx - cdef bint _closed diff --git a/dpctl/ipc/_ipc_memory.pyx b/dpctl/ipc/_ipc_memory.pyx deleted file mode 100644 index 1c6d6ed9d3..0000000000 --- a/dpctl/ipc/_ipc_memory.pyx +++ /dev/null @@ -1,264 +0,0 @@ -# Data Parallel Control (dpctl) -# -# Copyright 2020-2026 Intel Corporation -# -# Licensed under the Apache License, Version 2.0 (the "License"); -# you may not use this file except in compliance with the License. -# You may obtain a copy of the License at -# -# http://www.apache.org/licenses/LICENSE-2.0 -# -# Unless required by applicable law or agreed to in writing, software -# distributed under the License is distributed on an "AS IS" BASIS, -# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -# See the License for the specific language governing permissions and -# limitations under the License. - -# distutils: language = c++ -# cython: language_level=3 -# cython: linetrace=True - -"""Python bindings for SYCL IPC memory -(``sycl::ext::oneapi::experimental::ipc::memory``). - -Typical usage for exporting a USM pointer:: - - import dpctl - from dpctl.memory import MemoryUSMDevice - from dpctl.ipc import IPCMemoryHandle - - # Sender process - mem = MemoryUSMDevice(1024) - handle = IPCMemoryHandle(mem) - raw = handle.to_bytes() # send *raw* to another process - - # Receiver process (same system, different PID) - remote_mem = IPCMemoryHandle.open(raw, device) -""" - -from cpython.bytes cimport PyBytes_AS_STRING, PyBytes_FromStringAndSize -from libc.stdlib cimport free as libc_free - -from dpctl._backend cimport ( - DPCTLSyclContextRef, - DPCTLSyclDeviceRef, - DPCTLSyclUSMRef, - DPCTLIPCMem_GetHandle, - DPCTLIPCMem_OpenHandle, - DPCTLIPCMem_CloseHandle, - DPCTLIPCMem_FreeHandleData, - DPCTLQueue_GetContext, - DPCTLQueue_Delete, - DPCTLContext_Delete, - DPCTLDevice_Copy, -) - -from .._sycl_context cimport SyclContext -from .._sycl_device cimport SyclDevice -from .._sycl_queue cimport SyclQueue -from ..memory._memory cimport _Memory, MemoryUSMDevice - -import dpctl - - -__all__ = ["IPCMemoryHandle"] - - -cdef class IPCMemoryHandle: - """Wrapper around a SYCL IPC memory handle. - - Instances are created by passing a :class:`dpctl.memory.MemoryUSMDevice` - (or any ``_Memory`` subclass backed by a device USM pointer) to the - constructor. The resulting object exposes :meth:`to_bytes` which - returns an opaque ``bytes`` payload suitable for inter-process - transport (e.g. via pickle, ZMQ, shared-memory). - - On the receiving side, call :meth:`IPCMemoryHandle.open` with the - payload and a target device to obtain a - :class:`dpctl.memory.MemoryUSMDevice` backed by the IPC-mapped - memory. - - Parameters - ---------- - usm_memory : dpctl.memory._Memory - USM memory object whose device pointer to export. - context : dpctl.SyclContext, optional - SYCL context to use. Defaults to the context of *usm_memory*'s - queue. - - Raises - ------ - TypeError - If *usm_memory* is not a ``_Memory`` instance. - RuntimeError - If the SYCL runtime fails to produce an IPC handle. - """ - - def __cinit__(self): - self._handle_bytes = None - self._ctx = None - self._closed = False - - def __init__(self, _Memory usm_memory not None, SyclContext context=None): - cdef DPCTLSyclUSMRef ptr = usm_memory.get_data_ptr() - if ptr is NULL: - raise ValueError("USM memory object has a null pointer") - - cdef SyclQueue q = usm_memory.queue - cdef SyclDevice dev = q.sycl_device - if not dev.has_aspect_ext_oneapi_ipc_memory: - raise RuntimeError( - "Device does not support aspect::ext_oneapi_ipc_memory" - ) - - if context is None: - context = q.sycl_context - - cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() - cdef char *data_out = NULL - cdef size_t size_out = 0 - - cdef int rc = DPCTLIPCMem_GetHandle(ptr, ctx_ref, &data_out, &size_out) - if rc != 0: - raise RuntimeError( - "DPCTLIPCMem_GetHandle failed — device may not support " - "aspect::ext_oneapi_ipc_memory" - ) - - try: - self._handle_bytes = PyBytes_FromStringAndSize(data_out, - size_out) - finally: - DPCTLIPCMem_FreeHandleData(data_out) - - self._ctx = context - self._closed = False - - def to_bytes(self): - """Return the raw IPC handle data as ``bytes``. - - The returned object can be pickled, sent over a socket, or - written to shared memory for another process to consume via - :meth:`open`. - """ - if self._closed: - raise RuntimeError("IPC handle has already been closed") - return self._handle_bytes - - @staticmethod - def open(bytes handle_bytes not None, - SyclDevice device not None, - SyclContext context=None, - Py_ssize_t nbytes=0): - """Open an IPC handle in this process. - - Parameters - ---------- - handle_bytes : bytes - Opaque payload from :meth:`to_bytes` (possibly from another - process). - device : dpctl.SyclDevice - Device to map the memory on. - context : dpctl.SyclContext, optional - SYCL context to use. Defaults to the default context for - *device*'s platform. - nbytes : int, optional - Byte size of the original allocation. If 0, the size is - determined by the driver (if supported). - - Returns - ------- - dpctl.memory.MemoryUSMDevice - A USM device memory object backed by the IPC-mapped pointer. - The mapping is closed when the returned object is garbage - collected. - - Raises - ------ - RuntimeError - If the handle cannot be opened. - """ - cdef const char *raw = PyBytes_AS_STRING(handle_bytes) - cdef size_t raw_size = len(handle_bytes) - - if not device.has_aspect_ext_oneapi_ipc_memory: - raise RuntimeError( - "Device does not support aspect::ext_oneapi_ipc_memory" - ) - - if context is None: - context = device.sycl_platform.default_context - - cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() - cdef DPCTLSyclDeviceRef dev_ref = device.get_device_ref() - - cdef DPCTLSyclUSMRef mapped_ptr = DPCTLIPCMem_OpenHandle( - raw, raw_size, ctx_ref, dev_ref) - if mapped_ptr is NULL: - raise RuntimeError("DPCTLIPCMem_OpenHandle failed") - - # Build a MemoryUSMDevice around the mapped pointer. - # Use the device-cached queue so dpctl tracks the allocation. - cdef SyclQueue q - try: - q = dpctl.SyclQueue(context, device) - except Exception: - DPCTLIPCMem_CloseHandle(mapped_ptr, ctx_ref) - raise - - # Wrap as MemoryUSMDevice — nbytes must be known by the caller - # or the driver. We require nbytes > 0 for safety. - if nbytes <= 0: - DPCTLIPCMem_CloseHandle(mapped_ptr, ctx_ref) - raise ValueError("nbytes must be > 0 for IPC open") - - cdef object mem = MemoryUSMDevice.create_from_usm_pointer_size_qref( - mapped_ptr, nbytes, q.get_queue_ref()) - return mem - - @staticmethod - def close_mapping(_Memory usm_memory not None, - SyclContext context=None): - """Explicitly close an IPC mapping. - - After calling this, *usm_memory* is invalidated and must not be - used again. Its destructor will not attempt to free the pointer. - - Parameters - ---------- - usm_memory : dpctl.memory._Memory - The memory object returned by :meth:`open`. - context : dpctl.SyclContext, optional - Context used when opening. Defaults to the memory's queue - context. - """ - cdef DPCTLSyclUSMRef ptr = usm_memory.get_data_ptr() - if ptr is NULL: - return - - if context is None: - context = usm_memory.queue.sycl_context - - cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() - DPCTLIPCMem_CloseHandle(ptr, ctx_ref) - - # Prevent the _Memory destructor from calling sycl::free on the - # now-unmapped pointer. Setting _opaque_ptr to NULL makes - # __dealloc__ skip OpaqueSmartPtr_Delete. - usm_memory._opaque_ptr = NULL - usm_memory._memory_ptr = NULL - usm_memory.nbytes = 0 - - def close(self): - """Mark this handle as closed (driver resources already released - during construction via ``put``).""" - self._closed = True - - def __dealloc__(self): - self._closed = True - - def __repr__(self): - return ( - f"IPCMemoryHandle(size={len(self._handle_bytes)}, " - f"closed={self._closed})" - ) diff --git a/dpctl/memory/_memory.pyx b/dpctl/memory/_memory.pyx index 99223039ba..41c07fd9aa 100644 --- a/dpctl/memory/_memory.pyx +++ b/dpctl/memory/_memory.pyx @@ -39,6 +39,10 @@ from dpctl._backend cimport ( # noqa: E211 DPCTLDevice_Copy, DPCTLEvent_Delete, DPCTLEvent_Wait, + DPCTLIPCMem_CloseHandle, + DPCTLIPCMem_FreeHandleData, + DPCTLIPCMem_GetHandle, + DPCTLIPCMem_OpenHandle, DPCTLmalloc_device, DPCTLmalloc_host, DPCTLmalloc_shared, @@ -744,6 +748,154 @@ cdef class _Memory: _out = mem_ty(_mem) return _out + # ─── IPC memory methods ─────────────────────────────────────────── + + def get_ipc_handle(self): + """Export this USM allocation as IPC handle bytes. + + Returns an opaque ``bytes`` payload suitable for inter-process + transport (e.g. via pickle, ZMQ, shared-memory). The receiving + process can reconstruct a mapping via + :meth:`MemoryUSMDevice.open_ipc_handle`. + + Returns + ------- + bytes + Opaque IPC handle data. + + Raises + ------ + RuntimeError + If the device does not support IPC memory. + """ + cdef DPCTLSyclUSMRef ptr = self._memory_ptr + if ptr is NULL: + raise ValueError("USM memory object has a null pointer") + + cdef SyclDevice dev = self.sycl_device + if not dev.has_aspect_ext_oneapi_ipc_memory: + raise RuntimeError( + "Device does not support IPC memory " + "(aspect::ext_oneapi_ipc_memory)" + ) + + cdef SyclContext ctx = self._context + cdef DPCTLSyclContextRef ctx_ref = ctx.get_context_ref() + cdef char *data_out = NULL + cdef size_t size_out = 0 + + cdef int rc = DPCTLIPCMem_GetHandle(ptr, ctx_ref, &data_out, &size_out) + if rc != 0: + raise RuntimeError( + "DPCTLIPCMem_GetHandle failed — IPC handle export failed" + ) + + try: + return PyBytes_FromStringAndSize(data_out, size_out) + finally: + DPCTLIPCMem_FreeHandleData(data_out) + + @staticmethod + def open_ipc_handle(bytes handle_bytes not None, + SyclDevice device not None, + Py_ssize_t nbytes, + SyclContext context=None): + """Open an IPC handle and return a MemoryUSMDevice. + + Parameters + ---------- + handle_bytes : bytes + Opaque payload from :meth:`get_ipc_handle` (possibly from + another process). + device : dpctl.SyclDevice + Device to map the memory on. + nbytes : int + Byte size of the original allocation. Must be > 0. + context : dpctl.SyclContext, optional + SYCL context to use. Defaults to the default context for + *device*'s platform. + + Returns + ------- + dpctl.memory.MemoryUSMDevice + A USM device memory object backed by the IPC-mapped pointer. + Call :meth:`close_ipc_mapping` when done. + + Raises + ------ + RuntimeError + If the device does not support IPC memory or the handle + cannot be opened. + ValueError + If *nbytes* <= 0. + """ + if not device.has_aspect_ext_oneapi_ipc_memory: + raise RuntimeError( + "Device does not support IPC memory " + "(aspect::ext_oneapi_ipc_memory)" + ) + + if nbytes <= 0: + raise ValueError("nbytes must be > 0 for IPC open") + + cdef const char *raw = PyBytes_AS_STRING(handle_bytes) + cdef size_t raw_size = len(handle_bytes) + + if context is None: + context = device.sycl_platform.default_context + + cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() + cdef DPCTLSyclDeviceRef dev_ref = device.get_device_ref() + + cdef DPCTLSyclUSMRef mapped_ptr = DPCTLIPCMem_OpenHandle( + raw, raw_size, ctx_ref, dev_ref) + if mapped_ptr is NULL: + raise RuntimeError("DPCTLIPCMem_OpenHandle failed") + + # Build a SyclQueue for the device+context + cdef SyclQueue q + try: + q = dpctl.SyclQueue(context, device) + except Exception: + DPCTLIPCMem_CloseHandle(mapped_ptr, ctx_ref) + raise + + # Wrap as MemoryUSMDevice — we pass self as memory_owner= + # so that create_from_usm_pointer_size_qref does NOT wrap in + # OpaqueSmartPtr (which would call sycl::free on dealloc). + cdef object mem = _Memory.create_from_usm_pointer_size_qref( + mapped_ptr, nbytes, q.get_queue_ref(), memory_owner=True) + return mem + + def close_ipc_mapping(self, SyclContext context=None): + """Explicitly close an IPC mapping. + + After calling this method, the object is invalidated and must + not be used again. The destructor will not attempt to free the + pointer. + + Parameters + ---------- + context : dpctl.SyclContext, optional + Context used when opening. Defaults to the memory's queue + context. + """ + cdef DPCTLSyclUSMRef ptr = self._memory_ptr + if ptr is NULL: + return + + if context is None: + context = self._context + + cdef DPCTLSyclContextRef ctx_ref = context.get_context_ref() + DPCTLIPCMem_CloseHandle(ptr, ctx_ref) + + # Prevent __dealloc__ from calling sycl::free. + self._opaque_ptr = NULL + self._memory_ptr = NULL + self.nbytes = 0 + + cdef class MemoryUSMShared(_Memory): """ diff --git a/libsyclinterface/CMakeLists.txt b/libsyclinterface/CMakeLists.txt index 8225ad7c13..c68d16b134 100644 --- a/libsyclinterface/CMakeLists.txt +++ b/libsyclinterface/CMakeLists.txt @@ -251,15 +251,20 @@ file(GLOB_RECURSE sources list(REMOVE_ITEM sources "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_vector_templ.cpp" - + "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_sycl_ipc_memory_interface.cpp" + "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_sycl_ipc_memory_stubs.cpp" ) -# Conditionally exclude IPC memory source if not supported -if(NOT DPCTL_HAS_IPC_MEMORY) - list(REMOVE_ITEM sources +# IPC memory: use real implementation or stubs depending on header availability. +if(DPCTL_HAS_IPC_MEMORY) + list(APPEND sources "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_sycl_ipc_memory_interface.cpp" ) - message(STATUS "Excluding dpctl_sycl_ipc_memory_interface.cpp from build") +else() + list(APPEND sources + "${CMAKE_CURRENT_SOURCE_DIR}/source/dpctl_sycl_ipc_memory_stubs.cpp" + ) + message(STATUS "Using IPC memory stubs (header not found)") endif() file(GLOB_RECURSE helper_sources diff --git a/libsyclinterface/helper/source/dpctl_utils_helper.cpp b/libsyclinterface/helper/source/dpctl_utils_helper.cpp index f060598b31..347d036458 100644 --- a/libsyclinterface/helper/source/dpctl_utils_helper.cpp +++ b/libsyclinterface/helper/source/dpctl_utils_helper.cpp @@ -224,9 +224,11 @@ std::string DPCTL_AspectToStr(aspect aspectTy) case aspect::ext_oneapi_is_composite: ss << "is_composite"; break; +#ifdef SYCL_EXT_ONEAPI_INTER_PROCESS_COMMUNICATION case aspect::ext_oneapi_ipc_memory: ss << "ext_oneapi_ipc_memory"; break; +#endif default: throw std::runtime_error("Unsupported aspect type"); } @@ -302,9 +304,11 @@ aspect DPCTL_StrToAspectType(const std::string &aspectTyStr) else if (aspectTyStr == "is_composite") { aspectTy = aspect::ext_oneapi_is_composite; } +#ifdef SYCL_EXT_ONEAPI_INTER_PROCESS_COMMUNICATION else if (aspectTyStr == "ext_oneapi_ipc_memory") { aspectTy = aspect::ext_oneapi_ipc_memory; } +#endif else { // \todo handle the error throw std::runtime_error("Unsupported aspect type"); @@ -357,8 +361,10 @@ aspect DPCTL_DPCTLAspectTypeToSyclAspect(DPCTLSyclAspectType AspectTy) return aspect::ext_oneapi_is_component; case DPCTLSyclAspectType::is_composite: return aspect::ext_oneapi_is_composite; +#ifdef SYCL_EXT_ONEAPI_INTER_PROCESS_COMMUNICATION case DPCTLSyclAspectType::ext_oneapi_ipc_memory: return aspect::ext_oneapi_ipc_memory; +#endif default: throw std::runtime_error("Unsupported aspect type"); } @@ -409,8 +415,10 @@ DPCTLSyclAspectType DPCTL_SyclAspectToDPCTLAspectType(aspect Aspect) return DPCTLSyclAspectType::is_composite; case aspect::ext_oneapi_is_component: return DPCTLSyclAspectType::is_component; +#ifdef SYCL_EXT_ONEAPI_INTER_PROCESS_COMMUNICATION case aspect::ext_oneapi_ipc_memory: return DPCTLSyclAspectType::ext_oneapi_ipc_memory; +#endif default: throw std::runtime_error("Unsupported aspect type"); } diff --git a/libsyclinterface/include/syclinterface/dpctl_sycl_extension_interface.h b/libsyclinterface/include/syclinterface/dpctl_sycl_extension_interface.h index c7ff7463ea..fa0b1c15df 100644 --- a/libsyclinterface/include/syclinterface/dpctl_sycl_extension_interface.h +++ b/libsyclinterface/include/syclinterface/dpctl_sycl_extension_interface.h @@ -53,6 +53,9 @@ void DPCTLWorkGroupMemory_Delete(__dpctl_take DPCTLSyclWorkGroupMemoryRef Ref); DPCTL_API bool DPCTLWorkGroupMemory_Available(); +DPCTL_API +bool DPCTLIPCMem_Available(); + typedef struct DPCTLOpaqueSyclRawKernelArg *DPCTLSyclRawKernelArgRef; DPCTL_API diff --git a/libsyclinterface/source/dpctl_sycl_extension_interface.cpp b/libsyclinterface/source/dpctl_sycl_extension_interface.cpp index fa33fd8997..20638f4d11 100644 --- a/libsyclinterface/source/dpctl_sycl_extension_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_extension_interface.cpp @@ -63,6 +63,16 @@ bool DPCTLWorkGroupMemory_Available() #endif } +DPCTL_API +bool DPCTLIPCMem_Available() +{ +#ifdef SYCL_EXT_ONEAPI_INTER_PROCESS_COMMUNICATION + return true; +#else + return false; +#endif +} + using raw_kernel_arg_t = std::vector; DPCTL_API diff --git a/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp b/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp index babf00f002..c0530a0640 100644 --- a/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_ipc_memory_interface.cpp @@ -29,10 +29,11 @@ #include "dpctl_error_handlers.h" #include "dpctl_sycl_type_casters.hpp" #include +#include #include #include -#include #include +#include using namespace sycl; @@ -85,10 +86,6 @@ int DPCTLIPCMem_GetHandle(__dpctl_keep DPCTLSyclUSMRef Ptr, // Copy handle data into a malloc'd buffer for the caller. auto HandleData = Handle.data(); // std::vector - // Release driver-side resources immediately; the bytes are sufficient - // for the receiver to call open(). - ipc::put(Handle, *Ctx); - size_t Size = HandleData.size(); char *Buf = static_cast(std::malloc(Size)); if (!Buf) { @@ -139,6 +136,8 @@ DPCTLIPCMem_OpenHandle(const char *HandleData, void *MappedPtr = ipc::open(HData, *Ctx, *Dev); return wrap(MappedPtr); } catch (std::exception const &e) { + fprintf(stderr, "[DPCTLIPCMem_OpenHandle] SYCL exception: %s\n", + e.what()); error_handler(e, __FILE__, __func__, __LINE__); return nullptr; } @@ -166,7 +165,4 @@ void DPCTLIPCMem_CloseHandle(__dpctl_keep DPCTLSyclUSMRef MappedPtr, } } -void DPCTLIPCMem_FreeHandleData(char *Data) -{ - std::free(Data); -} +void DPCTLIPCMem_FreeHandleData(char *Data) { std::free(Data); } diff --git a/libsyclinterface/source/dpctl_sycl_ipc_memory_stubs.cpp b/libsyclinterface/source/dpctl_sycl_ipc_memory_stubs.cpp new file mode 100644 index 0000000000..417fa80674 --- /dev/null +++ b/libsyclinterface/source/dpctl_sycl_ipc_memory_stubs.cpp @@ -0,0 +1,60 @@ +//===- dpctl_sycl_ipc_memory_stubs.cpp - Stub IPC functions ---------------===// +// +// Data Parallel Control (dpctl) +// +// Copyright 2020-2026 Intel Corporation +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. +// +//===----------------------------------------------------------------------===// +/// +/// \file +/// Stub implementations of DPCTLIPCMem_* functions for builds where the +/// SYCL IPC memory extension is not available. These allow _memory.pyx to +/// always link; the functions return error codes at runtime. +/// +//===----------------------------------------------------------------------===// + +#include "dpctl_error_handlers.h" +#include "dpctl_sycl_ipc_memory_interface.h" +#include + +int DPCTLIPCMem_GetHandle(__dpctl_keep DPCTLSyclUSMRef, + __dpctl_keep const DPCTLSyclContextRef, + char **, + size_t *) +{ + error_handler("IPC memory not supported in this build.", __FILE__, __func__, + __LINE__); + return 1; +} + +__dpctl_give DPCTLSyclUSMRef +DPCTLIPCMem_OpenHandle(const char *, + size_t, + __dpctl_keep const DPCTLSyclContextRef, + __dpctl_keep const DPCTLSyclDeviceRef) +{ + error_handler("IPC memory not supported in this build.", __FILE__, __func__, + __LINE__); + return nullptr; +} + +void DPCTLIPCMem_CloseHandle(__dpctl_keep DPCTLSyclUSMRef, + __dpctl_keep const DPCTLSyclContextRef) +{ + error_handler("IPC memory not supported in this build.", __FILE__, __func__, + __LINE__); +} + +void DPCTLIPCMem_FreeHandleData(char *Data) { (void)Data; } diff --git a/setup.py b/setup.py index d4adc3d94c..f7c5d3ef9c 100644 --- a/setup.py +++ b/setup.py @@ -28,7 +28,6 @@ "dpctl.memory", "dpctl.program", "dpctl.utils", - "dpctl.ipc", ], package_data={ "dpctl": [ @@ -48,7 +47,6 @@ "*.pxd", "memory/*.pxd", "program/*.pxd", - "ipc/*.pxd", ] }, include_package_data=False,