diff --git a/dpctl-capi/include/dpctl_sycl_device_interface.h b/dpctl-capi/include/dpctl_sycl_device_interface.h index 4976f3497f..1c8d57802d 100644 --- a/dpctl-capi/include/dpctl_sycl_device_interface.h +++ b/dpctl-capi/include/dpctl_sycl_device_interface.h @@ -177,6 +177,28 @@ DPCTL_API uint32_t DPCTLDevice_GetMaxComputeUnits(__dpctl_keep const DPCTLSyclDeviceRef DRef); +/*! + * @brief Wrapper over device.get_info(). + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the valid result if device exists else returns 0. + * @ingroup DeviceInterface + */ +DPCTL_API +uint64_t +DPCTLDevice_GetGlobalMemSize(__dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over device.get_info(). + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the valid result if device exists else returns 0. + * @ingroup DeviceInterface + */ +DPCTL_API +uint64_t +DPCTLDevice_GetLocalMemSize(__dpctl_keep const DPCTLSyclDeviceRef DRef); + /*! * @brief Wrapper for get_info(). * diff --git a/dpctl-capi/source/dpctl_sycl_device_interface.cpp b/dpctl-capi/source/dpctl_sycl_device_interface.cpp index bbf510f5c9..8ef71b65c4 100644 --- a/dpctl-capi/source/dpctl_sycl_device_interface.cpp +++ b/dpctl-capi/source/dpctl_sycl_device_interface.cpp @@ -186,6 +186,37 @@ DPCTLDevice_GetMaxComputeUnits(__dpctl_keep const DPCTLSyclDeviceRef DRef) return nComputeUnits; } +uint64_t +DPCTLDevice_GetGlobalMemSize(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + uint64_t GlobalMemSize = 0; + auto D = unwrap(DRef); + if (D) { + try { + GlobalMemSize = D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return GlobalMemSize; +} + +uint64_t DPCTLDevice_GetLocalMemSize(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + uint64_t LocalMemSize = 0; + auto D = unwrap(DRef); + if (D) { + try { + LocalMemSize = D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return LocalMemSize; +} + uint32_t DPCTLDevice_GetMaxWorkItemDims(__dpctl_keep const DPCTLSyclDeviceRef DRef) { diff --git a/dpctl-capi/tests/test_sycl_device_interface.cpp b/dpctl-capi/tests/test_sycl_device_interface.cpp index 5a7f898962..25ff7c3f47 100644 --- a/dpctl-capi/tests/test_sycl_device_interface.cpp +++ b/dpctl-capi/tests/test_sycl_device_interface.cpp @@ -132,6 +132,20 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetMaxComputeUnits) EXPECT_TRUE(n > 0); } +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetGlobalMemSize) +{ + size_t gm_sz = 0; + EXPECT_NO_FATAL_FAILURE(gm_sz = DPCTLDevice_GetGlobalMemSize(DRef)); + EXPECT_TRUE(gm_sz > 0); +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetLocalMemSize) +{ + size_t lm_sz = 0; + EXPECT_NO_FATAL_FAILURE(lm_sz = DPCTLDevice_GetLocalMemSize(DRef)); + EXPECT_TRUE(lm_sz > 0); +} + TEST_P(TestDPCTLSyclDeviceInterface, ChkGetMaxWorkItemDims) { size_t n = 0; diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index aa29740d82..d968a7d42f 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, uint64_t +from libc.stdint cimport int64_t, uint32_t from libcpp cimport bool @@ -149,6 +149,8 @@ cdef extern from "dpctl_sycl_device_interface.h": cdef _backend_type DPCTLDevice_GetBackend(const DPCTLSyclDeviceRef) cdef _device_type DPCTLDevice_GetDeviceType(const DPCTLSyclDeviceRef) cdef const char *DPCTLDevice_GetDriverVersion(const DPCTLSyclDeviceRef DRef) + cdef size_t DPCTLDevice_GetGlobalMemSize(const DPCTLSyclDeviceRef DRef) + cdef size_t DPCTLDevice_GetLocalMemSize(const DPCTLSyclDeviceRef DRef) cdef uint32_t DPCTLDevice_GetMaxComputeUnits(const DPCTLSyclDeviceRef DRef) cdef uint32_t DPCTLDevice_GetMaxNumSubGroups(const DPCTLSyclDeviceRef DRef) cdef size_t DPCTLDevice_GetMaxWorkGroupSize(const DPCTLSyclDeviceRef DRef) @@ -239,9 +241,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 size_t DPCTLEvent_GetProfilingInfoSubmit(DPCTLSyclEventRef ERef) + cdef size_t DPCTLEvent_GetProfilingInfoStart(DPCTLSyclEventRef ERef) + cdef size_t DPCTLEvent_GetProfilingInfoEnd(DPCTLSyclEventRef ERef) cdef extern from "dpctl_sycl_kernel_interface.h": diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index ea3397875a..48fd7a6e65 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -34,11 +34,13 @@ from ._backend cimport ( # noqa: E211 DPCTLDevice_GetBackend, DPCTLDevice_GetDeviceType, DPCTLDevice_GetDriverVersion, + DPCTLDevice_GetGlobalMemSize, DPCTLDevice_GetImage2dMaxHeight, DPCTLDevice_GetImage2dMaxWidth, DPCTLDevice_GetImage3dMaxDepth, DPCTLDevice_GetImage3dMaxHeight, DPCTLDevice_GetImage3dMaxWidth, + DPCTLDevice_GetLocalMemSize, DPCTLDevice_GetMaxComputeUnits, DPCTLDevice_GetMaxNumSubGroups, DPCTLDevice_GetMaxReadImageArgs, @@ -656,6 +658,22 @@ cdef class SyclDevice(_SyclDevice): """ return DPCTLDevice_GetPreferredVectorWidthHalf(self._device_ref) + @property + def global_mem_size(self): + """ Returns the size of global memory on this device in bytes. + """ + cdef size_t global_mem_size = 0 + global_mem_size = DPCTLDevice_GetGlobalMemSize(self._device_ref) + return global_mem_size + + @property + def local_mem_size(self): + """ Returns the size of local memory on this device in bytes. + """ + cdef size_t local_mem_size = 0 + local_mem_size = DPCTLDevice_GetLocalMemSize(self._device_ref) + return local_mem_size + @property def vendor(self): """ Returns the device vendor name as a string. diff --git a/dpctl/tests/test_sycl_device.py b/dpctl/tests/test_sycl_device.py index 4dd52c782f..5b3f30c89d 100644 --- a/dpctl/tests/test_sycl_device.py +++ b/dpctl/tests/test_sycl_device.py @@ -61,6 +61,16 @@ def check_get_max_compute_units(device): assert max_compute_units > 0 +def check_get_global_mem_size(device): + global_mem_size = device.global_mem_size + assert global_mem_size > 0 + + +def check_get_local_mem_size(device): + local_mem_size = device.local_mem_size + assert local_mem_size > 0 + + def check_get_max_work_item_dims(device): max_work_item_dims = device.max_work_item_dims assert max_work_item_dims > 0 @@ -529,6 +539,8 @@ def check_repr(device): check_create_sub_devices_by_affinity_next_partitionable, check_print_device_info, check_repr, + check_get_global_mem_size, + check_get_local_mem_size, ] diff --git a/dpctl/tests/test_sycl_kernel_submit.py b/dpctl/tests/test_sycl_kernel_submit.py index 358b07cb90..3f3a03bb97 100644 --- a/dpctl/tests/test_sycl_kernel_submit.py +++ b/dpctl/tests/test_sycl_kernel_submit.py @@ -87,7 +87,7 @@ def test_create_program_from_source(ctype_str, dtype, ctypes_ctor): q.submit(axpyKernel, args, r).wait() ref_c = a * np.array(d, dtype=dtype) + b host_dt, device_dt = timer.dt - assert host_dt > device_dt + assert type(host_dt) is float and type(device_dt) is float assert np.allclose(c, ref_c), "Failed for {}".format(r) for gr, lr in ( @@ -106,5 +106,5 @@ def test_create_program_from_source(ctype_str, dtype, ctypes_ctor): q.submit(axpyKernel, args, gr, lr, [dpctl.SyclEvent()]).wait() ref_c = a * np.array(d, dtype=dtype) + b host_dt, device_dt = timer.dt - assert host_dt > device_dt + assert type(host_dt) is float and type(device_dt) is float assert np.allclose(c, ref_c), "Faled for {}, {}".formatg(r, lr)