Implementing USM memory management

dppl/_memory.pyx

@@ -0,0 +1,154 @@
+import dppl
+
+from cython.operator cimport dereference as deref
+
+from cpython.pycapsule cimport PyCapsule_GetPointer
+from cpython cimport Py_buffer
+
+cdef extern from "CL/sycl.hpp" namespace "cl::sycl::usm":
+    cdef enum alloc:
+       host 'cl::sycl::usm::alloc::host'
+       device 'cl::sycl::usm::alloc::device'
+       shared 'cl::sycl::usm::alloc::shared'
+       unknown 'cl::sycl::usm::alloc::unknown'
+
+cdef extern from "CL/sycl.hpp" namespace "cl::sycl":
+    cdef cppclass context nogil:
+       pass
+
+    cdef cppclass queue nogil:
+       context get_context() nogil
+       pass
+
+    cdef void* malloc_shared(Py_ssize_t, queue&) nogil
+    cdef void free(void *, queue&) nogil
+    cdef alloc get_pointer_type(void *, context&) nogil
+
+
+cdef class SyclQueue:
+    cdef object queue_cap
+    cdef queue q
+
+    def __cinit__(self):
+        cdef void* q_ptr
+        self.queue_cap = dppl.get_current_queue()
+        q_ptr = PyCapsule_GetPointer(self.queue_cap, NULL)
+        if (q_ptr):
+            self.q = deref(<queue *>q_ptr)
+        else:
+            raise ValueError("NULL pointer returned by the Capsule")
+
+    def get_pointer_type(self, Py_ssize_t p):
+        cdef context ctx = self.q.get_context()
+        cdef void * p_ptr = <void *> p
+
+        ptr_type = get_pointer_type(p_ptr, ctx)
+        if (ptr_type == alloc.shared):
+            return "shared"
+        elif (ptr_type == alloc.host):
+            return "host"
+        elif (ptr_type == alloc.device):
+            return "device"
+        else:
+            return "unknown"
+
+    property get_capsule:
+        def __get__(self):
+            return self.queue_cap
+
+    cdef queue get_queue(self):
+        return self.q
+
+
+cdef class Memory:
+    cdef void* _ptr
+    cdef Py_ssize_t nbytes
+    cdef object queue_cap
+
+    def __cinit__(self, Py_ssize_t nbytes):
+        cdef object q_cap
+        cdef void* queue_ptr
+        cdef void* p
+
+        self._ptr = NULL
+        self.queue_cap = None
+        self.nbytes = 0
+
+        if (nbytes > 0):
+            q_cap = dppl.get_current_queue()
+            queue_ptr = PyCapsule_GetPointer(q_cap, NULL)
+            p = malloc_shared(nbytes, deref(<queue *>queue_ptr))
+            if (p):
+                self._ptr = p
+                self.nbytes = nbytes
+                self.queue_cap = q_cap
+            else:
+                raise RuntimeError("Null memory pointer returned")
+        else:
+            raise ValueError("Non-positive number of bytes found.")
+
+    def __dealloc__(self):
+        cdef void* queue_ptr
+
+        if (self._ptr):
+            queue_ptr = PyCapsule_GetPointer(self.queue_cap, NULL)
+            free(self._ptr, deref(<queue *>queue_ptr))
+        self._ptr = NULL
+        self.nbytes = 0
+        self.queue_cap = None
+
+    def __getbuffer__(self, Py_buffer *buffer, int flags):
+        buffer.buf = <char *>self._ptr
+        buffer.format = 'B'                     # byte
+        buffer.internal = NULL                  # see References
+        buffer.itemsize = 1
+        buffer.len = self.nbytes
+        buffer.ndim = 1
+        buffer.obj = self
+        buffer.readonly = 0
+        buffer.shape = &self.nbytes
+        buffer.strides = &buffer.itemsize
+        buffer.suboffsets = NULL                # for pointer arrays only
+
+    property pointer:
+        def __get__(self):
+            return <object>(<Py_ssize_t>self._ptr)
+
+    property nbytes:
+        def __get__(self):
+            return self.nbytes
+
+    property _queue:
+        def __get__(self):
+            return self.queue_cap
+
+    def __repr__(self):
+        return "<Intel(R) USM allocated memory block of {} bytes at {}>".format(self.nbytes, hex(<object>(<Py_ssize_t>self._ptr)))
+
+    def _usm_type(self, qcaps=None):
+        cdef void *q_ptr
+        cdef alloc ptr_type
+
+        _cap = qcaps if (qcaps) else self.queue_cap
+        q_ptr = PyCapsule_GetPointer(_cap, NULL)
+        ptr_type = get_pointer_type(self._ptr, deref(<queue*>q_ptr).get_context())
+        if (ptr_type == alloc.shared):
+            return "shared"
+        elif (ptr_type == alloc.host):
+            return "host"
+        elif (ptr_type == alloc.device):
+            return "device"
+        else:
+            return "unknown"
+
+#    cdef void* _ptr
+#    cdef Py_ssize_t nbytes
+#    cdef object queue_cap
+
+    @staticmethod
+    cdef Memory create(void *p, Py_ssize_t nbytes, object queue_cap):
+        cdef Memory ret = Memory.__new__()
+        ret._ptr = p
+        ret.nbytes = nbytes
+        ret.q_cap = queue_cap
+        return ret

