From 729cbed60f50202c0c7f13f69365af769c81d9cc Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Wed, 11 Aug 2021 18:28:33 +0300 Subject: [PATCH 01/12] Add a new SyclEventRaw class (#520) --- dpctl/__init__.py | 3 +- dpctl/_backend.pxd | 2 + dpctl/_sycl_event.pxd | 17 ++++ dpctl/_sycl_event.pyx | 140 ++++++++++++++++++++++++++++++++- dpctl/tests/test_sycl_event.py | 72 +++++++++++++++++ 5 files changed, 232 insertions(+), 2 deletions(-) create mode 100644 dpctl/tests/test_sycl_event.py 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 4b6ca0303b..4f255fd2f9 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 64f4b30fac..9aa4d5f1a6 100644 --- a/dpctl/_sycl_event.pxd +++ b/dpctl/_sycl_event.pxd @@ -33,3 +33,20 @@ cdef public api 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 29b7733913..61cc18a48f 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__) @@ -71,3 +80,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") From 6c8fcca0cdc91f290bf0540b902b2236fa0c6a9f Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Fri, 13 Aug 2021 18:01:22 +0300 Subject: [PATCH 02/12] Add a execution_status method for SyclEventRaw (#522) --- dpctl/__init__.py | 3 ++- dpctl/_backend.pxd | 7 +++++++ dpctl/_sycl_event.pyx | 20 ++++++++++++++++++++ dpctl/enum_types.py | 9 +++++++++ dpctl/tests/test_sycl_event.py | 10 ++++++++++ 5 files changed, 48 insertions(+), 1 deletion(-) diff --git a/dpctl/__init__.py b/dpctl/__init__.py index a81f620642..5ef54fa06b 100644 --- a/dpctl/__init__.py +++ b/dpctl/__init__.py @@ -65,7 +65,7 @@ ) from ._version import get_versions -from .enum_types import backend_type, device_type +from .enum_types import backend_type, device_type, event_status_type __all__ = [ "SyclContext", @@ -113,6 +113,7 @@ __all__ += [ "device_type", "backend_type", + "event_status_type", ] __all__ += [ "get_include", diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 4f255fd2f9..69f7416ab4 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -104,6 +104,12 @@ cdef extern from "dpctl_sycl_enum_types.h": _L1_cache 'L1_cache', _next_partitionable 'next_partitionable', + ctypedef enum _event_status_type 'DPCTLSyclEventStatusType': + _UNKNOWN_STATUS 'DPCTL_UNKNOWN_STATUS' + _SUBMITTED 'DPCTL_SUBMITTED' + _RUNNING 'DPCTL_RUNNING' + _COMPLETE 'DPCTL_COMPLETE' + cdef extern from "dpctl_sycl_types.h": cdef struct DPCTLOpaqueSyclContext @@ -221,6 +227,7 @@ cdef extern from "dpctl_sycl_event_interface.h": cdef DPCTLSyclEventRef DPCTLEvent_Copy(const DPCTLSyclEventRef ERef) cdef void DPCTLEvent_Wait(DPCTLSyclEventRef ERef) cdef void DPCTLEvent_Delete(DPCTLSyclEventRef ERef) + cdef _event_status_type DPCTLEvent_GetCommandExecutionStatus(DPCTLSyclEventRef ERef) cdef extern from "dpctl_sycl_kernel_interface.h": diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index 61cc18a48f..91e1cd3d85 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -29,10 +29,14 @@ from ._backend cimport ( # noqa: E211 DPCTLEvent_Copy, DPCTLEvent_Create, DPCTLEvent_Delete, + DPCTLEvent_GetCommandExecutionStatus, DPCTLEvent_Wait, DPCTLSyclEventRef, + _event_status_type, ) +from .enum_types import backend_type, event_status_type + __all__ = [ "SyclEvent", "SyclEventRaw", @@ -209,3 +213,19 @@ cdef class SyclEventRaw(_SyclEventRaw): "SyclEventRef", &_event_capsule_deleter ) + + @property + def execution_status(self): + """ Returns the event status. + """ + cdef _event_status_type ESTy = DPCTLEvent_GetCommandExecutionStatus( + self._event_ref + ) + if ESTy == _event_status_type._SUBMITTED: + return event_status_type.submitted + elif ESTy == _event_status_type._RUNNING: + return event_status_type.running + elif ESTy == _event_status_type._COMPLETE: + return event_status_type.complete + else: + raise ValueError("Unknown event status.") diff --git a/dpctl/enum_types.py b/dpctl/enum_types.py index 2c2bd4edca..de11538417 100644 --- a/dpctl/enum_types.py +++ b/dpctl/enum_types.py @@ -25,6 +25,7 @@ __all__ = [ "device_type", "backend_type", + "event_status_type", ] @@ -71,3 +72,11 @@ class backend_type(Enum): host = auto() level_zero = auto() opencl = auto() + + +class event_status_type(Enum): + + unknown_status = auto() + submitted = auto() + running = auto() + complete = auto() diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py index 0a54849615..d45e29b4e7 100644 --- a/dpctl/tests/test_sycl_event.py +++ b/dpctl/tests/test_sycl_event.py @@ -23,6 +23,7 @@ import dpctl import dpctl.memory as dpctl_mem import dpctl.program as dpctl_prog +from dpctl import event_status_type as esty from ._helper import has_cpu @@ -70,3 +71,12 @@ def test_create_event_raw_from_capsule(): dpctl.SyclEventRaw(event_capsule) except ValueError: pytest.fail("Failed to create an event from capsule") + + +def test_execution_status(): + event = dpctl.SyclEventRaw() + try: + event_status = event.execution_status + except ValueError: + pytest.fail("Failed to get an event status") + assert event_status == esty.complete From 0b9f376a006cc3f1055bf80c8900976aef49050d Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Mon, 16 Aug 2021 22:12:01 +0300 Subject: [PATCH 03/12] Add a backend property for the SyclEventRaw class (#521) --- dpctl/_backend.pxd | 1 + dpctl/_sycl_event.pyx | 18 ++++++++++++++++++ dpctl/tests/test_sycl_event.py | 7 +++++++ 3 files changed, 26 insertions(+) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 69f7416ab4..25a8183b3f 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -228,6 +228,7 @@ cdef extern from "dpctl_sycl_event_interface.h": cdef void DPCTLEvent_Wait(DPCTLSyclEventRef ERef) cdef void DPCTLEvent_Delete(DPCTLSyclEventRef ERef) cdef _event_status_type DPCTLEvent_GetCommandExecutionStatus(DPCTLSyclEventRef ERef) + cdef _backend_type DPCTLEvent_GetBackend(DPCTLSyclEventRef ERef) cdef extern from "dpctl_sycl_kernel_interface.h": diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index 91e1cd3d85..c3df6bd2cc 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -29,9 +29,11 @@ from ._backend cimport ( # noqa: E211 DPCTLEvent_Copy, DPCTLEvent_Create, DPCTLEvent_Delete, + DPCTLEvent_GetBackend, DPCTLEvent_GetCommandExecutionStatus, DPCTLEvent_Wait, DPCTLSyclEventRef, + _backend_type, _event_status_type, ) @@ -229,3 +231,19 @@ cdef class SyclEventRaw(_SyclEventRaw): return event_status_type.complete else: raise ValueError("Unknown event status.") + + @property + def backend(self): + """ Returns the Sycl backend associated with the event. + """ + cdef _backend_type BE = DPCTLEvent_GetBackend(self._event_ref) + if BE == _backend_type._OPENCL: + return backend_type.opencl + elif BE == _backend_type._LEVEL_ZERO: + return backend_type.level_zero + elif BE == _backend_type._HOST: + return backend_type.host + elif BE == _backend_type._CUDA: + return backend_type.cuda + else: + raise ValueError("Unknown backend type.") diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py index d45e29b4e7..7bdbf8fe11 100644 --- a/dpctl/tests/test_sycl_event.py +++ b/dpctl/tests/test_sycl_event.py @@ -80,3 +80,10 @@ def test_execution_status(): except ValueError: pytest.fail("Failed to get an event status") assert event_status == esty.complete + + +def test_backend(): + try: + dpctl.SyclEventRaw().backend + except ValueError: + pytest.fail("Failed to get backend from event") From c4fc6f0636c05bca02386f1a7ef8e16dc8a80d8b Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Mon, 16 Aug 2021 23:52:48 +0300 Subject: [PATCH 04/12] Add a wait_list method for SyclEventRaw class (#527) --- dpctl/_backend.pxd | 9 +++++++ dpctl/_sycl_event.pyx | 22 ++++++++++++++++ dpctl/tests/test_sycl_event.py | 47 ++++++++++++++++++++++++++++++++++ 3 files changed, 78 insertions(+) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 25a8183b3f..1966334cc1 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -229,6 +229,15 @@ cdef extern from "dpctl_sycl_event_interface.h": cdef void DPCTLEvent_Delete(DPCTLSyclEventRef ERef) cdef _event_status_type DPCTLEvent_GetCommandExecutionStatus(DPCTLSyclEventRef ERef) cdef _backend_type DPCTLEvent_GetBackend(DPCTLSyclEventRef ERef) + cdef struct DPCTLEventVector + ctypedef DPCTLEventVector *DPCTLEventVectorRef + cdef void DPCTLEventVector_Delete(DPCTLEventVectorRef EVRef) + cdef size_t DPCTLEventVector_Size(DPCTLEventVectorRef EVRef) + cdef DPCTLSyclEventRef DPCTLEventVector_GetAt( + DPCTLEventVectorRef EVRef, + size_t index) + cdef DPCTLEventVectorRef DPCTLEvent_GetWaitList( + DPCTLSyclEventRef ERef) cdef extern from "dpctl_sycl_kernel_interface.h": diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index c3df6bd2cc..145db8dd6e 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -31,7 +31,12 @@ from ._backend cimport ( # noqa: E211 DPCTLEvent_Delete, DPCTLEvent_GetBackend, DPCTLEvent_GetCommandExecutionStatus, + DPCTLEvent_GetWaitList, DPCTLEvent_Wait, + DPCTLEventVector_Delete, + DPCTLEventVector_GetAt, + DPCTLEventVector_Size, + DPCTLEventVectorRef, DPCTLSyclEventRef, _backend_type, _event_status_type, @@ -247,3 +252,20 @@ cdef class SyclEventRaw(_SyclEventRaw): return backend_type.cuda else: raise ValueError("Unknown backend type.") + + def get_wait_list(self): + cdef DPCTLEventVectorRef EVRef = DPCTLEvent_GetWaitList( + self.get_event_ref() + ) + cdef size_t num_events + cdef size_t i + cdef DPCTLSyclEventRef ERef + if (EVRef is NULL): + raise ValueError("Internal error: NULL event vector encountered") + num_events = DPCTLEventVector_Size(EVRef) + events = [] + for i in range(num_events): + ERef = DPCTLEventVector_GetAt(EVRef, i) + events.append(SyclEventRaw._create(ERef)) + DPCTLEventVector_Delete(EVRef) + return events diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py index 7bdbf8fe11..f298405743 100644 --- a/dpctl/tests/test_sycl_event.py +++ b/dpctl/tests/test_sycl_event.py @@ -87,3 +87,50 @@ def test_backend(): dpctl.SyclEventRaw().backend except ValueError: pytest.fail("Failed to get backend from event") + + +@pytest.mark.skip(reason="event::get_wait_list() method returns wrong result") +def test_get_wait_list(): + if has_cpu(): + oclSrc = " \ + kernel void add_k(global float* a) { \ + size_t index = get_global_id(0); \ + a[index] = a[index] + 1; \ + } \ + kernel void sqrt_k(global float* a) { \ + size_t index = get_global_id(0); \ + a[index] = sqrt(a[index]); \ + } \ + kernel void sin_k(global float* a) { \ + size_t index = get_global_id(0); \ + a[index] = sin(a[index]); \ + }" + q = dpctl.SyclQueue("opencl:cpu") + prog = dpctl_prog.create_program_from_source(q, oclSrc) + addKernel = prog.get_sycl_kernel("add_k") + sqrtKernel = prog.get_sycl_kernel("sqrt_k") + sinKernel = prog.get_sycl_kernel("sin_k") + + bufBytes = 1024 * np.dtype("f").itemsize + abuf = dpctl_mem.MemoryUSMShared(bufBytes, queue=q) + a = np.ndarray((1024), buffer=abuf, dtype="f") + a[:] = np.arange(1024) + args = [] + + args.append(a.base) + r = [1024] + ev_1 = q.submit(addKernel, args, r) + ev_2 = q.submit(sqrtKernel, args, r, dEvents=[ev_1]) + ev_3 = q.submit(sinKernel, args, r, dEvents=[ev_2]) + + ev_raw = dpctl.SyclEventRaw(ev_3) + + try: + wait_list = ev_raw.get_wait_list() + except ValueError: + pytest.fail( + "Failed to get a list of waiting events from SyclEventRaw" + ) + assert len(wait_list) + else: + pytest.skip("No OpenCL CPU queues available") From 3a691a72a46fe5db3a5d6a0ba514651db4e6038b Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Tue, 17 Aug 2021 18:15:30 +0300 Subject: [PATCH 05/12] Remove _init_helper from _sycl_event.pxd (#530) --- dpctl/_sycl_event.pxd | 2 -- dpctl/_sycl_event.pyx | 8 +++----- 2 files changed, 3 insertions(+), 7 deletions(-) diff --git a/dpctl/_sycl_event.pxd b/dpctl/_sycl_event.pxd index 9aa4d5f1a6..46d01d7b09 100644 --- a/dpctl/_sycl_event.pxd +++ b/dpctl/_sycl_event.pxd @@ -42,8 +42,6 @@ cdef class _SyclEventRaw: 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) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index 145db8dd6e..efd40ff416 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -100,6 +100,8 @@ cdef void _event_capsule_deleter(object o): ) DPCTLEvent_Delete(ERef) +cdef void _init_helper(_SyclEventRaw event, DPCTLSyclEventRef ERef): + event._event_ref = ERef cdef class _SyclEventRaw: """ Python wrapper class for a ``cl::sycl::event``. @@ -113,14 +115,10 @@ 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) + _init_helper(ret, eref) return SyclEventRaw(ret) cdef int _init_event_default(self): From dd3d7c04713fbc21db61e192e60843e142b25693 Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Thu, 12 Aug 2021 06:51:25 -0500 Subject: [PATCH 06/12] Add 3 methods of SyclEventRaw class for profiling --- dpctl/_backend.pxd | 5 +++- dpctl/_sycl_event.pyx | 23 ++++++++++++++ dpctl/tests/test_sycl_event.py | 55 ++++++++++++++++++++++------------ 3 files changed, 63 insertions(+), 20 deletions(-) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 1966334cc1..f485fd304f 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -21,7 +21,7 @@ types defined by dpctl's C API. """ -from libc.stdint cimport int64_t, uint32_t +from libc.stdint cimport int64_t, uint32_t, uint64_t from libcpp cimport bool @@ -238,6 +238,9 @@ cdef extern from "dpctl_sycl_event_interface.h": size_t index) cdef DPCTLEventVectorRef DPCTLEvent_GetWaitList( DPCTLSyclEventRef ERef) + cdef uint64_t DPCTLEvent_GetProfilingInfoSubmit(DPCTLSyclEventRef ERef) + cdef uint64_t DPCTLEvent_GetProfilingInfoStart(DPCTLSyclEventRef ERef) + cdef uint64_t DPCTLEvent_GetProfilingInfoEnd(DPCTLSyclEventRef ERef) cdef extern from "dpctl_sycl_kernel_interface.h": diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index efd40ff416..e6ffd04ab3 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -24,6 +24,7 @@ import logging from cpython cimport pycapsule +from libc.stdint cimport uint64_t from ._backend cimport ( # noqa: E211 DPCTLEvent_Copy, @@ -31,6 +32,9 @@ from ._backend cimport ( # noqa: E211 DPCTLEvent_Delete, DPCTLEvent_GetBackend, DPCTLEvent_GetCommandExecutionStatus, + DPCTLEvent_GetProfilingInfoEnd, + DPCTLEvent_GetProfilingInfoStart, + DPCTLEvent_GetProfilingInfoSubmit, DPCTLEvent_GetWaitList, DPCTLEvent_Wait, DPCTLEventVector_Delete, @@ -267,3 +271,22 @@ cdef class SyclEventRaw(_SyclEventRaw): events.append(SyclEventRaw._create(ERef)) DPCTLEventVector_Delete(EVRef) return events + + def profiling_info_submit(self): + cdef uint64_t profiling_info_submit = 0 + profiling_info_submit = DPCTLEvent_GetProfilingInfoSubmit( + self._event_ref + ) + return profiling_info_submit + + @property + def profiling_info_start(self): + cdef uint64_t profiling_info_start = 0 + profiling_info_start = DPCTLEvent_GetProfilingInfoStart(self._event_ref) + return profiling_info_start + + @property + def profiling_info_end(self): + cdef uint64_t profiling_info_end = 0 + profiling_info_end = DPCTLEvent_GetProfilingInfoEnd(self._event_ref) + return profiling_info_end diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py index f298405743..4e0fe30f49 100644 --- a/dpctl/tests/test_sycl_event.py +++ b/dpctl/tests/test_sycl_event.py @@ -28,6 +28,32 @@ from ._helper import has_cpu +def produce_event(profiling=False): + oclSrc = " \ + kernel void add(global int* a) { \ + size_t index = get_global_id(0); \ + a[index] = a[index] + 1; \ + }" + if profiling: + q = dpctl.SyclQueue("opencl:cpu", property="enable_profiling") + else: + 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) + + return ev + + def test_create_default_event_raw(): try: dpctl.SyclEventRaw() @@ -37,25 +63,7 @@ def test_create_default_event_raw(): 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) - + ev = produce_event() try: dpctl.SyclEventRaw(ev) except ValueError: @@ -132,5 +140,14 @@ def test_get_wait_list(): "Failed to get a list of waiting events from SyclEventRaw" ) assert len(wait_list) + + +def test_profiling_info(): + if has_cpu(): + event = produce_event(profiling=True) + event_raw = dpctl.SyclEventRaw(event) + assert event_raw.profiling_info_submit + assert event_raw.profiling_info_start + assert event_raw.profiling_info_end else: pytest.skip("No OpenCL CPU queues available") From 1e45e4cc6388830cf56d0dfd0749ec7bcdd53a63 Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Mon, 23 Aug 2021 09:30:32 -0500 Subject: [PATCH 07/12] Add the SyclTimer class to the examples --- examples/python/dppy_kernel.py | 53 +++++++++++++++++++++++++++++ examples/python/sycl_timer.py | 62 ++++++++++++++++++++++++++++++++++ 2 files changed, 115 insertions(+) create mode 100644 examples/python/dppy_kernel.py create mode 100644 examples/python/sycl_timer.py diff --git a/examples/python/dppy_kernel.py b/examples/python/dppy_kernel.py new file mode 100644 index 0000000000..f07b274a1b --- /dev/null +++ b/examples/python/dppy_kernel.py @@ -0,0 +1,53 @@ +# 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. + + +import numba_dppy as dppy +import numpy as np +from sycl_timer import SyclTimer + +import dpctl + + +@dppy.kernel +def dppy_gemm(a, b, c): + i = dppy.get_global_id(0) + j = dppy.get_global_id(1) + if i >= c.shape[0] or j >= c.shape[1]: + return + c[i, j] = 0 + for k in range(c.shape[0]): + c[i, j] += a[i, k] * b[k, j] + + +X = 1024 +Y = 16 +global_size = X, X + +griddim = X, X +blockdim = Y, Y + +a = np.arange(X * X, dtype=np.float32).reshape(X, X) +b = np.array(np.random.random(X * X), dtype=np.float32).reshape(X, X) +c = np.ones_like(a).reshape(X, X) + +q = dpctl.SyclQueue("opencl:gpu", property="enable_profiling") +with dpctl.device_context(q): + timers = SyclTimer(time_scale=1) + with timers(q): + dppy_gemm[griddim, blockdim](a, b, c) + host_time, device_time = timers.dt() + print("Wall time: ", host_time, "\n", "Device time: ", device_time) diff --git a/examples/python/sycl_timer.py b/examples/python/sycl_timer.py new file mode 100644 index 0000000000..ff0325cdfd --- /dev/null +++ b/examples/python/sycl_timer.py @@ -0,0 +1,62 @@ +# 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. + + +import timeit + +import dpctl + + +class SyclTimer: + def __init__(self, host_time=timeit.default_timer, time_scale=1): + self.timer = host_time + self.time_scale = time_scale + + def __call__(self, queue=None): + if isinstance(queue, dpctl.SyclQueue): + if queue.has_enable_profiling: + self.queue = queue + else: + raise ValueError( + "The queue does not contain the enable_profiling property" + ) + else: + raise ValueError( + "The passed queue must be " + ) + return self.__enter__() + + def __enter__(self): + self.event_start = dpctl.SyclEventRaw(self.queue.submit_barrier()) + self.host_start = self.timer() + return self + + def __exit__(self, *args): + self.event_finish = dpctl.SyclEventRaw(self.queue.submit_barrier()) + self.host_finish = self.timer() + + def dt(self): + self.event_start.wait() + self.event_finish.wait() + return ( + (self.host_finish - self.host_start) * self.time_scale, + ( + self.event_finish.profiling_info_start + - self.event_start.profiling_info_end + ) + / 1e9 + * self.time_scale, + ) From fe5f49781bbf68b7c3d617d323105715e8bced8c Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Tue, 17 Aug 2021 14:22:56 -0500 Subject: [PATCH 08/12] Add static method wait for SyclEventRaw class --- dpctl/_backend.pxd | 1 + dpctl/_sycl_event.pxd | 4 ++-- dpctl/_sycl_event.pyx | 17 +++++++++++++++-- dpctl/tests/test_sycl_event.py | 17 +++++++++++++++++ 4 files changed, 35 insertions(+), 4 deletions(-) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index f485fd304f..4af83deb4e 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -226,6 +226,7 @@ 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_WaitAndThrow(DPCTLSyclEventRef ERef) cdef void DPCTLEvent_Delete(DPCTLSyclEventRef ERef) cdef _event_status_type DPCTLEvent_GetCommandExecutionStatus(DPCTLSyclEventRef ERef) cdef _backend_type DPCTLEvent_GetBackend(DPCTLSyclEventRef ERef) diff --git a/dpctl/_sycl_event.pxd b/dpctl/_sycl_event.pxd index 46d01d7b09..7d50afdd0d 100644 --- a/dpctl/_sycl_event.pxd +++ b/dpctl/_sycl_event.pxd @@ -46,5 +46,5 @@ cdef public class SyclEventRaw(_SyclEventRaw) [object PySyclEventRawObject, type 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) + cdef DPCTLSyclEventRef get_event_ref (self) + cdef void _wait (SyclEventRaw event) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index e6ffd04ab3..4bbe189168 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -37,6 +37,7 @@ from ._backend cimport ( # noqa: E211 DPCTLEvent_GetProfilingInfoSubmit, DPCTLEvent_GetWaitList, DPCTLEvent_Wait, + DPCTLEvent_WaitAndThrow, DPCTLEventVector_Delete, DPCTLEventVector_GetAt, DPCTLEventVector_Size, @@ -199,8 +200,20 @@ cdef class SyclEventRaw(_SyclEventRaw): """ return self._event_ref - cpdef void wait(self): - DPCTLEvent_Wait(self._event_ref) + @staticmethod + cdef void _wait(SyclEventRaw event): + DPCTLEvent_WaitAndThrow(event._event_ref) + + @staticmethod + def wait(event): + if isinstance(event, list): + for e in event: + SyclEventRaw._wait(e) + elif isinstance(event, SyclEventRaw): + SyclEventRaw._wait(event) + else: + raise ValueError("The passed argument is not a list \ + or a SyclEventRaw type.") def addressof_ref(self): """ Returns the address of the C API `DPCTLSyclEventRef` pointer as diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py index 4e0fe30f49..d1bae4ac0d 100644 --- a/dpctl/tests/test_sycl_event.py +++ b/dpctl/tests/test_sycl_event.py @@ -81,6 +81,23 @@ def test_create_event_raw_from_capsule(): pytest.fail("Failed to create an event from capsule") +def test_wait_with_event(): + event = dpctl.SyclEventRaw() + try: + dpctl.SyclEventRaw.wait(event) + except ValueError: + pytest.fail("Failed to wait for the event") + + +def test_wait_with_list(): + event_1 = dpctl.SyclEventRaw() + event_2 = dpctl.SyclEventRaw() + try: + dpctl.SyclEventRaw.wait([event_1, event_2]) + except ValueError: + pytest.fail("Failed to wait for events from the list") + + def test_execution_status(): event = dpctl.SyclEventRaw() try: From 9b060b8eeee2f7875a474c8f18b74a32273cf019 Mon Sep 17 00:00:00 2001 From: Oleksandr Pavlyk Date: Tue, 24 Aug 2021 09:48:55 -0500 Subject: [PATCH 09/12] Allow dpctl.SyclEventRaw.wait to take a sequence of events --- dpctl/_sycl_event.pyx | 12 +++++++++--- 1 file changed, 9 insertions(+), 3 deletions(-) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index 4bbe189168..e8b22c3cb0 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -25,6 +25,7 @@ import logging from cpython cimport pycapsule from libc.stdint cimport uint64_t +import collections.abc from ._backend cimport ( # noqa: E211 DPCTLEvent_Copy, @@ -206,14 +207,19 @@ cdef class SyclEventRaw(_SyclEventRaw): @staticmethod def wait(event): - if isinstance(event, list): + """ Waits for a given event or a sequence of events. + """ + if (isinstance(event, collections.abc.Sequence) and + all( (isinstance(el, SyclEventRaw) for el in event) )): for e in event: SyclEventRaw._wait(e) elif isinstance(event, SyclEventRaw): SyclEventRaw._wait(event) else: - raise ValueError("The passed argument is not a list \ - or a SyclEventRaw type.") + raise TypeError( + "The passed argument is not a SyclEventRaw type or " + "a sequence of such objects" + ) def addressof_ref(self): """ Returns the address of the C API `DPCTLSyclEventRef` pointer as From 2775d29c5c643362a83af6a15991687777f25d19 Mon Sep 17 00:00:00 2001 From: Oleksandr Pavlyk Date: Tue, 24 Aug 2021 09:53:45 -0500 Subject: [PATCH 10/12] Fix flake8 flop --- dpctl/_sycl_event.pyx | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index e8b22c3cb0..e359b9db49 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -210,7 +210,7 @@ cdef class SyclEventRaw(_SyclEventRaw): """ Waits for a given event or a sequence of events. """ if (isinstance(event, collections.abc.Sequence) and - all( (isinstance(el, SyclEventRaw) for el in event) )): + all((isinstance(el, SyclEventRaw) for el in event))): for e in event: SyclEventRaw._wait(e) elif isinstance(event, SyclEventRaw): From 40b9058290499f8f5c1e571c9f32e0deefa9a41f Mon Sep 17 00:00:00 2001 From: vlad-perevezentsev Date: Wed, 25 Aug 2021 15:53:38 +0300 Subject: [PATCH 11/12] Add docstrings for the SyclEvent class (#551) --- dpctl/_sycl_event.pxd | 4 ++ dpctl/_sycl_event.pyx | 106 ++++++++++++++++++++++++++++++++++++++++-- 2 files changed, 106 insertions(+), 4 deletions(-) diff --git a/dpctl/_sycl_event.pxd b/dpctl/_sycl_event.pxd index 7d50afdd0d..1019ddd68b 100644 --- a/dpctl/_sycl_event.pxd +++ b/dpctl/_sycl_event.pxd @@ -36,10 +36,14 @@ cdef public api class SyclEvent [object PySyclEventObject, type PySyclEventType] cdef class _SyclEventRaw: + """ Data owner for SyclEvent + """ cdef DPCTLSyclEventRef _event_ref cdef public class SyclEventRaw(_SyclEventRaw) [object PySyclEventRawObject, type PySyclEventRawType]: + """ Python wrapper class for a ``cl::sycl::event`` + """ @staticmethod cdef SyclEventRaw _create (DPCTLSyclEventRef event) cdef int _init_event_default(self) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index e359b9db49..c38a8c69a1 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -107,10 +107,11 @@ cdef void _event_capsule_deleter(object o): DPCTLEvent_Delete(ERef) cdef void _init_helper(_SyclEventRaw event, DPCTLSyclEventRef ERef): + "Populate attributes of class from opaque reference ERef" event._event_ref = ERef cdef class _SyclEventRaw: - """ Python wrapper class for a ``cl::sycl::event``. + """ Data owner for SyclEvent """ def __dealloc__(self): @@ -118,11 +119,74 @@ cdef class _SyclEventRaw: cdef class SyclEventRaw(_SyclEventRaw): - """ Python wrapper class for a ``cl::sycl::event``. + """ + SyclEvent(arg=None) + Python class representing ``cl::sycl::event``. There are multiple + ways to create a :class:`dpctl.SyclEventRaw` object: + + - Invoking the constructor with no arguments creates a ready event + using the default constructor of the ``cl::sycl::event``. + + :Example: + .. code-block:: python + + import dpctl + + # Create a default SyclEventRaw + e = dpctl.SyclEventRaw() + + - Invoking the constuctor with a :class:`dpctl.SyclEvent` object + creates an event by copying the passed object. + + :Example: + .. code-block:: python + + import dpctl + + # Create a SyclEventRaw by passing SyclEvent + q = dpctl.SyclQueue() + e = q.submit_barrier() + e_r = dpctl.SyclEventRaw(e) + + - Invoking the constuctor with a :class:`dpctl.SyclEventRaw` object + creates an event by copying the passed object. + + :Example: + .. code-block:: python + + import dpctl + + # Create a SyclEventRaw by passing SyclEventRaw + e = dpctl.SyclEventRaw() + e_r = dpctl.SyclEventRaw(e) + + - Invoking the constuctor with a named ``PyCapsule`` with name + **"SyclEventRef"** that carries a pointer to a ``sycl::event`` + object. The capsule will be renamed upon successful consumption + to ensure one-time use. A new named capsule can be constructed by + using :func:`dpctl.SyclEventRaw._get_capsule` method. + + Args: + arg (optional): Defaults to ``None``. + The argument can be a :class:`dpctl.SyclEvent` + instance, a :class:`dpctl.SyclEventRaw` instance, or a + named ``PyCapsule`` called **"SyclEventRef"**. + + Raises: + ValueError: If the :class:`dpctl.SyclEventRaw` object creation failed. + TypeError: In case of incorrect arguments given to constructors, + unexpected types of input arguments, or in the case the input + capsule contained a null pointer or could not be renamed. """ @staticmethod cdef SyclEventRaw _create(DPCTLSyclEventRef eref): + """" + This function calls DPCTLEvent_Delete(eref). + + The user of this function must pass a copy to keep the + eref argument alive. + """ cdef _SyclEventRaw ret = _SyclEventRaw.__new__(_SyclEventRaw) _init_helper(ret, eref) return SyclEventRaw(ret) @@ -232,6 +296,20 @@ cdef class SyclEventRaw(_SyclEventRaw): return self._event_ref def _get_capsule(self): + """ + Returns a copy of the underlying ``cl::sycl::event`` pointer as a void + pointer inside a named ``PyCapsule`` that has the name + **SyclEventRef**. The ownership of the pointer inside the capsule is + passed to the caller, and pointer is deleted when the capsule goes out + of scope. + Returns: + :class:`pycapsule`: A capsule object storing a copy of the + ``cl::sycl::event`` pointer belonging to thus + :class:`dpctl.SyclEventRaw` instance. + Raises: + ValueError: If the ``DPCTLEvent_Copy`` fails to copy the + ``cl::sycl::event`` pointer. + """ cdef DPCTLSyclEventRef ERef = NULL ERef = DPCTLEvent_Copy(self._event_ref) if (ERef is NULL): @@ -244,7 +322,7 @@ cdef class SyclEventRaw(_SyclEventRaw): @property def execution_status(self): - """ Returns the event status. + """ Returns the event_status_type enum value for this event. """ cdef _event_status_type ESTy = DPCTLEvent_GetCommandExecutionStatus( self._event_ref @@ -260,7 +338,11 @@ cdef class SyclEventRaw(_SyclEventRaw): @property def backend(self): - """ Returns the Sycl backend associated with the event. + """Returns the backend_type enum value for the device + associated with this event. + + Returns: + backend_type: The backend for the device. """ cdef _backend_type BE = DPCTLEvent_GetBackend(self._event_ref) if BE == _backend_type._OPENCL: @@ -275,6 +357,10 @@ cdef class SyclEventRaw(_SyclEventRaw): raise ValueError("Unknown backend type.") def get_wait_list(self): + """ + Returns the list of :class:`dpctl.SyclEventRaw` objects that depend + on this event. + """ cdef DPCTLEventVectorRef EVRef = DPCTLEvent_GetWaitList( self.get_event_ref() ) @@ -292,6 +378,10 @@ cdef class SyclEventRaw(_SyclEventRaw): return events def profiling_info_submit(self): + """ + Returns the 64-bit time value in nanoseconds + when ``cl::sycl::command_group`` was submitted to the queue. + """ cdef uint64_t profiling_info_submit = 0 profiling_info_submit = DPCTLEvent_GetProfilingInfoSubmit( self._event_ref @@ -300,12 +390,20 @@ cdef class SyclEventRaw(_SyclEventRaw): @property def profiling_info_start(self): + """ + Returns the 64-bit time value in nanoseconds + when ``cl::sycl::command_group`` started execution on the device. + """ cdef uint64_t profiling_info_start = 0 profiling_info_start = DPCTLEvent_GetProfilingInfoStart(self._event_ref) return profiling_info_start @property def profiling_info_end(self): + """ + Returns the 64-bit time value in nanoseconds + when ``cl::sycl::command_group`` finished execution on the device. + """ cdef uint64_t profiling_info_end = 0 profiling_info_end = DPCTLEvent_GetProfilingInfoEnd(self._event_ref) return profiling_info_end From b9f4fe7dd62d572e2cd092d2164ae264cfe2d74f Mon Sep 17 00:00:00 2001 From: Oleksandr Pavlyk Date: Wed, 25 Aug 2021 15:27:06 -0500 Subject: [PATCH 12/12] SyclEventRaw -> SyclEvent Backward compatibility is still preserved. --- dpctl/__init__.py | 2 +- dpctl/_sycl_event.pxd | 31 ++++---- dpctl/_sycl_event.pyx | 129 ++++++++++----------------------- dpctl/tests/test_sycl_event.py | 57 ++++++--------- examples/python/dppy_kernel.py | 10 ++- examples/python/sycl_timer.py | 7 +- 6 files changed, 83 insertions(+), 153 deletions(-) diff --git a/dpctl/__init__.py b/dpctl/__init__.py index 5ef54fa06b..c033a75de2 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, SyclEventRaw +from dpctl._sycl_event import SyclEvent from dpctl._sycl_platform import SyclPlatform, get_platforms, lsplatform from dpctl._sycl_queue import ( SyclKernelInvalidRangeError, diff --git a/dpctl/_sycl_event.pxd b/dpctl/_sycl_event.pxd index 1019ddd68b..72d8e4eb3d 100644 --- a/dpctl/_sycl_event.pxd +++ b/dpctl/_sycl_event.pxd @@ -23,32 +23,27 @@ from ._backend cimport DPCTLSyclEventRef -cdef public api class SyclEvent [object PySyclEventObject, type PySyclEventType]: - ''' Wrapper class for a Sycl Event - ''' - cdef DPCTLSyclEventRef _event_ref - cdef list _args - - @staticmethod - cdef SyclEvent _create (DPCTLSyclEventRef e, list args) - cdef DPCTLSyclEventRef get_event_ref (self) - cpdef void wait (self) - - -cdef class _SyclEventRaw: +cdef public api class _SyclEvent [ + object Py_SyclEventObject, + type Py_SyclEventType +]: """ Data owner for SyclEvent """ cdef DPCTLSyclEventRef _event_ref + cdef object args -cdef public class SyclEventRaw(_SyclEventRaw) [object PySyclEventRawObject, type PySyclEventRawType]: +cdef public api class SyclEvent(_SyclEvent) [ + object PySyclEventObject, + type PySyclEventType +]: """ Python wrapper class for a ``cl::sycl::event`` """ @staticmethod - cdef SyclEventRaw _create (DPCTLSyclEventRef event) + cdef SyclEvent _create (DPCTLSyclEventRef event, object args=*) 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__SyclEvent(self, _SyclEvent other) cdef int _init_event_from_capsule(self, object caps) cdef DPCTLSyclEventRef get_event_ref (self) - cdef void _wait (SyclEventRaw event) + cdef void _wait (SyclEvent event) + cpdef void wait (self) diff --git a/dpctl/_sycl_event.pyx b/dpctl/_sycl_event.pyx index c38a8c69a1..3e81a188ee 100644 --- a/dpctl/_sycl_event.pyx +++ b/dpctl/_sycl_event.pyx @@ -52,7 +52,6 @@ from .enum_types import backend_type, event_status_type __all__ = [ "SyclEvent", - "SyclEventRaw", ] _logger = logging.getLogger(__name__) @@ -65,39 +64,6 @@ cdef api DPCTLSyclEventRef get_event_ref(SyclEvent ev): return ev.get_event_ref() -cdef class SyclEvent: - """ Python wrapper class for cl::sycl::event. - """ - - @staticmethod - cdef SyclEvent _create(DPCTLSyclEventRef eref, list args): - cdef SyclEvent ret = SyclEvent.__new__(SyclEvent) - ret._event_ref = eref - ret._args = args - return ret - - def __dealloc__(self): - self.wait() - DPCTLEvent_Delete(self._event_ref) - - 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 int(self._event_ref) - cdef void _event_capsule_deleter(object o): cdef DPCTLSyclEventRef ERef = NULL if pycapsule.PyCapsule_IsValid(o, "SyclEventRef"): @@ -106,23 +72,27 @@ cdef void _event_capsule_deleter(object o): ) DPCTLEvent_Delete(ERef) -cdef void _init_helper(_SyclEventRaw event, DPCTLSyclEventRef ERef): + +cdef void _init_helper(_SyclEvent event, DPCTLSyclEventRef ERef): "Populate attributes of class from opaque reference ERef" event._event_ref = ERef -cdef class _SyclEventRaw: + +cdef class _SyclEvent: """ Data owner for SyclEvent """ def __dealloc__(self): + DPCTLEvent_Wait(self._event_ref) DPCTLEvent_Delete(self._event_ref) + self.args = None -cdef class SyclEventRaw(_SyclEventRaw): +cdef class SyclEvent(_SyclEvent): """ SyclEvent(arg=None) Python class representing ``cl::sycl::event``. There are multiple - ways to create a :class:`dpctl.SyclEventRaw` object: + ways to create a :class:`dpctl.SyclEvent` object: - Invoking the constructor with no arguments creates a ready event using the default constructor of the ``cl::sycl::event``. @@ -132,81 +102,53 @@ cdef class SyclEventRaw(_SyclEventRaw): import dpctl - # Create a default SyclEventRaw - e = dpctl.SyclEventRaw() - - - Invoking the constuctor with a :class:`dpctl.SyclEvent` object - creates an event by copying the passed object. - - :Example: - .. code-block:: python - - import dpctl - - # Create a SyclEventRaw by passing SyclEvent - q = dpctl.SyclQueue() - e = q.submit_barrier() - e_r = dpctl.SyclEventRaw(e) - - - Invoking the constuctor with a :class:`dpctl.SyclEventRaw` object - creates an event by copying the passed object. - - :Example: - .. code-block:: python - - import dpctl - - # Create a SyclEventRaw by passing SyclEventRaw - e = dpctl.SyclEventRaw() - e_r = dpctl.SyclEventRaw(e) + # Create a default SyclEvent + e = dpctl.SyclEvent() - Invoking the constuctor with a named ``PyCapsule`` with name **"SyclEventRef"** that carries a pointer to a ``sycl::event`` object. The capsule will be renamed upon successful consumption to ensure one-time use. A new named capsule can be constructed by - using :func:`dpctl.SyclEventRaw._get_capsule` method. + using :func:`dpctl.SyclEvent._get_capsule` method. Args: arg (optional): Defaults to ``None``. The argument can be a :class:`dpctl.SyclEvent` - instance, a :class:`dpctl.SyclEventRaw` instance, or a + instance, a :class:`dpctl.SyclEvent` instance, or a named ``PyCapsule`` called **"SyclEventRef"**. Raises: - ValueError: If the :class:`dpctl.SyclEventRaw` object creation failed. + ValueError: If the :class:`dpctl.SyclEvent` object creation failed. TypeError: In case of incorrect arguments given to constructors, unexpected types of input arguments, or in the case the input capsule contained a null pointer or could not be renamed. """ @staticmethod - cdef SyclEventRaw _create(DPCTLSyclEventRef eref): + cdef SyclEvent _create(DPCTLSyclEventRef eref, object args=None): """" This function calls DPCTLEvent_Delete(eref). The user of this function must pass a copy to keep the eref argument alive. """ - cdef _SyclEventRaw ret = _SyclEventRaw.__new__(_SyclEventRaw) + cdef _SyclEvent ret = _SyclEvent.__new__(_SyclEvent) _init_helper(ret, eref) - return SyclEventRaw(ret) + ret.args=args + return SyclEvent(ret) cdef int _init_event_default(self): self._event_ref = DPCTLEvent_Create() if (self._event_ref is NULL): return -1 + self.args=None return 0 - cdef int _init_event_from__SyclEventRaw(self, _SyclEventRaw other): + cdef int _init_event_from__SyclEvent(self, _SyclEvent 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 + self.args = other.args return 0 cdef int _init_event_from_capsule(self, object cap): @@ -226,6 +168,7 @@ cdef class SyclEventRaw(_SyclEventRaw): if (ERef_copy is NULL): return -3 self._event_ref = ERef_copy + self.args = None return 0 else: return -128 @@ -234,10 +177,8 @@ cdef class SyclEventRaw(_SyclEventRaw): 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 type(arg) is _SyclEvent: + ret = self._init_event_from__SyclEvent(<_SyclEvent> arg) elif pycapsule.PyCapsule_IsValid(arg, "SyclEventRef"): ret = self._init_event_from_capsule(arg) else: @@ -266,22 +207,22 @@ cdef class SyclEventRaw(_SyclEventRaw): return self._event_ref @staticmethod - cdef void _wait(SyclEventRaw event): + cdef void _wait(SyclEvent event): DPCTLEvent_WaitAndThrow(event._event_ref) @staticmethod - def wait(event): + def wait_for(event): """ Waits for a given event or a sequence of events. """ if (isinstance(event, collections.abc.Sequence) and - all((isinstance(el, SyclEventRaw) for el in event))): + all((isinstance(el, SyclEvent) for el in event))): for e in event: - SyclEventRaw._wait(e) - elif isinstance(event, SyclEventRaw): - SyclEventRaw._wait(event) + SyclEvent._wait(e) + elif isinstance(event, SyclEvent): + SyclEvent._wait(event) else: raise TypeError( - "The passed argument is not a SyclEventRaw type or " + "The passed argument is not a SyclEvent type or " "a sequence of such objects" ) @@ -305,7 +246,7 @@ cdef class SyclEventRaw(_SyclEventRaw): Returns: :class:`pycapsule`: A capsule object storing a copy of the ``cl::sycl::event`` pointer belonging to thus - :class:`dpctl.SyclEventRaw` instance. + :class:`dpctl.SyclEvent` instance. Raises: ValueError: If the ``DPCTLEvent_Copy`` fails to copy the ``cl::sycl::event`` pointer. @@ -358,7 +299,7 @@ cdef class SyclEventRaw(_SyclEventRaw): def get_wait_list(self): """ - Returns the list of :class:`dpctl.SyclEventRaw` objects that depend + Returns the list of :class:`dpctl.SyclEvent` objects that depend on this event. """ cdef DPCTLEventVectorRef EVRef = DPCTLEvent_GetWaitList( @@ -373,7 +314,7 @@ cdef class SyclEventRaw(_SyclEventRaw): events = [] for i in range(num_events): ERef = DPCTLEventVector_GetAt(EVRef, i) - events.append(SyclEventRaw._create(ERef)) + events.append(SyclEvent._create(ERef, args=None)) DPCTLEventVector_Delete(EVRef) return events @@ -407,3 +348,7 @@ cdef class SyclEventRaw(_SyclEventRaw): cdef uint64_t profiling_info_end = 0 profiling_info_end = DPCTLEvent_GetProfilingInfoEnd(self._event_ref) return profiling_info_end + + cpdef void wait(self): + "Synchronously wait for completion of this event." + DPCTLEvent_Wait(self._event_ref) diff --git a/dpctl/tests/test_sycl_event.py b/dpctl/tests/test_sycl_event.py index d1bae4ac0d..289060ec1b 100644 --- a/dpctl/tests/test_sycl_event.py +++ b/dpctl/tests/test_sycl_event.py @@ -14,7 +14,7 @@ # See the License for the specific language governing permissions and # limitations under the License. -""" Defines unit test cases for the SyclEventRaw class. +""" Defines unit test cases for the SyclEvent class. """ import numpy as np @@ -54,52 +54,46 @@ def produce_event(profiling=False): return ev -def test_create_default_event_raw(): +def test_create_default_event(): try: - dpctl.SyclEventRaw() + dpctl.SyclEvent() except ValueError: pytest.fail("Failed to create a default event") -def test_create_event_raw_from_SyclEvent(): - if has_cpu(): - ev = produce_event() - 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(): +def test_create_event_from_capsule(): try: - event = dpctl.SyclEventRaw() + event = dpctl.SyclEvent() event_capsule = event._get_capsule() - dpctl.SyclEventRaw(event_capsule) + dpctl.SyclEvent(event_capsule) except ValueError: pytest.fail("Failed to create an event from capsule") def test_wait_with_event(): - event = dpctl.SyclEventRaw() + event = dpctl.SyclEvent() try: - dpctl.SyclEventRaw.wait(event) + dpctl.SyclEvent.wait_for(event) + except ValueError: + pytest.fail("Failed to wait_for(event)") + event = dpctl.SyclEvent() + try: + event.wait() except ValueError: pytest.fail("Failed to wait for the event") def test_wait_with_list(): - event_1 = dpctl.SyclEventRaw() - event_2 = dpctl.SyclEventRaw() + event_1 = dpctl.SyclEvent() + event_2 = dpctl.SyclEvent() try: - dpctl.SyclEventRaw.wait([event_1, event_2]) + dpctl.SyclEvent.wait_for([event_1, event_2]) except ValueError: pytest.fail("Failed to wait for events from the list") def test_execution_status(): - event = dpctl.SyclEventRaw() + event = dpctl.SyclEvent() try: event_status = event.execution_status except ValueError: @@ -109,7 +103,7 @@ def test_execution_status(): def test_backend(): try: - dpctl.SyclEventRaw().backend + dpctl.SyclEvent().backend except ValueError: pytest.fail("Failed to get backend from event") @@ -148,23 +142,18 @@ def test_get_wait_list(): ev_2 = q.submit(sqrtKernel, args, r, dEvents=[ev_1]) ev_3 = q.submit(sinKernel, args, r, dEvents=[ev_2]) - ev_raw = dpctl.SyclEventRaw(ev_3) - try: - wait_list = ev_raw.get_wait_list() + wait_list = ev_3.get_wait_list() except ValueError: - pytest.fail( - "Failed to get a list of waiting events from SyclEventRaw" - ) + pytest.fail("Failed to get a list of waiting events from SyclEvent") assert len(wait_list) def test_profiling_info(): if has_cpu(): event = produce_event(profiling=True) - event_raw = dpctl.SyclEventRaw(event) - assert event_raw.profiling_info_submit - assert event_raw.profiling_info_start - assert event_raw.profiling_info_end + assert event.profiling_info_submit + assert event.profiling_info_start + assert event.profiling_info_end else: pytest.skip("No OpenCL CPU queues available") diff --git a/examples/python/dppy_kernel.py b/examples/python/dppy_kernel.py index f07b274a1b..3384fd5183 100644 --- a/examples/python/dppy_kernel.py +++ b/examples/python/dppy_kernel.py @@ -15,17 +15,17 @@ # limitations under the License. -import numba_dppy as dppy +import numba_dppy import numpy as np from sycl_timer import SyclTimer import dpctl -@dppy.kernel +@numba_dppy.kernel def dppy_gemm(a, b, c): - i = dppy.get_global_id(0) - j = dppy.get_global_id(1) + i = numba_dppy.get_global_id(0) + j = numba_dppy.get_global_id(1) if i >= c.shape[0] or j >= c.shape[1]: return c[i, j] = 0 @@ -49,5 +49,7 @@ def dppy_gemm(a, b, c): timers = SyclTimer(time_scale=1) with timers(q): dppy_gemm[griddim, blockdim](a, b, c) + cc = np.dot(a, b) host_time, device_time = timers.dt() print("Wall time: ", host_time, "\n", "Device time: ", device_time) + print(np.allclose(c, cc)) diff --git a/examples/python/sycl_timer.py b/examples/python/sycl_timer.py index ff0325cdfd..60422b9ebc 100644 --- a/examples/python/sycl_timer.py +++ b/examples/python/sycl_timer.py @@ -40,12 +40,12 @@ def __call__(self, queue=None): return self.__enter__() def __enter__(self): - self.event_start = dpctl.SyclEventRaw(self.queue.submit_barrier()) + self.event_start = self.queue.submit_barrier() self.host_start = self.timer() return self def __exit__(self, *args): - self.event_finish = dpctl.SyclEventRaw(self.queue.submit_barrier()) + self.event_finish = self.queue.submit_barrier() self.host_finish = self.timer() def dt(self): @@ -57,6 +57,5 @@ def dt(self): self.event_finish.profiling_info_start - self.event_start.profiling_info_end ) - / 1e9 - * self.time_scale, + * (1e-9 * self.time_scale), )