diff --git a/dpctl/__init__.py b/dpctl/__init__.py index 409b93ef79..a81f620642 100644 --- a/dpctl/__init__.py +++ b/dpctl/__init__.py @@ -46,7 +46,7 @@ select_gpu_device, select_host_device, ) -from dpctl._sycl_event import SyclEvent +from dpctl._sycl_event import SyclEvent, SyclEventRaw from dpctl._sycl_platform import SyclPlatform, get_platforms, lsplatform from dpctl._sycl_queue import ( SyclKernelInvalidRangeError, @@ -88,6 +88,7 @@ ] __all__ += [ "SyclEvent", + "SyclEventRaw", ] __all__ += [ "get_platforms", diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 14917afe78..d8022fe5c6 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -217,6 +217,8 @@ cdef extern from "dpctl_sycl_device_selector_interface.h": cdef extern from "dpctl_sycl_event_interface.h": + cdef DPCTLSyclEventRef DPCTLEvent_Create() + cdef DPCTLSyclEventRef DPCTLEvent_Copy(const DPCTLSyclEventRef ERef) cdef void DPCTLEvent_Wait(DPCTLSyclEventRef ERef) cdef void DPCTLEvent_Delete(DPCTLSyclEventRef ERef) diff --git a/dpctl/_sycl_event.pxd b/dpctl/_sycl_event.pxd index 7f397cb716..7049eda955 100644 --- a/dpctl/_sycl_event.pxd +++ b/dpctl/_sycl_event.pxd @@ -33,3 +33,20 @@ cdef public class SyclEvent [object PySyclEventObject, type PySyclEventType]: cdef SyclEvent _create (DPCTLSyclEventRef e, list args) cdef DPCTLSyclEventRef get_event_ref (self) cpdef void wait (self) + + +cdef class _SyclEventRaw: + cdef DPCTLSyclEventRef _event_ref + + +cdef public class SyclEventRaw(_SyclEventRaw) [object PySyclEventRawObject, type PySyclEventRawType]: + @staticmethod + cdef SyclEventRaw _create (DPCTLSyclEventRef event) + @staticmethod + cdef void _init_helper(_SyclEventRaw event, DPCTLSyclEventRef ERef) + cdef int _init_event_default(self) + cdef int _init_event_from__SyclEventRaw(self, _SyclEventRaw other) + cdef int _init_event_from_SyclEvent(self, SyclEvent event) + cdef int _init_event_from_capsule(self, object caps) + cdef DPCTLSyclEventRef get_event_ref (self) + cpdef void wait (self) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index eac541fbb1..83cd19368b 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -23,10 +23,19 @@ import logging -from ._backend cimport DPCTLEvent_Delete, DPCTLEvent_Wait, DPCTLSyclEventRef +from cpython cimport pycapsule + +from ._backend cimport ( # noqa: E211 + DPCTLEvent_Copy, + DPCTLEvent_Create, + DPCTLEvent_Delete, + DPCTLEvent_Wait, + DPCTLSyclEventRef, +) __all__ = [ "SyclEvent", + "SyclEventRaw", ] _logger = logging.getLogger(__name__) @@ -64,3 +73,132 @@ cdef class SyclEvent: SyclEvent cast to a size_t. """ return int(self._event_ref) + +cdef void _event_capsule_deleter(object o): + cdef DPCTLSyclEventRef ERef = NULL + if pycapsule.PyCapsule_IsValid(o, "SyclEventRef"): + ERef = pycapsule.PyCapsule_GetPointer( + o, "SyclEventRef" + ) + DPCTLEvent_Delete(ERef) + + +cdef class _SyclEventRaw: + """ Python wrapper class for a ``cl::sycl::event``. + """ + + def __dealloc__(self): + DPCTLEvent_Delete(self._event_ref) + + +cdef class SyclEventRaw(_SyclEventRaw): + """ Python wrapper class for a ``cl::sycl::event``. + """ + + @staticmethod + cdef void _init_helper(_SyclEventRaw event, DPCTLSyclEventRef ERef): + event._event_ref = ERef + + @staticmethod + cdef SyclEventRaw _create(DPCTLSyclEventRef eref): + cdef _SyclEventRaw ret = _SyclEventRaw.__new__(_SyclEventRaw) + SyclEventRaw._init_helper(ret, eref) + return SyclEventRaw(ret) + + cdef int _init_event_default(self): + self._event_ref = DPCTLEvent_Create() + if (self._event_ref is NULL): + return -1 + return 0 + + cdef int _init_event_from__SyclEventRaw(self, _SyclEventRaw other): + self._event_ref = DPCTLEvent_Copy(other._event_ref) + if (self._event_ref is NULL): + return -1 + return 0 + + cdef int _init_event_from_SyclEvent(self, SyclEvent event): + self._event_ref = DPCTLEvent_Copy(event._event_ref) + if (self._event_ref is NULL): + return -1 + return 0 + + cdef int _init_event_from_capsule(self, object cap): + cdef DPCTLSyclEventRef ERef = NULL + cdef DPCTLSyclEventRef ERef_copy = NULL + cdef int ret = 0 + if pycapsule.PyCapsule_IsValid(cap, "SyclEventRef"): + ERef = pycapsule.PyCapsule_GetPointer( + cap, "SyclEventRef" + ) + if (ERef is NULL): + return -2 + ret = pycapsule.PyCapsule_SetName(cap, "used_SyclEventRef") + if (ret): + return -2 + ERef_copy = DPCTLEvent_Copy(ERef) + if (ERef_copy is NULL): + return -3 + self._event_ref = ERef_copy + return 0 + else: + return -128 + + def __cinit__(self, arg=None): + cdef int ret = 0 + if arg is None: + ret = self._init_event_default() + elif type(arg) is _SyclEventRaw: + ret = self._init_event_from__SyclEventRaw(<_SyclEventRaw> arg) + elif isinstance(arg, SyclEvent): + ret = self._init_event_from_SyclEvent( arg) + elif pycapsule.PyCapsule_IsValid(arg, "SyclEventRef"): + ret = self._init_event_from_capsule(arg) + else: + raise TypeError( + "Invalid argument." + ) + if (ret < 0): + if (ret == -1): + raise ValueError("Event failed to be created.") + elif (ret == -2): + raise TypeError( + "Input capsule {} contains a null pointer or could not be" + " renamed".format(arg) + ) + elif (ret == -3): + raise ValueError( + "Internal Error: Could not create a copy of a sycl event." + ) + raise ValueError( + "Unrecognized error code ({}) encountered.".format(ret) + ) + + cdef DPCTLSyclEventRef get_event_ref(self): + """ Returns the `DPCTLSyclEventRef` pointer for this class. + """ + return self._event_ref + + cpdef void wait(self): + DPCTLEvent_Wait(self._event_ref) + + def addressof_ref(self): + """ Returns the address of the C API `DPCTLSyclEventRef` pointer as + a size_t. + + Returns: + The address of the `DPCTLSyclEventRef` object used to create this + `SyclEvent` cast to a size_t. + """ + return self._event_ref + + def _get_capsule(self): + cdef DPCTLSyclEventRef ERef = NULL + ERef = DPCTLEvent_Copy(self._event_ref) + if (ERef is NULL): + raise ValueError("SyclEvent copy failed.") + return pycapsule.PyCapsule_New( + ERef, + "SyclEventRef", + &_event_capsule_deleter + ) diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py new file mode 100644 index 0000000000..0a54849615 --- /dev/null +++ b/dpctl/tests/test_sycl_event.py @@ -0,0 +1,72 @@ +# Data Parallel Control (dpctl) +# +# Copyright 2020-2021 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. + +""" Defines unit test cases for the SyclEventRaw class. +""" + +import numpy as np +import pytest + +import dpctl +import dpctl.memory as dpctl_mem +import dpctl.program as dpctl_prog + +from ._helper import has_cpu + + +def test_create_default_event_raw(): + try: + dpctl.SyclEventRaw() + except ValueError: + pytest.fail("Failed to create a default event") + + +def test_create_event_raw_from_SyclEvent(): + if has_cpu(): + oclSrc = " \ + kernel void add(global int* a) { \ + size_t index = get_global_id(0); \ + a[index] = a[index] + 1; \ + }" + q = dpctl.SyclQueue("opencl:cpu") + prog = dpctl_prog.create_program_from_source(q, oclSrc) + addKernel = prog.get_sycl_kernel("add") + + bufBytes = 1024 * np.dtype("i").itemsize + abuf = dpctl_mem.MemoryUSMShared(bufBytes, queue=q) + a = np.ndarray((1024), buffer=abuf, dtype="i") + a[:] = np.arange(1024) + args = [] + + args.append(a.base) + r = [1024] + ev = q.submit(addKernel, args, r) + + try: + dpctl.SyclEventRaw(ev) + except ValueError: + pytest.fail("Failed to create an event from SyclEvent") + else: + pytest.skip("No OpenCL CPU queues available") + + +def test_create_event_raw_from_capsule(): + try: + event = dpctl.SyclEventRaw() + event_capsule = event._get_capsule() + dpctl.SyclEventRaw(event_capsule) + except ValueError: + pytest.fail("Failed to create an event from capsule")