diff --git a/dpctl/_backend.pxd b/dpctl/_backend.pxd index 5f306d70a0..b817ae573b 100644 --- a/dpctl/_backend.pxd +++ b/dpctl/_backend.pxd @@ -186,7 +186,14 @@ cdef extern from "syclinterface/dpctl_sycl_device_interface.h": 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, _aspect_type) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthChar(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthShort(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthInt(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthLong(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthFloat(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthDouble(const DPCTLSyclDeviceRef DRef) + cdef uint32_t DPCTLDevice_GetNativeVectorWidthHalf(const DPCTLSyclDeviceRef DRef) + cdef bool DPCTLDevice_HasAspect(const DPCTLSyclDeviceRef, _aspect_type) cdef uint32_t DPCTLDevice_GetMaxReadImageArgs(const DPCTLSyclDeviceRef DRef) cdef uint32_t DPCTLDevice_GetMaxWriteImageArgs(const DPCTLSyclDeviceRef DRef) cdef size_t DPCTLDevice_GetImage2dMaxWidth(const DPCTLSyclDeviceRef DRef) diff --git a/dpctl/_sycl_device.pyx b/dpctl/_sycl_device.pyx index 76d9e99acc..a2e1b18239 100644 --- a/dpctl/_sycl_device.pyx +++ b/dpctl/_sycl_device.pyx @@ -54,6 +54,13 @@ from ._backend cimport ( # noqa: E211 DPCTLDevice_GetMaxWorkItemSizes3d, DPCTLDevice_GetMaxWriteImageArgs, DPCTLDevice_GetName, + DPCTLDevice_GetNativeVectorWidthChar, + DPCTLDevice_GetNativeVectorWidthDouble, + DPCTLDevice_GetNativeVectorWidthFloat, + DPCTLDevice_GetNativeVectorWidthHalf, + DPCTLDevice_GetNativeVectorWidthInt, + DPCTLDevice_GetNativeVectorWidthLong, + DPCTLDevice_GetNativeVectorWidthShort, DPCTLDevice_GetParentDevice, DPCTLDevice_GetPartitionMaxSubDevices, DPCTLDevice_GetPlatform, @@ -942,6 +949,55 @@ cdef class SyclDevice(_SyclDevice): """ return DPCTLDevice_GetPreferredVectorWidthHalf(self._device_ref) + @property + def native_vector_width_char(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthChar(self._device_ref) + + @property + def native_vector_width_short(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthShort(self._device_ref) + + @property + def native_vector_width_int(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthInt(self._device_ref) + + @property + def native_vector_width_long(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthLong(self._device_ref) + + @property + def native_vector_width_float(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthFloat(self._device_ref) + + @property + def native_vector_width_double(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthDouble(self._device_ref) + + @property + def native_vector_width_half(self): + """ Returns the native ISA vector width size for built-in scalar + types that can be put into vectors. + """ + return DPCTLDevice_GetNativeVectorWidthHalf(self._device_ref) + @property def global_mem_size(self): """ Returns the size of global memory on this device in bytes. diff --git a/dpctl/tests/_device_attributes_checks.py b/dpctl/tests/_device_attributes_checks.py index 49666dfce6..c2d2bbc218 100644 --- a/dpctl/tests/_device_attributes_checks.py +++ b/dpctl/tests/_device_attributes_checks.py @@ -378,6 +378,55 @@ def check_preferred_vector_width_half(device): pytest.fail("preferred_vector_width_half call failed") +def check_native_vector_width_char(device): + try: + device.native_vector_width_char + except Exception: + pytest.fail("native_vector_width_char call failed") + + +def check_native_vector_width_short(device): + try: + device.native_vector_width_short + except Exception: + pytest.fail("native_vector_width_short call failed") + + +def check_native_vector_width_int(device): + try: + device.native_vector_width_int + except Exception: + pytest.fail("native_vector_width_int call failed") + + +def check_native_vector_width_long(device): + try: + device.native_vector_width_long + except Exception: + pytest.fail("native_vector_width_long call failed") + + +def check_native_vector_width_float(device): + try: + device.native_vector_width_float + except Exception: + pytest.fail("native_vector_width_float call failed") + + +def check_native_vector_width_double(device): + try: + device.native_vector_width_double + except Exception: + pytest.fail("native_vector_width_double call failed") + + +def check_native_vector_width_half(device): + try: + device.native_vector_width_half + except Exception: + pytest.fail("native_vector_width_half call failed") + + def check_create_sub_devices_equally(device): try: n = int(device.max_compute_units / 2) @@ -618,6 +667,13 @@ def check_global_mem_cache_line_size(device): check_preferred_vector_width_float, check_preferred_vector_width_double, check_preferred_vector_width_half, + check_native_vector_width_char, + check_native_vector_width_short, + check_native_vector_width_int, + check_native_vector_width_long, + check_native_vector_width_float, + check_native_vector_width_double, + check_native_vector_width_half, check_has_aspect_cpu, check_has_aspect_gpu, check_has_aspect_accelerator, diff --git a/libsyclinterface/include/dpctl_sycl_device_interface.h b/libsyclinterface/include/dpctl_sycl_device_interface.h index bcce8cc592..50c4d1f1f2 100644 --- a/libsyclinterface/include/dpctl_sycl_device_interface.h +++ b/libsyclinterface/include/dpctl_sycl_device_interface.h @@ -575,6 +575,97 @@ DPCTL_API uint32_t DPCTLDevice_GetPreferredVectorWidthHalf( __dpctl_keep const DPCTLSyclDeviceRef DRef); +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * types that can be put into vectors. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t DPCTLDevice_GetNativeVectorWidthChar( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * types that can be put into vectors. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t DPCTLDevice_GetNativeVectorWidthShort( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * types that can be put into vectors. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t +DPCTLDevice_GetNativeVectorWidthInt(__dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * types that can be put into vectors. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t DPCTLDevice_GetNativeVectorWidthLong( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * type. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t DPCTLDevice_GetNativeVectorWidthFloat( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * device.get_info. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * types that can be put into vectors. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t DPCTLDevice_GetNativeVectorWidthDouble( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + +/*! + * @brief Wrapper over + * ``device.get_info``. + * + * @param DRef Opaque pointer to a ``sycl::device`` + * @return Returns the native ISA vector width size for built-in scalar + * types that can be put into vectors. + * @ingroup DeviceInterface + */ +DPCTL_API +uint32_t DPCTLDevice_GetNativeVectorWidthHalf( + __dpctl_keep const DPCTLSyclDeviceRef DRef); + /*! * @brief Wrapper over * device.get_info diff --git a/libsyclinterface/source/dpctl_sycl_device_interface.cpp b/libsyclinterface/source/dpctl_sycl_device_interface.cpp index 02e4239a3f..bcdd46a9fd 100644 --- a/libsyclinterface/source/dpctl_sycl_device_interface.cpp +++ b/libsyclinterface/source/dpctl_sycl_device_interface.cpp @@ -430,116 +430,117 @@ bool DPCTLDevice_GetSubGroupIndependentForwardProgress( return SubGroupProgress; } -uint32_t DPCTLDevice_GetPreferredVectorWidthChar( - __dpctl_keep const DPCTLSyclDeviceRef DRef) +namespace { - size_t vector_width_char = 0; + +template +uint32_t get_uint32_descriptor(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + uint32_t descr_val = 0; auto D = unwrap(DRef); if (D) { try { - vector_width_char = - D->get_info(); + descr_val = D->get_info(); } catch (std::exception const &e) { error_handler(e, __FILE__, __func__, __LINE__); } } - return vector_width_char; + return descr_val; +} + +} // end of anonymous namespace + +uint32_t DPCTLDevice_GetPreferredVectorWidthChar( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor( + DRef); } 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 (std::exception const &e) { - error_handler(e, __FILE__, __func__, __LINE__); - } - } - return vector_width_short; + return get_uint32_descriptor( + DRef); } 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 (std::exception const &e) { - error_handler(e, __FILE__, __func__, __LINE__); - } - } - return vector_width_int; + return get_uint32_descriptor( + DRef); } 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 (std::exception const &e) { - error_handler(e, __FILE__, __func__, __LINE__); - } - } - return vector_width_long; + return get_uint32_descriptor( + DRef); } 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 (std::exception const &e) { - error_handler(e, __FILE__, __func__, __LINE__); - } - } - return vector_width_float; + return get_uint32_descriptor( + DRef); } 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 (std::exception const &e) { - error_handler(e, __FILE__, __func__, __LINE__); - } - } - return vector_width_double; + return get_uint32_descriptor( + DRef); } 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 (std::exception const &e) { - error_handler(e, __FILE__, __func__, __LINE__); - } - } - return vector_width_half; + return get_uint32_descriptor( + DRef); +} + +// +uint32_t +DPCTLDevice_GetNativeVectorWidthChar(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor(DRef); +} + +uint32_t DPCTLDevice_GetNativeVectorWidthShort( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor(DRef); +} + +uint32_t +DPCTLDevice_GetNativeVectorWidthInt(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor(DRef); +} + +uint32_t +DPCTLDevice_GetNativeVectorWidthLong(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor(DRef); +} + +uint32_t DPCTLDevice_GetNativeVectorWidthFloat( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor(DRef); +} + +uint32_t DPCTLDevice_GetNativeVectorWidthDouble( + __dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor( + DRef); +} + +uint32_t +DPCTLDevice_GetNativeVectorWidthHalf(__dpctl_keep const DPCTLSyclDeviceRef DRef) +{ + return get_uint32_descriptor(DRef); } __dpctl_give DPCTLSyclDeviceRef diff --git a/libsyclinterface/tests/test_sycl_device_interface.cpp b/libsyclinterface/tests/test_sycl_device_interface.cpp index 7d202044d1..7e92c8c9de 100644 --- a/libsyclinterface/tests/test_sycl_device_interface.cpp +++ b/libsyclinterface/tests/test_sycl_device_interface.cpp @@ -256,7 +256,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetSubGroupIndependentForwardProgress) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthChar) { - size_t vector_width_char = 0; + uint32_t vector_width_char = 0; EXPECT_NO_FATAL_FAILURE(vector_width_char = DPCTLDevice_GetPreferredVectorWidthChar(DRef)); EXPECT_TRUE(vector_width_char != 0); @@ -264,7 +264,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthChar) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthShort) { - size_t vector_width_short = 0; + uint32_t vector_width_short = 0; EXPECT_NO_FATAL_FAILURE(vector_width_short = DPCTLDevice_GetPreferredVectorWidthShort(DRef)); EXPECT_TRUE(vector_width_short != 0); @@ -272,7 +272,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthShort) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthInt) { - size_t vector_width_int = 0; + uint32_t vector_width_int = 0; EXPECT_NO_FATAL_FAILURE(vector_width_int = DPCTLDevice_GetPreferredVectorWidthInt(DRef)); EXPECT_TRUE(vector_width_int != 0); @@ -280,7 +280,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthInt) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthLong) { - size_t vector_width_long = 0; + uint32_t vector_width_long = 0; EXPECT_NO_FATAL_FAILURE(vector_width_long = DPCTLDevice_GetPreferredVectorWidthLong(DRef)); EXPECT_TRUE(vector_width_long != 0); @@ -288,7 +288,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthLong) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthFloat) { - size_t vector_width_float = 0; + uint32_t vector_width_float = 0; EXPECT_NO_FATAL_FAILURE(vector_width_float = DPCTLDevice_GetPreferredVectorWidthFloat(DRef)); EXPECT_TRUE(vector_width_float != 0); @@ -296,7 +296,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthFloat) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthDouble) { - size_t vector_width_double = 0; + uint32_t vector_width_double = 0; EXPECT_NO_FATAL_FAILURE( vector_width_double = DPCTLDevice_GetPreferredVectorWidthDouble(DRef)); if (DPCTLDevice_HasAspect(DRef, DPCTL_SyclAspectToDPCTLAspectType( @@ -313,7 +313,7 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthDouble) TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthHalf) { - size_t vector_width_half = 0; + uint32_t vector_width_half = 0; EXPECT_NO_FATAL_FAILURE(vector_width_half = DPCTLDevice_GetPreferredVectorWidthHalf(DRef)); if (DPCTLDevice_HasAspect(DRef, DPCTL_SyclAspectToDPCTLAspectType( @@ -326,6 +326,80 @@ TEST_P(TestDPCTLSyclDeviceInterface, ChkGetPreferredVectorWidthHalf) } } +// +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthChar) +{ + uint32_t vector_width_char = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_char = + DPCTLDevice_GetNativeVectorWidthChar(DRef)); + EXPECT_TRUE(vector_width_char != 0); +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthShort) +{ + uint32_t vector_width_short = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_short = + DPCTLDevice_GetNativeVectorWidthShort(DRef)); + EXPECT_TRUE(vector_width_short != 0); +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthInt) +{ + uint32_t vector_width_int = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_int = + DPCTLDevice_GetNativeVectorWidthInt(DRef)); + EXPECT_TRUE(vector_width_int != 0); +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthLong) +{ + uint32_t vector_width_long = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_long = + DPCTLDevice_GetNativeVectorWidthLong(DRef)); + EXPECT_TRUE(vector_width_long != 0); +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthFloat) +{ + uint32_t vector_width_float = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_float = + DPCTLDevice_GetNativeVectorWidthFloat(DRef)); + EXPECT_TRUE(vector_width_float != 0); +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthDouble) +{ + uint32_t vector_width_double = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_double = + DPCTLDevice_GetNativeVectorWidthDouble(DRef)); + if (DPCTLDevice_HasAspect(DRef, DPCTL_SyclAspectToDPCTLAspectType( + DPCTL_StrToAspectType("fp64")))) + { + EXPECT_TRUE(vector_width_double != 0); + } + else { + // FIXME: DPC++ 2023 RT must have a bug, since it returns 1 for + // devices without aspect::fp64 + // EXPECT_TRUE(vector_width_double == 0); + } +} + +TEST_P(TestDPCTLSyclDeviceInterface, ChkGetNativeVectorWidthHalf) +{ + uint32_t vector_width_half = 0; + EXPECT_NO_FATAL_FAILURE(vector_width_half = + DPCTLDevice_GetNativeVectorWidthHalf(DRef)); + if (DPCTLDevice_HasAspect(DRef, DPCTL_SyclAspectToDPCTLAspectType( + DPCTL_StrToAspectType("fp16")))) + { + EXPECT_TRUE(vector_width_half != 0); + } + else { + EXPECT_TRUE(vector_width_half == 0); + } +} +// + TEST_P(TestDPCTLSyclDeviceInterface, ChkGetMaxReadImageArgs) { size_t max_read_image_args = 0;