dppl/tests/__init__.py

@@ -1 +1,2 @@
 from .test_dump_functions import *
+from .dppl_tests import *

dppl/tests/dppl_tests/__init__.py

@@ -1 +1,2 @@
-from .test_sycl_queue_manager import *
+from .test_sycl_queue_manager import *
+from .test_sycl_memory_manager import *

dppl/tests/dppl_tests/test_sycl_memory_manager.py

@@ -0,0 +1,59 @@
+##===---------- test_sycl_queue_manager.py - dppl  -------*- Python -*-----===##
+##
+##               Python Data Parallel Processing Library (PyDPPL)
+##
+## Copyright 2020 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 unittest
+import dppl
+import dppl._memory as mem
+
+
+class TestMemory (unittest.TestCase):
+    # @unittest.skipIf(not dppl.has_sycl_platforms, "No SYCL platforms available")
+    def test_memory_create (self):
+        nbytes = 1024
+        mobj = mem.Memory(nbytes)
+        self.assertEqual(mobj.nbytes, nbytes)
+
+    def _create_memory (self):
+        nbytes = 1024
+        mobj = mem.Memory(nbytes)
+        return mobj
+
+    def test_memory_without_context (self):
+        mobj = self._create_memory()
+
+        # Without context
+        self.assertEqual(mem.SyclQueue().get_pointer_type(mobj.pointer), 'shared')
+        self.assertEqual(mobj._usm_type(), 'shared')
+
+    def test_memory_cpu_context (self):
+        mobj = self._create_memory()
+
+        # CPU context
+        with dppl.device_context(dppl.device_type.cpu):
+            self.assertEqual(mem.SyclQueue().get_pointer_type(mobj.pointer), 'unknown')
+            self.assertEqual(mobj._usm_type(), 'shared')
+
+    def test_memory_gpu_context (self):
+        mobj = self._create_memory()
+
+        # GPU context
+        with dppl.device_context(dppl.device_type.gpu):
+            self.assertEqual(mem.SyclQueue().get_pointer_type(mobj.pointer), 'unknown')
+            self.assertEqual(mobj._usm_type(), 'shared')

setup.py

@@ -25,8 +25,8 @@
 import sys
 import versioneer
 
-from Cython.Build import cythonize
 from setuptools import setup, Extension, find_packages
+from Cython.Build import cythonize
 
 import numpy as np
 
@@ -78,7 +78,7 @@ def get_other_cxxflags():
         # what compiler we are using.
         return ['/Ox', '/std:c++17']
 
-def getpyexts():
+def extensions():
     # Security flags
     eca = get_sdl_cflags()
     ela = get_sdl_ldflags()
@@ -104,17 +104,22 @@ def getpyexts():
     elif IS_WIN:
         runtime_library_dirs = []
 
-    exts = cythonize(Extension('dppl._sycl_core',
-                               [os.path.abspath('dppl/sycl_core.pyx'),],
-                                depends=[dppl_sycl_interface_include,],
-                                include_dirs=[np.get_include(),
-                                              dppl_sycl_interface_include],
-                                extra_compile_args=eca + get_other_cxxflags(),
-                                extra_link_args=ela,
-                                libraries=libs,
-                                library_dirs=librarys,
-                                runtime_library_dirs=runtime_library_dirs,
-                                language='c++'))
+    extension_args = {
+        "depends": [dppl_sycl_interface_include,],
+        "include_dirs": [np.get_include(), dppl_sycl_interface_include],
+        "extra_compile_args": eca + get_other_cxxflags(),
+        "extra_link_args": ela, "libraries": libs, "library_dirs": librarys,
+        "runtime_library_dirs": runtime_library_dirs, "language": 'c++',
+    }
+
+    extensions = [
+        Extension('dppl._sycl_core', [os.path.abspath('dppl/sycl_core.pyx'),],
+            **extension_args),
+        Extension('dppl._memory', [os.path.abspath('dppl/_memory.pyx'),],
+            **extension_args),
+    ]
+
+    exts = cythonize(extensions)
     return exts
 
 setup(
@@ -126,7 +131,7 @@ def getpyexts():
     author="Intel Corporation",
     url='https://github.com/IntelPython/PyDPPL',
     packages=find_packages(include=["dppl", "dppl.*"]),
-    ext_modules = getpyexts(),
+    ext_modules = extensions(),
     setup_requires=requirements,
     cffi_modules=[
        "./dppl/opencl_core.py:ffi"