From 8507a501d597232c24d5b36a94bbe3dbb7dbd198 Mon Sep 17 00:00:00 2001 From: etotmeni Date: Fri, 26 Feb 2021 07:25:46 -0600 Subject: [PATCH 1/6] Add sub_group_independent_forward_progress and preferred_vector_width_char --- .../include/dpctl_sycl_device_interface.h | 24 +++++++ .../source/dpctl_sycl_device_interface.cpp | 34 +++++++++ .../tests/test_sycl_device_interface.cpp | 70 +++++++++++++++++++ 3 files changed, 128 insertions(+) diff --git a/dpctl-capi/include/dpctl_sycl_device_interface.h b/dpctl-capi/include/dpctl_sycl_device_interface.h index 79d380d9eb..378172e743 100644 --- a/dpctl-capi/include/dpctl_sycl_device_interface.h +++ b/dpctl-capi/include/dpctl_sycl_device_interface.h @@ -252,3 +252,27 @@ DPCTL_API bool DPCTLDevice_AreEq(__dpctl_keep const DPCTLSyclDeviceRef DevRef1, __dpctl_keep const DPCTLSyclDeviceRef DevRef2); DPCTL_C_EXTERN_C_END + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns true if the device supports independent forward progress of + * sub-groups with respect to other sub-groups in the same work-group. + */ +DPCTL_API +bool DPCTLDevice_GetSubGroupIndependentForwardProgress( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthChar( + __dpctl_keep const DPCTLSyclDeviceRef DRef); diff --git a/dpctl-capi/source/dpctl_sycl_device_interface.cpp b/dpctl-capi/source/dpctl_sycl_device_interface.cpp index d534dd8d7a..c523bd8a05 100644 --- a/dpctl-capi/source/dpctl_sycl_device_interface.cpp +++ b/dpctl-capi/source/dpctl_sycl_device_interface.cpp @@ -395,3 +395,37 @@ bool DPCTLDevice_AreEq(__dpctl_keep const DPCTLSyclDeviceRef DevRef1, return false; return (*unwrap(DevRef1) == *unwrap(DevRef2)); } + +bool DPCTLDevice_GetSubGroupIndependentForwardProgress( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + bool SubGroupProgress = false; + auto D = unwrap(DRef); + if (D) { + try { + SubGroupProgress = D->get_info< + info::device::sub_group_independent_forward_progress>(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return SubGroupProgress; +} + +uint32_t DPCTLDevice_GetPreferredVectorWidthChar( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_char = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_char = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_char; +} diff --git a/dpctl-capi/tests/test_sycl_device_interface.cpp b/dpctl-capi/tests/test_sycl_device_interface.cpp index 65032446ba..1d9fe91723 100644 --- a/dpctl-capi/tests/test_sycl_device_interface.cpp +++ b/dpctl-capi/tests/test_sycl_device_interface.cpp @@ -391,3 +391,73 @@ TEST_F(TestDPCTLSyclDeviceInterface, CheckLevel0GPU_IsGPU) EXPECT_TRUE(DPCTLDevice_IsGPU(OpenCL_Level0_gpu)); } + +TEST_F(TestDPCTLSyclDeviceInterface, + CheckOCLCPU_GetSubGroupIndependentForwardProgress) +{ + if (!OpenCL_cpu) + GTEST_SKIP_("Skipping as no OpenCL CPU device found."); + + auto sub_group_progress = + DPCTLDevice_GetSubGroupIndependentForwardProgress(OpenCL_cpu); + auto D = reinterpret_cast(OpenCL_cpu); + auto get_sub_group_progress = + D->get_info(); + EXPECT_TRUE(get_sub_group_progress == sub_group_progress); +} + +TEST_F(TestDPCTLSyclDeviceInterface, + CheckOCLGPU_GetSubGroupIndependentForwardProgress) +{ + if (!OpenCL_gpu) + GTEST_SKIP_("Skipping as no OpenCL GPU device found."); + + auto sub_group_progress = + DPCTLDevice_GetSubGroupIndependentForwardProgress(OpenCL_gpu); + auto D = reinterpret_cast(OpenCL_gpu); + auto get_sub_group_progress = + D->get_info(); + EXPECT_TRUE(get_sub_group_progress == sub_group_progress); +} + +TEST_F(TestDPCTLSyclDeviceInterface, + CheckLevel0GPU_GetSubGroupIndependentForwardProgress) +{ + if (!OpenCL_Level0_gpu) + GTEST_SKIP_("Skipping as no Level0 GPU device found."); + + auto sub_group_progress = + DPCTLDevice_GetSubGroupIndependentForwardProgress(OpenCL_Level0_gpu); + auto D = reinterpret_cast(OpenCL_Level0_gpu); + auto get_sub_group_progress = + D->get_info(); + EXPECT_TRUE(get_sub_group_progress == sub_group_progress); +} + +TEST_F(TestDPCTLSyclDeviceInterface, CheckOCLCPU_GetPreferredVectorWidthChar) +{ + if (!OpenCL_cpu) + GTEST_SKIP_("Skipping as no OpenCL CPU device found."); + + auto VectorWidthChar = DPCTLDevice_GetPreferredVectorWidthChar(OpenCL_cpu); + EXPECT_TRUE(VectorWidthChar != 0); +} + +TEST_F(TestDPCTLSyclDeviceInterface, CheckOCLGPU_GetPreferredVectorWidthChar) +{ + if (!OpenCL_gpu) + GTEST_SKIP_("Skipping as no OpenCL CPU device found."); + + auto VectorWidthChar = DPCTLDevice_GetPreferredVectorWidthChar(OpenCL_gpu); + EXPECT_TRUE(VectorWidthChar != 0); +} + +TEST_F(TestDPCTLSyclDeviceInterface, CheckLevel0GPU_GetPreferredVectorWidthChar) +{ + if (!OpenCL_Level0_gpu) + GTEST_SKIP_("Skipping as no Level0 GPU device found."); + + auto VectorWidthChar = + DPCTLDevice_GetPreferredVectorWidthChar(OpenCL_Level0_gpu); + EXPECT_TRUE(VectorWidthChar != 0); +} From 952cb1842aec58314e438e7b2d16c2cc44205bd2 Mon Sep 17 00:00:00 2001 From: etotmeni Date: Tue, 2 Mar 2021 07:10:44 -0600 Subject: [PATCH 2/6] Add tests --- .../tests/test_sycl_device_interface.cpp | 30 +++++++++++++++++++ 1 file changed, 30 insertions(+) diff --git a/dpctl-capi/tests/test_sycl_device_interface.cpp b/dpctl-capi/tests/test_sycl_device_interface.cpp index b913317d84..65028decf8 100644 --- a/dpctl-capi/tests/test_sycl_device_interface.cpp +++ b/dpctl-capi/tests/test_sycl_device_interface.cpp @@ -288,6 +288,36 @@ TEST_P(TestDPCTLSyclDeviceInterface, Chk_IsHost) EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); } +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetSubGroupIndependentForwardProgress) +{ + DPCTLSyclDeviceRef DRef = nullptr; + bool sub_group_progress = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE( + sub_group_progress = + DPCTLDevice_GetSubGroupIndependentForwardProgress(DRef)); + auto D = reinterpret_cast(DRef); + auto get_sub_group_progress = + D->get_info(); + EXPECT_TRUE(get_sub_group_progress == sub_group_progress); + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthChar) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_char = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE(vector_width_char = + DPCTLDevice_GetPreferredVectorWidthChar(DRef)); + EXPECT_TRUE(vector_width_char != 0); + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + INSTANTIATE_TEST_SUITE_P(DPCTLDevice_Fns, TestDPCTLSyclDeviceInterface, ::testing::Values("opencl", From e153c768ff9785ff51d8588e9ec0a680008fb2d0 Mon Sep 17 00:00:00 2001 From: etotmeni Date: Tue, 2 Mar 2021 07:53:11 -0600 Subject: [PATCH 3/6] Add python api --- dpctl/_backend.pxd | 2 ++ dpctl/_sycl_device.pxd | 2 ++ dpctl/_sycl_device.pyx | 26 ++++++++++++++++++++++++++ dpctl/tests/test_sycl_device.py | 14 ++++++++++++++ 4 files changed, 44 insertions(+) diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 4e28a9968b..c6bcb7f518 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -126,6 +126,8 @@ cdef extern from "dpctl_sycl_device_interface.h": cdef bool DPCTLDevice_IsGPU(const DPCTLSyclDeviceRef DRef) cdef bool DPCTLDevice_IsHost(const DPCTLSyclDeviceRef DRef) cdef bool DPCTLDevice_IsHostUnifiedMemory(const DPCTLSyclDeviceRef DRef) + cdef bool DPCTLDevice_GetSubGroupIndependentForwardProgress(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthChar(const DPCTLSyclDeviceRef DRef) cdef extern from "dpctl_sycl_device_selector_interface.h": diff --git a/dpctl/_sycl_device.pxd b/dpctl/_sycl_device.pxd index 2e0c76a296..89504651af 100644 --- a/dpctl/_sycl_device.pxd +++ b/dpctl/_sycl_device.pxd @@ -63,6 +63,8 @@ cdef class _SyclDevice: cpdef is_cpu(self) cpdef is_gpu(self) cpdef is_host(self) + cpdef get_sub_group_independent_forward_progress(self) + cpdef get_preferred_vector_width_char(self) cdef class SyclDevice(_SyclDevice): diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index e8ea50bfc5..9c75794ce2 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -56,6 +56,8 @@ from ._backend cimport ( DPCTLSyclDeviceRef, DPCTLSyclDeviceSelectorRef, DPCTLSyclDeviceType, + DPCTLDevice_GetSubGroupIndependentForwardProgress, + DPCTLDevice_GetPreferredVectorWidthChar ) from . import backend_type, device_type import warnings @@ -259,6 +261,18 @@ cdef class _SyclDevice: """ return int(self._device_ref) + cpdef get_sub_group_independent_forward_progress(self): + """ Returns true if the device supports independent forward progress of + sub-groups with respect to other sub-groups in the same work-group. + """ + return self._get_sub_group_independent_forward_progress + + cpdef get_preferred_vector_width_char(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._get_preferred_vector_width_char + @property def __name__(self): return "SyclDevice" @@ -325,6 +339,12 @@ cdef class SyclDevice(_SyclDevice): device._cpu_device = DPCTLDevice_IsCPU(DRef) device._gpu_device = DPCTLDevice_IsGPU(DRef) device._host_device = DPCTLDevice_IsHost(DRef) + device._get_sub_group_independent_forward_progress = ( + DPCTLDevice_GetSubGroupIndependentForwardProgress(DRef) + ) + device._get_preferred_vector_width_char = ( + DPCTLDevice_GetPreferredVectorWidthChar(DRef) + ) @staticmethod cdef SyclDevice _create(DPCTLSyclDeviceRef dref): @@ -351,6 +371,12 @@ cdef class SyclDevice(_SyclDevice): self._cpu_device = other._cpu_device self._gpu_device = other._gpu_device self._host_device = other._host_device + self._get_sub_group_independent_forward_progress = ( + other._get_sub_group_independent_forward_progress + ) + self._get_preferred_vector_width_char = ( + other._get_preferred_vector_width_char + ) cdef int _init_from_selector(self, DPCTLSyclDeviceSelectorRef DSRef): # Initialize the attributes of the SyclDevice object diff --git a/dpctl/tests/test_sycl_device.py b/dpctl/tests/test_sycl_device.py index 7df90c6df3..5eaac36200 100644 --- a/dpctl/tests/test_sycl_device.py +++ b/dpctl/tests/test_sycl_device.py @@ -128,6 +128,20 @@ def check_is_host(device): pytest.fail("is_hostcall failed") +def check_get_sub_group_independent_forward_progress(device): + try: + device.get_sub_group_independent_forward_progress() + except Exception: + pytest.fail("get_sub_group_independent_forward_progress call failed") + + +def check_get_preferred_vector_width_char(device): + try: + device.get_preferred_vector_width_char() + except Exception: + pytest.fail("get_preferred_vector_width_char call failed") + + list_of_checks = [ check_get_max_compute_units, check_get_max_work_item_dims, From 7baa7a111eb2eee88a77bdf3fc96d6ca35835aba Mon Sep 17 00:00:00 2001 From: etotmeni Date: Wed, 3 Mar 2021 08:28:46 -0600 Subject: [PATCH 4/6] Fixes + tests --- dpctl/_sycl_device.pxd | 2 ++ dpctl/_sycl_device.pyx | 16 ++++++---------- dpctl/tests/test_sycl_device.py | 2 ++ 3 files changed, 10 insertions(+), 10 deletions(-) diff --git a/dpctl/_sycl_device.pxd b/dpctl/_sycl_device.pxd index 89504651af..33b5f5f1bc 100644 --- a/dpctl/_sycl_device.pxd +++ b/dpctl/_sycl_device.pxd @@ -46,6 +46,8 @@ cdef class _SyclDevice: cdef uint32_t _max_num_sub_groups cdef bool _int64_base_atomics cdef bool _int64_extended_atomics + cdef bool _sub_group_independent_forward_progress + cdef uint32_t _preferred_vector_width_char cdef DPCTLSyclDeviceRef get_device_ref(self) cpdef get_backend(self) cpdef get_device_name(self) diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index 9c75794ce2..e8e4b4beb7 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -265,13 +265,13 @@ cdef class _SyclDevice: """ Returns true if the device supports independent forward progress of sub-groups with respect to other sub-groups in the same work-group. """ - return self._get_sub_group_independent_forward_progress + return self._sub_group_independent_forward_progress cpdef get_preferred_vector_width_char(self): """ Returns the preferred native vector width size for built-in scalar types that can be put into vectors. """ - return self._get_preferred_vector_width_char + return self._preferred_vector_width_char @property def __name__(self): @@ -339,10 +339,10 @@ cdef class SyclDevice(_SyclDevice): device._cpu_device = DPCTLDevice_IsCPU(DRef) device._gpu_device = DPCTLDevice_IsGPU(DRef) device._host_device = DPCTLDevice_IsHost(DRef) - device._get_sub_group_independent_forward_progress = ( + device._sub_group_independent_forward_progress = ( DPCTLDevice_GetSubGroupIndependentForwardProgress(DRef) ) - device._get_preferred_vector_width_char = ( + device._preferred_vector_width_char = ( DPCTLDevice_GetPreferredVectorWidthChar(DRef) ) @@ -371,12 +371,8 @@ cdef class SyclDevice(_SyclDevice): self._cpu_device = other._cpu_device self._gpu_device = other._gpu_device self._host_device = other._host_device - self._get_sub_group_independent_forward_progress = ( - other._get_sub_group_independent_forward_progress - ) - self._get_preferred_vector_width_char = ( - other._get_preferred_vector_width_char - ) + self._sub_group_independent_forward_progress = other._sub_group_independent_forward_progress + self._preferred_vector_width_char = other._preferred_vector_width_char cdef int _init_from_selector(self, DPCTLSyclDeviceSelectorRef DSRef): # Initialize the attributes of the SyclDevice object diff --git a/dpctl/tests/test_sycl_device.py b/dpctl/tests/test_sycl_device.py index 5eaac36200..9134e811a7 100644 --- a/dpctl/tests/test_sycl_device.py +++ b/dpctl/tests/test_sycl_device.py @@ -154,6 +154,8 @@ def check_get_preferred_vector_width_char(device): check_is_cpu, check_is_gpu, check_is_host, + check_get_sub_group_independent_forward_progress, + check_get_preferred_vector_width_char, ] From 68620f67f26d828cd602de80b2d718376deded06 Mon Sep 17 00:00:00 2001 From: etotmeni Date: Wed, 10 Mar 2021 07:22:54 -0600 Subject: [PATCH 5/6] Add preferred_vector_width funcs --- .../include/dpctl_sycl_device_interface.h | 72 +++++++++++++ .../source/dpctl_sycl_device_interface.cpp | 102 ++++++++++++++++++ .../tests/test_sycl_device_interface.cpp | 93 ++++++++++++++++ dpctl/_backend.pxd | 6 ++ dpctl/_sycl_device.pxd | 12 +++ dpctl/_sycl_device.pyx | 68 +++++++++++- dpctl/tests/test_sycl_device.py | 48 +++++++++ 7 files changed, 400 insertions(+), 1 deletion(-) diff --git a/dpctl-capi/include/dpctl_sycl_device_interface.h b/dpctl-capi/include/dpctl_sycl_device_interface.h index 5984a0b5b1..293c511f07 100644 --- a/dpctl-capi/include/dpctl_sycl_device_interface.h +++ b/dpctl-capi/include/dpctl_sycl_device_interface.h @@ -290,3 +290,75 @@ bool DPCTLDevice_GetSubGroupIndependentForwardProgress( DPCTL_API uint32_t DPCTLDevice_GetPreferredVectorWidthChar( __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthShort( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthInt( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthLong( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthFloat( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthDouble( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a sycl::device + * @return Returns the preferred native vector width size for built-in scalar + * types that can be put into vectors. + */ +DPCTL_API +uint32_t DPCTLDevice_GetPreferredVectorWidthHalf( + __dpctl_keep const DPCTLSyclDeviceRef DRef); diff --git a/dpctl-capi/source/dpctl_sycl_device_interface.cpp b/dpctl-capi/source/dpctl_sycl_device_interface.cpp index 285d18c586..ff37d4834b 100644 --- a/dpctl-capi/source/dpctl_sycl_device_interface.cpp +++ b/dpctl-capi/source/dpctl_sycl_device_interface.cpp @@ -421,3 +421,105 @@ uint32_t DPCTLDevice_GetPreferredVectorWidthChar( } return vector_width_char; } + +uint32_t DPCTLDevice_GetPreferredVectorWidthShort( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_short = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_short = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_short; +} + +uint32_t DPCTLDevice_GetPreferredVectorWidthInt( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_int = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_int = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_int; +} + +uint32_t DPCTLDevice_GetPreferredVectorWidthLong( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_long = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_long = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_long; +} + +uint32_t DPCTLDevice_GetPreferredVectorWidthFloat( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_float = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_float = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_float; +} + +uint32_t DPCTLDevice_GetPreferredVectorWidthDouble( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_double = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_double = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_double; +} + +uint32_t DPCTLDevice_GetPreferredVectorWidthHalf( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + size_t vector_width_half = 0; + auto D = unwrap(DRef); + if (D) { + try { + vector_width_half = + D->get_info(); + } catch (runtime_error const &re) { + // \todo log error + std::cerr << re.what() << '\n'; + } + } + return vector_width_half; +} diff --git a/dpctl-capi/tests/test_sycl_device_interface.cpp b/dpctl-capi/tests/test_sycl_device_interface.cpp index e742161b7f..36a3f85c66 100644 --- a/dpctl-capi/tests/test_sycl_device_interface.cpp +++ b/dpctl-capi/tests/test_sycl_device_interface.cpp @@ -24,6 +24,7 @@ /// //===----------------------------------------------------------------------===// +#include "../helper/include/dpctl_utils_helper.h" #include "dpctl_sycl_device_interface.h" #include "dpctl_sycl_device_selector_interface.h" #include "dpctl_sycl_platform_interface.h" @@ -285,6 +286,98 @@ TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthChar) EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); } +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthShort) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_short = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE(vector_width_short = + DPCTLDevice_GetPreferredVectorWidthShort(DRef)); + EXPECT_TRUE(vector_width_short != 0); + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthInt) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_int = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE(vector_width_int = + DPCTLDevice_GetPreferredVectorWidthInt(DRef)); + EXPECT_TRUE(vector_width_int != 0); + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthLong) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_long = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE(vector_width_long = + DPCTLDevice_GetPreferredVectorWidthLong(DRef)); + EXPECT_TRUE(vector_width_long != 0); + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthFloat) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_float = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE(vector_width_float = + DPCTLDevice_GetPreferredVectorWidthFloat(DRef)); + EXPECT_TRUE(vector_width_float != 0); + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthDouble) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_double = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE( + vector_width_double = DPCTLDevice_GetPreferredVectorWidthDouble(DRef)); + if (DPCTLDevice_HasAspect(DRef, DPCTL_SyclAspectToDPCTLAspectType( + DPCTL_StrToAspectType("fp64")))) + { + EXPECT_TRUE(vector_width_double != 0); + } + else { + EXPECT_TRUE(vector_width_double == 0); + } + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + +TEST_P(TestDPCTLSyclDeviceInterface, Chk_GetPreferredVectorWidthHalf) +{ + DPCTLSyclDeviceRef DRef = nullptr; + size_t vector_width_half = 0; + EXPECT_NO_FATAL_FAILURE(DRef = DPCTLDevice_CreateFromSelector(DSRef)); + if (!DRef) + GTEST_SKIP_("Device not found"); + EXPECT_NO_FATAL_FAILURE(vector_width_half = + DPCTLDevice_GetPreferredVectorWidthHalf(DRef)); + if (DPCTLDevice_HasAspect(DRef, DPCTL_SyclAspectToDPCTLAspectType( + DPCTL_StrToAspectType("fp16")))) + { + EXPECT_TRUE(vector_width_half != 0); + } + else { + EXPECT_TRUE(vector_width_half == 0); + } + EXPECT_NO_FATAL_FAILURE(DPCTLDevice_Delete(DRef)); +} + INSTANTIATE_TEST_SUITE_P(DPCTLDevice_Fns, TestDPCTLSyclDeviceInterface, ::testing::Values("opencl", diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 03bddb1823..ec01afdce0 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -147,6 +147,12 @@ cdef extern from "dpctl_sycl_device_interface.h": cdef bool DPCTLDevice_IsHostUnifiedMemory(const DPCTLSyclDeviceRef DRef) cdef bool DPCTLDevice_GetSubGroupIndependentForwardProgress(const DPCTLSyclDeviceRef DRef) cdef uint32_t DPCTLDevice_GetPreferredVectorWidthChar(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthShort(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthInt(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthLong(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthFloat(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthDouble(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetPreferredVectorWidthHalf(const DPCTLSyclDeviceRef DRef) cpdef bool DPCTLDevice_HasAspect( const DPCTLSyclDeviceRef DRef, DPCTLSyclAspectType AT) diff --git a/dpctl/_sycl_device.pxd b/dpctl/_sycl_device.pxd index fb12e0c711..3243c8c329 100644 --- a/dpctl/_sycl_device.pxd +++ b/dpctl/_sycl_device.pxd @@ -46,6 +46,12 @@ cdef class _SyclDevice: cdef uint32_t _max_num_sub_groups cdef bool _sub_group_independent_forward_progress cdef uint32_t _preferred_vector_width_char + cdef uint32_t _preferred_vector_width_short + cdef uint32_t _preferred_vector_width_int + cdef uint32_t _preferred_vector_width_long + cdef uint32_t _preferred_vector_width_float + cdef uint32_t _preferred_vector_width_double + cdef uint32_t _preferred_vector_width_half cdef DPCTLSyclDeviceRef get_device_ref(self) cpdef get_backend(self) cpdef get_device_name(self) @@ -63,6 +69,12 @@ cdef class _SyclDevice: cpdef is_host(self) cpdef get_sub_group_independent_forward_progress(self) cpdef get_preferred_vector_width_char(self) + cpdef get_preferred_vector_width_short(self) + cpdef get_preferred_vector_width_int(self) + cpdef get_preferred_vector_width_long(self) + cpdef get_preferred_vector_width_float(self) + cpdef get_preferred_vector_width_double(self) + cpdef get_preferred_vector_width_half(self) cdef class SyclDevice(_SyclDevice): diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index b642926f03..324d4d6ead 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -53,7 +53,13 @@ from ._backend cimport ( DPCTLDevice_HasAspect, DPCTLSyclDeviceType, DPCTLDevice_GetSubGroupIndependentForwardProgress, - DPCTLDevice_GetPreferredVectorWidthChar + DPCTLDevice_GetPreferredVectorWidthChar, + DPCTLDevice_GetPreferredVectorWidthShort, + DPCTLDevice_GetPreferredVectorWidthInt, + DPCTLDevice_GetPreferredVectorWidthLong, + DPCTLDevice_GetPreferredVectorWidthFloat, + DPCTLDevice_GetPreferredVectorWidthDouble, + DPCTLDevice_GetPreferredVectorWidthHalf, ) from . import backend_type, device_type import warnings @@ -254,6 +260,42 @@ cdef class _SyclDevice: """ return self._preferred_vector_width_char + cpdef get_preferred_vector_width_short(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._preferred_vector_width_short + + cpdef get_preferred_vector_width_int(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._preferred_vector_width_int + + cpdef get_preferred_vector_width_long(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._preferred_vector_width_long + + cpdef get_preferred_vector_width_float(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._preferred_vector_width_float + + cpdef get_preferred_vector_width_double(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._preferred_vector_width_double + + cpdef get_preferred_vector_width_half(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return self._preferred_vector_width_half + @property def __name__(self): return "SyclDevice" @@ -322,6 +364,24 @@ cdef class SyclDevice(_SyclDevice): device._preferred_vector_width_char = ( DPCTLDevice_GetPreferredVectorWidthChar(DRef) ) + device._preferred_vector_width_short = ( + DPCTLDevice_GetPreferredVectorWidthShort(DRef) + ) + device._preferred_vector_width_int = ( + DPCTLDevice_GetPreferredVectorWidthInt(DRef) + ) + device._preferred_vector_width_long = ( + DPCTLDevice_GetPreferredVectorWidthLong(DRef) + ) + device._preferred_vector_width_float = ( + DPCTLDevice_GetPreferredVectorWidthFloat(DRef) + ) + device._preferred_vector_width_double = ( + DPCTLDevice_GetPreferredVectorWidthDouble(DRef) + ) + device._preferred_vector_width_half = ( + DPCTLDevice_GetPreferredVectorWidthHalf(DRef) + ) @staticmethod cdef SyclDevice _create(DPCTLSyclDeviceRef dref): @@ -348,6 +408,12 @@ cdef class SyclDevice(_SyclDevice): self._host_device = other._host_device self._sub_group_independent_forward_progress = other._sub_group_independent_forward_progress self._preferred_vector_width_char = other._preferred_vector_width_char + self._preferred_vector_width_short = other._preferred_vector_width_short + self._preferred_vector_width_int = other._preferred_vector_width_int + self._preferred_vector_width_long = other._preferred_vector_width_long + self._preferred_vector_width_float = other._preferred_vector_width_float + self._preferred_vector_width_double = other._preferred_vector_width_double + self._preferred_vector_width_half = other._preferred_vector_width_half cdef int _init_from_selector(self, DPCTLSyclDeviceSelectorRef DSRef): # Initialize the attributes of the SyclDevice object diff --git a/dpctl/tests/test_sycl_device.py b/dpctl/tests/test_sycl_device.py index 708bc80810..0265add5c8 100644 --- a/dpctl/tests/test_sycl_device.py +++ b/dpctl/tests/test_sycl_device.py @@ -254,6 +254,48 @@ def check_get_preferred_vector_width_char(device): pytest.fail("get_preferred_vector_width_char call failed") +def check_get_preferred_vector_width_short(device): + try: + device.get_preferred_vector_width_short() + except Exception: + pytest.fail("get_preferred_vector_width_short call failed") + + +def check_get_preferred_vector_width_int(device): + try: + device.get_preferred_vector_width_int() + except Exception: + pytest.fail("get_preferred_vector_width_int call failed") + + +def check_get_preferred_vector_width_long(device): + try: + device.get_preferred_vector_width_long() + except Exception: + pytest.fail("get_preferred_vector_width_long call failed") + + +def check_get_preferred_vector_width_float(device): + try: + device.get_preferred_vector_width_float() + except Exception: + pytest.fail("get_preferred_vector_width_float call failed") + + +def check_get_preferred_vector_width_double(device): + try: + device.get_preferred_vector_width_double() + except Exception: + pytest.fail("get_preferred_vector_width_double call failed") + + +def check_get_preferred_vector_width_half(device): + try: + device.get_preferred_vector_width_half() + except Exception: + pytest.fail("get_preferred_vector_width_half call failed") + + list_of_checks = [ check_get_max_compute_units, check_get_max_work_item_dims, @@ -266,6 +308,12 @@ def check_get_preferred_vector_width_char(device): check_is_host, check_get_sub_group_independent_forward_progress, check_get_preferred_vector_width_char, + check_get_preferred_vector_width_short, + check_get_preferred_vector_width_int, + check_get_preferred_vector_width_long, + check_get_preferred_vector_width_float, + check_get_preferred_vector_width_double, + check_get_preferred_vector_width_half, check_has_aspect_host, check_has_aspect_cpu, check_has_aspect_gpu, From c0106e14095cd06393d0da825814026c0236e459 Mon Sep 17 00:00:00 2001 From: etotmeni Date: Thu, 25 Mar 2021 06:06:40 -0500 Subject: [PATCH 6/6] Added property --- dpctl/_sycl_device.pyx | 56 +++++++++++++++++++++++++++++++++ dpctl/tests/test_sycl_device.py | 32 +++++++++---------- 2 files changed, 72 insertions(+), 16 deletions(-) diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index 5ebd234c7b..7a53566707 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -465,6 +465,62 @@ cdef class SyclDevice(_SyclDevice): ) return max_num_sub_groups + @property + def sub_group_independent_forward_progress(self): + """ Returns true if the device supports independent forward progress of + sub-groups with respect to other sub-groups in the same work-group. + """ + return DPCTLDevice_GetSubGroupIndependentForwardProgress(self._device_ref) + + @property + def preferred_vector_width_char(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthChar(self._device_ref) + + @property + def preferred_vector_width_short(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthShort(self._device_ref) + + @property + def preferred_vector_width_int(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthInt(self._device_ref) + + @property + def preferred_vector_width_long(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthLong(self._device_ref) + + @property + def preferred_vector_width_float(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthFloat(self._device_ref) + + @property + def preferred_vector_width_double(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthDouble(self._device_ref) + + @property + def preferred_vector_width_half(self): + """ Returns the preferred native vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetPreferredVectorWidthHalf(self._device_ref) + @property def vendor_name(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 92b5c43cf8..4d2211c323 100644 --- a/dpctl/tests/test_sycl_device.py +++ b/dpctl/tests/test_sycl_device.py @@ -242,58 +242,58 @@ def check_is_host(device): def check_get_sub_group_independent_forward_progress(device): try: - device.get_sub_group_independent_forward_progress() + device.sub_group_independent_forward_progress except Exception: - pytest.fail("get_sub_group_independent_forward_progress call failed") + pytest.fail("sub_group_independent_forward_progress call failed") def check_get_preferred_vector_width_char(device): try: - device.get_preferred_vector_width_char() + device.preferred_vector_width_char except Exception: - pytest.fail("get_preferred_vector_width_char call failed") + pytest.fail("preferred_vector_width_char call failed") def check_get_preferred_vector_width_short(device): try: - device.get_preferred_vector_width_short() + device.preferred_vector_width_short except Exception: - pytest.fail("get_preferred_vector_width_short call failed") + pytest.fail("preferred_vector_width_short call failed") def check_get_preferred_vector_width_int(device): try: - device.get_preferred_vector_width_int() + device.preferred_vector_width_int except Exception: - pytest.fail("get_preferred_vector_width_int call failed") + pytest.fail("preferred_vector_width_int call failed") def check_get_preferred_vector_width_long(device): try: - device.get_preferred_vector_width_long() + device.preferred_vector_width_long except Exception: - pytest.fail("get_preferred_vector_width_long call failed") + pytest.fail("preferred_vector_width_long call failed") def check_get_preferred_vector_width_float(device): try: - device.get_preferred_vector_width_float() + device.preferred_vector_width_float except Exception: - pytest.fail("get_preferred_vector_width_float call failed") + pytest.fail("preferred_vector_width_float call failed") def check_get_preferred_vector_width_double(device): try: - device.get_preferred_vector_width_double() + device.preferred_vector_width_double except Exception: - pytest.fail("get_preferred_vector_width_double call failed") + pytest.fail("preferred_vector_width_double call failed") def check_get_preferred_vector_width_half(device): try: - device.get_preferred_vector_width_half() + device.preferred_vector_width_half except Exception: - pytest.fail("get_preferred_vector_width_half call failed") + pytest.fail("preferred_vector_width_half call failed") list_of_checks = [