From ae9c5489c8aaff992a11fe126faa859d3bc8608c Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 12 May 2026 05:26:06 -0700 Subject: [PATCH 01/12] Change _slicing.pxi to .pxd/.pyx format --- dpnp/tensor/_slicing.pxd | 42 ++++++++++++++++++++++ dpnp/tensor/{_slicing.pxi => _slicing.pyx} | 5 +++ 2 files changed, 47 insertions(+) create mode 100644 dpnp/tensor/_slicing.pxd rename dpnp/tensor/{_slicing.pxi => _slicing.pyx} (99%) diff --git a/dpnp/tensor/_slicing.pxd b/dpnp/tensor/_slicing.pxd new file mode 100644 index 00000000000..e556ff57f4a --- /dev/null +++ b/dpnp/tensor/_slicing.pxd @@ -0,0 +1,42 @@ +# ***************************************************************************** +# Copyright (c) 2026, Intel Corporation +# All rights reserved. +# +# Redistribution and use in source and binary forms, with or without +# modification, are permitted provided that the following conditions are met: +# - Redistributions of source code must retain the above copyright notice, +# this list of conditions and the following disclaimer. +# - Redistributions in binary form must reproduce the above copyright notice, +# this list of conditions and the following disclaimer in the documentation +# and/or other materials provided with the distribution. +# - Neither the name of the copyright holder nor the names of its contributors +# may be used to endorse or promote products derived from this software +# without specific prior written permission. +# +# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" +# AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE +# IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE +# ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE +# LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR +# CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF +# SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS +# INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN +# CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) +# ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF +# THE POSSIBILITY OF SUCH DAMAGE. +# ***************************************************************************** + +# distutils: language = c++ +# cython: language_level=3 + +cdef bint _is_buffer(object o) + +cdef Py_ssize_t _slice_len( + Py_ssize_t sl_start, + Py_ssize_t sl_stop, + Py_ssize_t sl_step +) + +cdef bint _is_integral(object x) except * + +cdef bint _is_boolean(object x) except * diff --git a/dpnp/tensor/_slicing.pxi b/dpnp/tensor/_slicing.pyx similarity index 99% rename from dpnp/tensor/_slicing.pxi rename to dpnp/tensor/_slicing.pyx index 2f22894c4b1..2e785951bc6 100644 --- a/dpnp/tensor/_slicing.pxi +++ b/dpnp/tensor/_slicing.pyx @@ -26,11 +26,16 @@ # THE POSSIBILITY OF SUCH DAMAGE. # ***************************************************************************** +# distutils: language = c++ +# cython: language_level=3 + import numbers from operator import index from cpython.buffer cimport PyObject_CheckBuffer from numpy import ndarray +from ._usmarray cimport usm_ndarray + cdef bint _is_buffer(object o): return PyObject_CheckBuffer(o) From e08ccbb3b7211ece893824ea98eff9c04d810252 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 12 May 2026 05:27:41 -0700 Subject: [PATCH 02/12] Change _stride_utils.pxi to .pxd/.pyx format --- dpnp/tensor/_stride_utils.pxd | 58 +++++++++++++++++++ .../{_stride_utils.pxi => _stride_utils.pyx} | 2 +- 2 files changed, 59 insertions(+), 1 deletion(-) create mode 100644 dpnp/tensor/_stride_utils.pxd rename dpnp/tensor/{_stride_utils.pxi => _stride_utils.pyx} (99%) diff --git a/dpnp/tensor/_stride_utils.pxd b/dpnp/tensor/_stride_utils.pxd new file mode 100644 index 00000000000..d936f0d864d --- /dev/null +++ b/dpnp/tensor/_stride_utils.pxd @@ -0,0 +1,58 @@ +# ***************************************************************************** +# Copyright (c) 2026, Intel Corporation +# All rights reserved. +# +# Redistribution and use in source and binary forms, with or without +# modification, are permitted provided that the following conditions are met: +# - Redistributions of source code must retain the above copyright notice, +# this list of conditions and the following disclaimer. +# - Redistributions in binary form must reproduce the above copyright notice, +# this list of conditions and the following disclaimer in the documentation +# and/or other materials provided with the distribution. +# - Neither the name of the copyright holder nor the names of its contributors +# may be used to endorse or promote products derived from this software +# without specific prior written permission. +# +# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" +# AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE +# IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE +# ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE +# LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR +# CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF +# SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS +# INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN +# CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) +# ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF +# THE POSSIBILITY OF SUCH DAMAGE. +# ***************************************************************************** + +# distutils: language = c++ +# cython: language_level=3 + +cdef int ERROR_MALLOC +cdef int ERROR_INTERNAL +cdef int ERROR_INCORRECT_ORDER +cdef int ERROR_UNEXPECTED_STRIDES + +cdef int USM_ARRAY_C_CONTIGUOUS +cdef int USM_ARRAY_F_CONTIGUOUS +cdef int USM_ARRAY_WRITABLE + +cdef Py_ssize_t shape_to_elem_count(int nd, Py_ssize_t *shape_arr) + +cdef int _from_input_shape_strides( + int nd, object shape, object strides, int itemsize, char order, + Py_ssize_t **shape_ptr, Py_ssize_t **strides_ptr, + Py_ssize_t *nelems, Py_ssize_t *min_disp, Py_ssize_t *max_disp, + int *contig +) + +cdef object _make_int_tuple(int nd, const Py_ssize_t *ary) + +cdef object _make_reversed_int_tuple(int nd, const Py_ssize_t *ary) + +cdef object _c_contig_strides(int nd, Py_ssize_t *shape) + +cdef object _f_contig_strides(int nd, Py_ssize_t *shape) + +cdef object _swap_last_two(tuple t) diff --git a/dpnp/tensor/_stride_utils.pxi b/dpnp/tensor/_stride_utils.pyx similarity index 99% rename from dpnp/tensor/_stride_utils.pxi rename to dpnp/tensor/_stride_utils.pyx index 3caf8dd8fd1..ff2cd72a1ba 100644 --- a/dpnp/tensor/_stride_utils.pxi +++ b/dpnp/tensor/_stride_utils.pyx @@ -29,7 +29,7 @@ # distutils: language = c++ # cython: language_level=3 -from cpython.mem cimport PyMem_Malloc +from cpython.mem cimport PyMem_Free, PyMem_Malloc from cpython.ref cimport Py_INCREF from cpython.tuple cimport PyTuple_New, PyTuple_SetItem From ec340fd82545a8a33fb122b6d4470bd4f3fa14cd Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 12 May 2026 05:28:37 -0700 Subject: [PATCH 03/12] Change _types.pxi to .pxd/.pyx format --- dpnp/tensor/_types.pxd | 58 ++++++++++++++++++++++++++ dpnp/tensor/{_types.pxi => _types.pyx} | 5 +++ 2 files changed, 63 insertions(+) create mode 100644 dpnp/tensor/_types.pxd rename dpnp/tensor/{_types.pxi => _types.pyx} (98%) diff --git a/dpnp/tensor/_types.pxd b/dpnp/tensor/_types.pxd new file mode 100644 index 00000000000..0e271bf5a7b --- /dev/null +++ b/dpnp/tensor/_types.pxd @@ -0,0 +1,58 @@ +# ***************************************************************************** +# Copyright (c) 2026, Intel Corporation +# All rights reserved. +# +# Redistribution and use in source and binary forms, with or without +# modification, are permitted provided that the following conditions are met: +# - Redistributions of source code must retain the above copyright notice, +# this list of conditions and the following disclaimer. +# - Redistributions in binary form must reproduce the above copyright notice, +# this list of conditions and the following disclaimer in the documentation +# and/or other materials provided with the distribution. +# - Neither the name of the copyright holder nor the names of its contributors +# may be used to endorse or promote products derived from this software +# without specific prior written permission. +# +# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" +# AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE +# IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE +# ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE +# LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR +# CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF +# SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS +# INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN +# CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) +# ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF +# THE POSSIBILITY OF SUCH DAMAGE. +# ***************************************************************************** + +# distutils: language = c++ +# cython: language_level=3 + +cdef int UAR_BOOL +cdef int UAR_BYTE +cdef int UAR_UBYTE +cdef int UAR_SHORT +cdef int UAR_USHORT +cdef int UAR_INT +cdef int UAR_UINT +cdef int UAR_LONG +cdef int UAR_ULONG +cdef int UAR_LONGLONG +cdef int UAR_ULONGLONG +cdef int UAR_FLOAT +cdef int UAR_DOUBLE +cdef int UAR_CFLOAT +cdef int UAR_CDOUBLE +cdef int UAR_TYPE_SENTINEL +cdef int UAR_HALF + +cdef int type_bytesize(int typenum) + +cdef str _make_typestr(int typenum) + +cdef int typenum_from_format(str s) + +cdef int descr_to_typenum(object dtype) + +cdef int dtype_to_typenum(dtype) diff --git a/dpnp/tensor/_types.pxi b/dpnp/tensor/_types.pyx similarity index 98% rename from dpnp/tensor/_types.pxi rename to dpnp/tensor/_types.pyx index 090750658f4..f25ad9908b4 100644 --- a/dpnp/tensor/_types.pxi +++ b/dpnp/tensor/_types.pyx @@ -26,6 +26,11 @@ # THE POSSIBILITY OF SUCH DAMAGE. # ***************************************************************************** +# distutils: language = c++ +# cython: language_level=3 + +import numpy as np + # these typenum values are aligned to values in NumPy cdef: int UAR_BOOL = 0 # pragma: no cover From 2ad80a8d0aead2d7e466210a14c9d35381ccec2c Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 12 May 2026 05:29:51 -0700 Subject: [PATCH 04/12] Update _usmarray.pyx imports --- dpnp/tensor/_usmarray.pyx | 53 ++++++++++++++++++++++++++++++++++++--- 1 file changed, 49 insertions(+), 4 deletions(-) diff --git a/dpnp/tensor/_usmarray.pyx b/dpnp/tensor/_usmarray.pyx index c696056d53c..08467417dff 100644 --- a/dpnp/tensor/_usmarray.pyx +++ b/dpnp/tensor/_usmarray.pyx @@ -46,7 +46,6 @@ from ._print import usm_ndarray_repr, usm_ndarray_str cimport dpctl as c_dpctl cimport dpctl.memory as c_dpmem from cpython.mem cimport PyMem_Free -from cpython.tuple cimport PyTuple_New, PyTuple_SetItem from . cimport _dlpack as c_dlpack @@ -56,9 +55,55 @@ from . import _flags from ._dlpack import get_build_dlpack_version from ._tensor_impl import default_device_fp_type -include "_stride_utils.pxi" -include "_types.pxi" -include "_slicing.pxi" +from ._slicing cimport * +from ._stride_utils cimport ( + ERROR_INCORRECT_ORDER, + ERROR_INTERNAL, + ERROR_MALLOC, + ERROR_UNEXPECTED_STRIDES, + _c_contig_strides, + _f_contig_strides, + _from_input_shape_strides, + _make_int_tuple, + _make_reversed_int_tuple, + _swap_last_two, + shape_to_elem_count, +) +from ._types cimport ( + _make_typestr, + descr_to_typenum, + dtype_to_typenum, + type_bytesize, + typenum_from_format, +) + +from ._slicing import _basic_slice_meta + + +# Local storage for `cdef public api` constants +# declared in _usmarray.pxd +cdef int USM_ARRAY_C_CONTIGUOUS = 1 +cdef int USM_ARRAY_F_CONTIGUOUS = 2 +cdef int USM_ARRAY_WRITABLE = 4 + +cdef: + int UAR_BOOL = 0 + int UAR_BYTE = 1 + int UAR_UBYTE = 2 + int UAR_SHORT = 3 + int UAR_USHORT = 4 + int UAR_INT = 5 + int UAR_UINT = 6 + int UAR_LONG = 7 + int UAR_ULONG = 8 + int UAR_LONGLONG = 9 + int UAR_ULONGLONG = 10 + int UAR_FLOAT = 11 + int UAR_DOUBLE = 12 + int UAR_CFLOAT = 14 + int UAR_CDOUBLE = 15 + int UAR_TYPE_SENTINEL = 17 + int UAR_HALF = 23 class DLDeviceType(IntEnum): From bfb3bc28b00daf6d4f8299ce4c978f8da781642f Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 12 May 2026 05:41:58 -0700 Subject: [PATCH 05/12] Update changelog --- CHANGELOG.md | 2 ++ 1 file changed, 2 insertions(+) diff --git a/CHANGELOG.md b/CHANGELOG.md index 752bf2ad4b3..791c631b3d8 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -10,6 +10,8 @@ and this project adheres to [Semantic Versioning](https://semver.org/spec/v2.0.0 ### Changed +* Replaced `.pxi` includes in `dpnp.tensor` with modular `.pxd`/`.pyx` Cython imports[#2913](https://github.com/IntelPython/dpnp/pull/2913) + ### Deprecated ### Removed From ef9c4f10d3f8a0d216ec545a231469696e60e297 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 13 May 2026 13:47:58 -0700 Subject: [PATCH 06/12] qwe --- CHANGELOG.md | 2 +- dpnp/tensor/_usmarray.pyx | 10 +++++-- .../kernels/elementwise_functions/expm1.hpp | 4 +++ dpnp/tests/tensor/elementwise/test_expm1.py | 28 +++++++++++++++++++ 4 files changed, 40 insertions(+), 4 deletions(-) diff --git a/CHANGELOG.md b/CHANGELOG.md index a63e9bb05f4..55ef3f92d4d 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -11,7 +11,7 @@ and this project adheres to [Semantic Versioning](https://semver.org/spec/v2.0.0 ### Changed * Changed `dpnp.meshgrid` and `dpnp.tensor.meshgrid` to return a tuple instead of a list, aligning with NumPy 2.5+ behavior and 2025.12 version of the Python array API standard [#2854](https://github.com/IntelPython/dpnp/pull/2854) -* Replaced `.pxi` includes in `dpnp.tensor` with modular `.pxd`/`.pyx` Cython imports[#2913](https://github.com/IntelPython/dpnp/pull/2913) +* Replaced `.pxi` includes in `dpnp.tensor` with modular `.pxd`/`.pyx` Cython imports [#2913](https://github.com/IntelPython/dpnp/pull/2913) ### Deprecated diff --git a/dpnp/tensor/_usmarray.pyx b/dpnp/tensor/_usmarray.pyx index 08467417dff..eb1777fcc59 100644 --- a/dpnp/tensor/_usmarray.pyx +++ b/dpnp/tensor/_usmarray.pyx @@ -55,7 +55,13 @@ from . import _flags from ._dlpack import get_build_dlpack_version from ._tensor_impl import default_device_fp_type -from ._slicing cimport * +from ._slicing cimport ( + _is_buffer, + _is_boolean, + _is_integral, + _slice_len, +) +from ._slicing import _basic_slice_meta from ._stride_utils cimport ( ERROR_INCORRECT_ORDER, ERROR_INTERNAL, @@ -77,8 +83,6 @@ from ._types cimport ( typenum_from_format, ) -from ._slicing import _basic_slice_meta - # Local storage for `cdef public api` constants # declared in _usmarray.pxd diff --git a/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp b/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp index e9e2a704cf0..735cb1fada5 100644 --- a/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp +++ b/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp @@ -121,6 +121,10 @@ struct Expm1Functor } } + if (x == realT(0) && y == realT(0)) { + return resT{realT(0), y}; + } + // x, y finite numbers const realT cosY_val = sycl::cos(y); const realT sinY_val = (y == 0) ? y : sycl::sin(y); diff --git a/dpnp/tests/tensor/elementwise/test_expm1.py b/dpnp/tests/tensor/elementwise/test_expm1.py index bb665c42456..0e1a8205dab 100644 --- a/dpnp/tests/tensor/elementwise/test_expm1.py +++ b/dpnp/tests/tensor/elementwise/test_expm1.py @@ -185,3 +185,31 @@ def test_expm1_special_cases(): tol = dpt.finfo(X.dtype).resolution with np.errstate(invalid="ignore"): assert_allclose(dpt.asnumpy(dpt.expm1(X)), res, atol=tol, rtol=tol) + + +@pytest.mark.parametrize("dtype", ["c8", "c16"]) +def test_expm1_zero_special_cases(dtype): + q = get_queue_or_skip() + skip_if_dtype_not_supported(dtype, q) + + x = [ + complex(+0.0, +0.0), + complex(-0.0, +0.0), + complex(+0.0, -0.0), + complex(-0.0, -0.0), + ] + expected = [ + complex(0.0, 0.0), + complex(0.0, 0.0), + complex(0.0, -0.0), + complex(0.0, -0.0), + ] + + xf = dpt.asarray(x, dtype=dtype, sycl_queue=q) + Y = dpt.asnumpy(dpt.expm1(xf)) + + for i in range(len(x)): + assert Y[i].real == expected[i].real + assert Y[i].imag == expected[i].imag + assert not np.signbit(Y[i].real) + assert np.signbit(Y[i].imag) == np.signbit(expected[i].imag) From 6debec83289b46cc22cbd0bdd67d857513ee8520 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Fri, 15 May 2026 04:51:07 -0700 Subject: [PATCH 07/12] Move usm_ndarray constants to usm_ndarray_constants.h --- dpnp/backend/include/dpnp4pybind11.hpp | 5 +- dpnp/tensor/_stride_utils.pxd | 9 +-- dpnp/tensor/_stride_utils.pyx | 4 -- dpnp/tensor/_types.pxd | 35 ++++++------ dpnp/tensor/_types.pyx | 19 ------- dpnp/tensor/_usmarray.pxd | 41 +++++++------- dpnp/tensor/_usmarray.pyx | 30 +--------- dpnp/tensor/include/usm_ndarray_constants.h | 62 +++++++++++++++++++++ 8 files changed, 112 insertions(+), 93 deletions(-) create mode 100644 dpnp/tensor/include/usm_ndarray_constants.h diff --git a/dpnp/backend/include/dpnp4pybind11.hpp b/dpnp/backend/include/dpnp4pybind11.hpp index 3150d63146f..4c46a53297e 100644 --- a/dpnp/backend/include/dpnp4pybind11.hpp +++ b/dpnp/backend/include/dpnp4pybind11.hpp @@ -32,9 +32,10 @@ #include "dpctl4pybind11.hpp" // Include generated Cython headers for usm_ndarray -// (struct definition and constants only) #include "dpnp/tensor/_usmarray.h" #include "dpnp/tensor/_usmarray_api.h" +// Include usm_ndarray constants (flags, type numbers) +#include "../../tensor/include/usm_ndarray_constants.h" #include #include @@ -135,7 +136,7 @@ class dpnp_capi this->PyUSMArrayType_ = &PyUSMArrayType; - // constants + // constants from usm_ndarray_constants.h this->USM_ARRAY_C_CONTIGUOUS_ = USM_ARRAY_C_CONTIGUOUS; this->USM_ARRAY_F_CONTIGUOUS_ = USM_ARRAY_F_CONTIGUOUS; this->USM_ARRAY_WRITABLE_ = USM_ARRAY_WRITABLE; diff --git a/dpnp/tensor/_stride_utils.pxd b/dpnp/tensor/_stride_utils.pxd index d936f0d864d..b4f5eac4cd1 100644 --- a/dpnp/tensor/_stride_utils.pxd +++ b/dpnp/tensor/_stride_utils.pxd @@ -29,15 +29,16 @@ # distutils: language = c++ # cython: language_level=3 +cdef extern from "usm_ndarray_constants.h": + int USM_ARRAY_C_CONTIGUOUS + int USM_ARRAY_F_CONTIGUOUS + int USM_ARRAY_WRITABLE + cdef int ERROR_MALLOC cdef int ERROR_INTERNAL cdef int ERROR_INCORRECT_ORDER cdef int ERROR_UNEXPECTED_STRIDES -cdef int USM_ARRAY_C_CONTIGUOUS -cdef int USM_ARRAY_F_CONTIGUOUS -cdef int USM_ARRAY_WRITABLE - cdef Py_ssize_t shape_to_elem_count(int nd, Py_ssize_t *shape_arr) cdef int _from_input_shape_strides( diff --git a/dpnp/tensor/_stride_utils.pyx b/dpnp/tensor/_stride_utils.pyx index ff2cd72a1ba..7d770e30e70 100644 --- a/dpnp/tensor/_stride_utils.pyx +++ b/dpnp/tensor/_stride_utils.pyx @@ -39,10 +39,6 @@ cdef int ERROR_INTERNAL = -1 cdef int ERROR_INCORRECT_ORDER = 2 cdef int ERROR_UNEXPECTED_STRIDES = 3 -cdef int USM_ARRAY_C_CONTIGUOUS = 1 -cdef int USM_ARRAY_F_CONTIGUOUS = 2 -cdef int USM_ARRAY_WRITABLE = 4 - cdef Py_ssize_t shape_to_elem_count(int nd, Py_ssize_t *shape_arr): """ diff --git a/dpnp/tensor/_types.pxd b/dpnp/tensor/_types.pxd index 0e271bf5a7b..5230688d4ea 100644 --- a/dpnp/tensor/_types.pxd +++ b/dpnp/tensor/_types.pxd @@ -29,23 +29,24 @@ # distutils: language = c++ # cython: language_level=3 -cdef int UAR_BOOL -cdef int UAR_BYTE -cdef int UAR_UBYTE -cdef int UAR_SHORT -cdef int UAR_USHORT -cdef int UAR_INT -cdef int UAR_UINT -cdef int UAR_LONG -cdef int UAR_ULONG -cdef int UAR_LONGLONG -cdef int UAR_ULONGLONG -cdef int UAR_FLOAT -cdef int UAR_DOUBLE -cdef int UAR_CFLOAT -cdef int UAR_CDOUBLE -cdef int UAR_TYPE_SENTINEL -cdef int UAR_HALF +cdef extern from "usm_ndarray_constants.h": + int UAR_BOOL + int UAR_BYTE + int UAR_UBYTE + int UAR_SHORT + int UAR_USHORT + int UAR_INT + int UAR_UINT + int UAR_LONG + int UAR_ULONG + int UAR_LONGLONG + int UAR_ULONGLONG + int UAR_FLOAT + int UAR_DOUBLE + int UAR_CFLOAT + int UAR_CDOUBLE + int UAR_TYPE_SENTINEL + int UAR_HALF cdef int type_bytesize(int typenum) diff --git a/dpnp/tensor/_types.pyx b/dpnp/tensor/_types.pyx index f25ad9908b4..34cea76cfd0 100644 --- a/dpnp/tensor/_types.pyx +++ b/dpnp/tensor/_types.pyx @@ -31,25 +31,6 @@ import numpy as np -# these typenum values are aligned to values in NumPy -cdef: - int UAR_BOOL = 0 # pragma: no cover - int UAR_BYTE = 1 # pragma: no cover - int UAR_UBYTE = 2 # pragma: no cover - int UAR_SHORT = 3 # pragma: no cover - int UAR_USHORT = 4 # pragma: no cover - int UAR_INT = 5 # pragma: no cover - int UAR_UINT = 6 # pragma: no cover - int UAR_LONG = 7 # pragma: no cover - int UAR_ULONG = 8 # pragma: no cover - int UAR_LONGLONG = 9 # pragma: no cover - int UAR_ULONGLONG = 10 # pragma: no cover - int UAR_FLOAT = 11 # pragma: no cover - int UAR_DOUBLE = 12 # pragma: no cover - int UAR_CFLOAT = 14 # pragma: no cover - int UAR_CDOUBLE = 15 # pragma: no cover - int UAR_TYPE_SENTINEL = 17 # pragma: no cover - int UAR_HALF = 23 # pragma: no cover cdef int type_bytesize(int typenum): """ diff --git a/dpnp/tensor/_usmarray.pxd b/dpnp/tensor/_usmarray.pxd index ccb8f4c796b..516a31cba71 100644 --- a/dpnp/tensor/_usmarray.pxd +++ b/dpnp/tensor/_usmarray.pxd @@ -32,27 +32,28 @@ cimport dpctl -cdef public api int USM_ARRAY_C_CONTIGUOUS -cdef public api int USM_ARRAY_F_CONTIGUOUS -cdef public api int USM_ARRAY_WRITABLE +cdef extern from "usm_ndarray_constants.h": + int USM_ARRAY_C_CONTIGUOUS + int USM_ARRAY_F_CONTIGUOUS + int USM_ARRAY_WRITABLE -cdef public api int UAR_BOOL -cdef public api int UAR_BYTE -cdef public api int UAR_UBYTE -cdef public api int UAR_SHORT -cdef public api int UAR_USHORT -cdef public api int UAR_INT -cdef public api int UAR_UINT -cdef public api int UAR_LONG -cdef public api int UAR_ULONG -cdef public api int UAR_LONGLONG -cdef public api int UAR_ULONGLONG -cdef public api int UAR_FLOAT -cdef public api int UAR_DOUBLE -cdef public api int UAR_CFLOAT -cdef public api int UAR_CDOUBLE -cdef public api int UAR_TYPE_SENTINEL -cdef public api int UAR_HALF + int UAR_BOOL + int UAR_BYTE + int UAR_UBYTE + int UAR_SHORT + int UAR_USHORT + int UAR_INT + int UAR_UINT + int UAR_LONG + int UAR_ULONG + int UAR_LONGLONG + int UAR_ULONGLONG + int UAR_FLOAT + int UAR_DOUBLE + int UAR_CFLOAT + int UAR_CDOUBLE + int UAR_TYPE_SENTINEL + int UAR_HALF cdef api class usm_ndarray [object PyUSMArrayObject, type PyUSMArrayType]: diff --git a/dpnp/tensor/_usmarray.pyx b/dpnp/tensor/_usmarray.pyx index eb1777fcc59..01ff970c5e1 100644 --- a/dpnp/tensor/_usmarray.pyx +++ b/dpnp/tensor/_usmarray.pyx @@ -56,12 +56,14 @@ from ._dlpack import get_build_dlpack_version from ._tensor_impl import default_device_fp_type from ._slicing cimport ( - _is_buffer, _is_boolean, + _is_buffer, _is_integral, _slice_len, ) + from ._slicing import _basic_slice_meta + from ._stride_utils cimport ( ERROR_INCORRECT_ORDER, ERROR_INTERNAL, @@ -84,32 +86,6 @@ from ._types cimport ( ) -# Local storage for `cdef public api` constants -# declared in _usmarray.pxd -cdef int USM_ARRAY_C_CONTIGUOUS = 1 -cdef int USM_ARRAY_F_CONTIGUOUS = 2 -cdef int USM_ARRAY_WRITABLE = 4 - -cdef: - int UAR_BOOL = 0 - int UAR_BYTE = 1 - int UAR_UBYTE = 2 - int UAR_SHORT = 3 - int UAR_USHORT = 4 - int UAR_INT = 5 - int UAR_UINT = 6 - int UAR_LONG = 7 - int UAR_ULONG = 8 - int UAR_LONGLONG = 9 - int UAR_ULONGLONG = 10 - int UAR_FLOAT = 11 - int UAR_DOUBLE = 12 - int UAR_CFLOAT = 14 - int UAR_CDOUBLE = 15 - int UAR_TYPE_SENTINEL = 17 - int UAR_HALF = 23 - - class DLDeviceType(IntEnum): """ An :class:`enum.IntEnum` for the types of DLDevices supported by the DLPack diff --git a/dpnp/tensor/include/usm_ndarray_constants.h b/dpnp/tensor/include/usm_ndarray_constants.h new file mode 100644 index 00000000000..dbe1f9da8b8 --- /dev/null +++ b/dpnp/tensor/include/usm_ndarray_constants.h @@ -0,0 +1,62 @@ +/* ***************************************************************************** + * Copyright (c) 2026, Intel Corporation + * All rights reserved. + * + * Redistribution and use in source and binary forms, with or without + * modification, are permitted provided that the following conditions are met: + * - Redistributions of source code must retain the above copyright notice, + * this list of conditions and the following disclaimer. + * - Redistributions in binary form must reproduce the above copyright notice, + * this list of conditions and the following disclaimer in the documentation + * and/or other materials provided with the distribution. + * - Neither the name of the copyright holder nor the names of its contributors + * may be used to endorse or promote products derived from this software + * without specific prior written permission. + * + * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" + * AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE + * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE + * ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE + * LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR + * CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF + * SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS + * INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN + * CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) + * ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF + * THE POSSIBILITY OF SUCH DAMAGE. + * ****************************************************************************/ + +#ifndef DPNP_TENSOR_USM_NDARRAY_CONSTANTS_H +#define DPNP_TENSOR_USM_NDARRAY_CONSTANTS_H + +/* Array contiguity flags */ +enum +{ + USM_ARRAY_C_CONTIGUOUS = 1, + USM_ARRAY_F_CONTIGUOUS = 2, + USM_ARRAY_WRITABLE = 4 +}; + +/* These typenum values are aligned to values in NumPy */ +enum +{ + UAR_BOOL = 0, + UAR_BYTE = 1, + UAR_UBYTE = 2, + UAR_SHORT = 3, + UAR_USHORT = 4, + UAR_INT = 5, + UAR_UINT = 6, + UAR_LONG = 7, + UAR_ULONG = 8, + UAR_LONGLONG = 9, + UAR_ULONGLONG = 10, + UAR_FLOAT = 11, + UAR_DOUBLE = 12, + UAR_CFLOAT = 14, + UAR_CDOUBLE = 15, + UAR_TYPE_SENTINEL = 17, + UAR_HALF = 23 +}; + +#endif /* DPNP_TENSOR_USM_NDARRAY_CONSTANTS_H */ From 4c1155c80fc8f43dd4fabf4164a6c619de9de785 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 19 May 2026 04:28:31 -0700 Subject: [PATCH 08/12] Remove accidentally added changes for expm1 --- .../kernels/elementwise_functions/expm1.hpp | 4 --- dpnp/tests/tensor/elementwise/test_expm1.py | 28 ------------------- 2 files changed, 32 deletions(-) diff --git a/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp b/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp index 735cb1fada5..e9e2a704cf0 100644 --- a/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp +++ b/dpnp/tensor/libtensor/include/kernels/elementwise_functions/expm1.hpp @@ -121,10 +121,6 @@ struct Expm1Functor } } - if (x == realT(0) && y == realT(0)) { - return resT{realT(0), y}; - } - // x, y finite numbers const realT cosY_val = sycl::cos(y); const realT sinY_val = (y == 0) ? y : sycl::sin(y); diff --git a/dpnp/tests/tensor/elementwise/test_expm1.py b/dpnp/tests/tensor/elementwise/test_expm1.py index 0e1a8205dab..bb665c42456 100644 --- a/dpnp/tests/tensor/elementwise/test_expm1.py +++ b/dpnp/tests/tensor/elementwise/test_expm1.py @@ -185,31 +185,3 @@ def test_expm1_special_cases(): tol = dpt.finfo(X.dtype).resolution with np.errstate(invalid="ignore"): assert_allclose(dpt.asnumpy(dpt.expm1(X)), res, atol=tol, rtol=tol) - - -@pytest.mark.parametrize("dtype", ["c8", "c16"]) -def test_expm1_zero_special_cases(dtype): - q = get_queue_or_skip() - skip_if_dtype_not_supported(dtype, q) - - x = [ - complex(+0.0, +0.0), - complex(-0.0, +0.0), - complex(+0.0, -0.0), - complex(-0.0, -0.0), - ] - expected = [ - complex(0.0, 0.0), - complex(0.0, 0.0), - complex(0.0, -0.0), - complex(0.0, -0.0), - ] - - xf = dpt.asarray(x, dtype=dtype, sycl_queue=q) - Y = dpt.asnumpy(dpt.expm1(xf)) - - for i in range(len(x)): - assert Y[i].real == expected[i].real - assert Y[i].imag == expected[i].imag - assert not np.signbit(Y[i].real) - assert np.signbit(Y[i].imag) == np.signbit(expected[i].imag) From 0240b5f33b993da756db8d2fe544aa288c61c44b Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 27 May 2026 05:01:44 -0700 Subject: [PATCH 09/12] Initialize cdef public api constants from usm_ndarray_constants.h --- dpnp/backend/include/dpnp4pybind11.hpp | 58 +++++++++---------- dpnp/tensor/_stride_utils.pxd | 10 +++- dpnp/tensor/_stride_utils.pyx | 4 ++ dpnp/tensor/_types.pxd | 52 +++++++++++------ dpnp/tensor/_types.pyx | 19 +++++++ dpnp/tensor/_usmarray.pxd | 62 ++++++++++++++------- dpnp/tensor/_usmarray.pyx | 24 ++++++++ dpnp/tensor/include/usm_ndarray_constants.h | 40 ++++++------- 8 files changed, 180 insertions(+), 89 deletions(-) diff --git a/dpnp/backend/include/dpnp4pybind11.hpp b/dpnp/backend/include/dpnp4pybind11.hpp index 986392409be..a64a5074beb 100644 --- a/dpnp/backend/include/dpnp4pybind11.hpp +++ b/dpnp/backend/include/dpnp4pybind11.hpp @@ -193,46 +193,46 @@ class dpnp_capi this->UsmNDArray_MakeFromPtr_ = UsmNDArray_MakeFromPtr; // constants from usm_ndarray_constants.h - this->USM_ARRAY_C_CONTIGUOUS_ = USM_ARRAY_C_CONTIGUOUS; - this->USM_ARRAY_F_CONTIGUOUS_ = USM_ARRAY_F_CONTIGUOUS; - this->USM_ARRAY_WRITABLE_ = USM_ARRAY_WRITABLE; - this->UAR_BOOL_ = UAR_BOOL; - this->UAR_BYTE_ = UAR_BYTE; - this->UAR_UBYTE_ = UAR_UBYTE; - this->UAR_SHORT_ = UAR_SHORT; - this->UAR_USHORT_ = UAR_USHORT; - this->UAR_INT_ = UAR_INT; - this->UAR_UINT_ = UAR_UINT; - this->UAR_LONG_ = UAR_LONG; - this->UAR_ULONG_ = UAR_ULONG; - this->UAR_LONGLONG_ = UAR_LONGLONG; - this->UAR_ULONGLONG_ = UAR_ULONGLONG; - this->UAR_FLOAT_ = UAR_FLOAT; - this->UAR_DOUBLE_ = UAR_DOUBLE; - this->UAR_CFLOAT_ = UAR_CFLOAT; - this->UAR_CDOUBLE_ = UAR_CDOUBLE; - this->UAR_TYPE_SENTINEL_ = UAR_TYPE_SENTINEL; - this->UAR_HALF_ = UAR_HALF; + this->USM_ARRAY_C_CONTIGUOUS_ = USM_ARRAY_C_CONTIGUOUS_VALUE; + this->USM_ARRAY_F_CONTIGUOUS_ = USM_ARRAY_F_CONTIGUOUS_VALUE; + this->USM_ARRAY_WRITABLE_ = USM_ARRAY_WRITABLE_VALUE; + this->UAR_BOOL_ = UAR_BOOL_VALUE; + this->UAR_BYTE_ = UAR_BYTE_VALUE; + this->UAR_UBYTE_ = UAR_UBYTE_VALUE; + this->UAR_SHORT_ = UAR_SHORT_VALUE; + this->UAR_USHORT_ = UAR_USHORT_VALUE; + this->UAR_INT_ = UAR_INT_VALUE; + this->UAR_UINT_ = UAR_UINT_VALUE; + this->UAR_LONG_ = UAR_LONG_VALUE; + this->UAR_ULONG_ = UAR_ULONG_VALUE; + this->UAR_LONGLONG_ = UAR_LONGLONG_VALUE; + this->UAR_ULONGLONG_ = UAR_ULONGLONG_VALUE; + this->UAR_FLOAT_ = UAR_FLOAT_VALUE; + this->UAR_DOUBLE_ = UAR_DOUBLE_VALUE; + this->UAR_CFLOAT_ = UAR_CFLOAT_VALUE; + this->UAR_CDOUBLE_ = UAR_CDOUBLE_VALUE; + this->UAR_TYPE_SENTINEL_ = UAR_TYPE_SENTINEL_VALUE; + this->UAR_HALF_ = UAR_HALF_VALUE; // deduced disjoint types - this->UAR_INT8_ = UAR_BYTE; - this->UAR_UINT8_ = UAR_UBYTE; - this->UAR_INT16_ = UAR_SHORT; - this->UAR_UINT16_ = UAR_USHORT; + this->UAR_INT8_ = UAR_BYTE_VALUE; + this->UAR_UINT8_ = UAR_UBYTE_VALUE; + this->UAR_INT16_ = UAR_SHORT_VALUE; + this->UAR_UINT16_ = UAR_USHORT_VALUE; this->UAR_INT32_ = platform_typeid_lookup( - UAR_LONG, UAR_INT, UAR_SHORT); + UAR_LONG_VALUE, UAR_INT_VALUE, UAR_SHORT_VALUE); this->UAR_UINT32_ = platform_typeid_lookup(UAR_ULONG, UAR_UINT, - UAR_USHORT); + unsigned short>( + UAR_ULONG_VALUE, UAR_UINT_VALUE, UAR_USHORT_VALUE); this->UAR_INT64_ = platform_typeid_lookup( - UAR_LONG, UAR_LONGLONG, UAR_INT); + UAR_LONG_VALUE, UAR_LONGLONG_VALUE, UAR_INT_VALUE); this->UAR_UINT64_ = platform_typeid_lookup( - UAR_ULONG, UAR_ULONGLONG, UAR_UINT); + UAR_ULONG_VALUE, UAR_ULONGLONG_VALUE, UAR_UINT_VALUE); py::object py_default_usm_memory = ::dpctl::detail::dpctl_capi::get().default_usm_memory_pyobj(); diff --git a/dpnp/tensor/_stride_utils.pxd b/dpnp/tensor/_stride_utils.pxd index b4f5eac4cd1..9aead47d76c 100644 --- a/dpnp/tensor/_stride_utils.pxd +++ b/dpnp/tensor/_stride_utils.pxd @@ -30,9 +30,13 @@ # cython: language_level=3 cdef extern from "usm_ndarray_constants.h": - int USM_ARRAY_C_CONTIGUOUS - int USM_ARRAY_F_CONTIGUOUS - int USM_ARRAY_WRITABLE + int USM_ARRAY_C_CONTIGUOUS_VALUE + int USM_ARRAY_F_CONTIGUOUS_VALUE + int USM_ARRAY_WRITABLE_VALUE + +cdef int USM_ARRAY_C_CONTIGUOUS +cdef int USM_ARRAY_F_CONTIGUOUS +cdef int USM_ARRAY_WRITABLE cdef int ERROR_MALLOC cdef int ERROR_INTERNAL diff --git a/dpnp/tensor/_stride_utils.pyx b/dpnp/tensor/_stride_utils.pyx index 7d770e30e70..0a6a1d6e5bb 100644 --- a/dpnp/tensor/_stride_utils.pyx +++ b/dpnp/tensor/_stride_utils.pyx @@ -34,6 +34,10 @@ from cpython.ref cimport Py_INCREF from cpython.tuple cimport PyTuple_New, PyTuple_SetItem +cdef int USM_ARRAY_C_CONTIGUOUS = USM_ARRAY_C_CONTIGUOUS_VALUE +cdef int USM_ARRAY_F_CONTIGUOUS = USM_ARRAY_F_CONTIGUOUS_VALUE +cdef int USM_ARRAY_WRITABLE = USM_ARRAY_WRITABLE_VALUE + cdef int ERROR_MALLOC = 1 cdef int ERROR_INTERNAL = -1 cdef int ERROR_INCORRECT_ORDER = 2 diff --git a/dpnp/tensor/_types.pxd b/dpnp/tensor/_types.pxd index 5230688d4ea..6d644bc2166 100644 --- a/dpnp/tensor/_types.pxd +++ b/dpnp/tensor/_types.pxd @@ -30,23 +30,41 @@ # cython: language_level=3 cdef extern from "usm_ndarray_constants.h": - int UAR_BOOL - int UAR_BYTE - int UAR_UBYTE - int UAR_SHORT - int UAR_USHORT - int UAR_INT - int UAR_UINT - int UAR_LONG - int UAR_ULONG - int UAR_LONGLONG - int UAR_ULONGLONG - int UAR_FLOAT - int UAR_DOUBLE - int UAR_CFLOAT - int UAR_CDOUBLE - int UAR_TYPE_SENTINEL - int UAR_HALF + int UAR_BOOL_VALUE + int UAR_BYTE_VALUE + int UAR_UBYTE_VALUE + int UAR_SHORT_VALUE + int UAR_USHORT_VALUE + int UAR_INT_VALUE + int UAR_UINT_VALUE + int UAR_LONG_VALUE + int UAR_ULONG_VALUE + int UAR_LONGLONG_VALUE + int UAR_ULONGLONG_VALUE + int UAR_FLOAT_VALUE + int UAR_DOUBLE_VALUE + int UAR_CFLOAT_VALUE + int UAR_CDOUBLE_VALUE + int UAR_TYPE_SENTINEL_VALUE + int UAR_HALF_VALUE + +cdef int UAR_BOOL +cdef int UAR_BYTE +cdef int UAR_UBYTE +cdef int UAR_SHORT +cdef int UAR_USHORT +cdef int UAR_INT +cdef int UAR_UINT +cdef int UAR_LONG +cdef int UAR_ULONG +cdef int UAR_LONGLONG +cdef int UAR_ULONGLONG +cdef int UAR_FLOAT +cdef int UAR_DOUBLE +cdef int UAR_CFLOAT +cdef int UAR_CDOUBLE +cdef int UAR_TYPE_SENTINEL +cdef int UAR_HALF cdef int type_bytesize(int typenum) diff --git a/dpnp/tensor/_types.pyx b/dpnp/tensor/_types.pyx index 34cea76cfd0..5bda0a50f40 100644 --- a/dpnp/tensor/_types.pyx +++ b/dpnp/tensor/_types.pyx @@ -32,6 +32,25 @@ import numpy as np +cdef int UAR_BOOL = UAR_BOOL_VALUE +cdef int UAR_BYTE = UAR_BYTE_VALUE +cdef int UAR_UBYTE = UAR_UBYTE_VALUE +cdef int UAR_SHORT = UAR_SHORT_VALUE +cdef int UAR_USHORT = UAR_USHORT_VALUE +cdef int UAR_INT = UAR_INT_VALUE +cdef int UAR_UINT = UAR_UINT_VALUE +cdef int UAR_LONG = UAR_LONG_VALUE +cdef int UAR_ULONG = UAR_ULONG_VALUE +cdef int UAR_LONGLONG = UAR_LONGLONG_VALUE +cdef int UAR_ULONGLONG = UAR_ULONGLONG_VALUE +cdef int UAR_FLOAT = UAR_FLOAT_VALUE +cdef int UAR_DOUBLE = UAR_DOUBLE_VALUE +cdef int UAR_CFLOAT = UAR_CFLOAT_VALUE +cdef int UAR_CDOUBLE = UAR_CDOUBLE_VALUE +cdef int UAR_TYPE_SENTINEL = UAR_TYPE_SENTINEL_VALUE +cdef int UAR_HALF = UAR_HALF_VALUE + + cdef int type_bytesize(int typenum): """ NPY_BOOL=0 : 1 diff --git a/dpnp/tensor/_usmarray.pxd b/dpnp/tensor/_usmarray.pxd index 516a31cba71..4179f09ed45 100644 --- a/dpnp/tensor/_usmarray.pxd +++ b/dpnp/tensor/_usmarray.pxd @@ -33,27 +33,49 @@ cimport dpctl cdef extern from "usm_ndarray_constants.h": - int USM_ARRAY_C_CONTIGUOUS - int USM_ARRAY_F_CONTIGUOUS - int USM_ARRAY_WRITABLE + int USM_ARRAY_C_CONTIGUOUS_VALUE + int USM_ARRAY_F_CONTIGUOUS_VALUE + int USM_ARRAY_WRITABLE_VALUE - int UAR_BOOL - int UAR_BYTE - int UAR_UBYTE - int UAR_SHORT - int UAR_USHORT - int UAR_INT - int UAR_UINT - int UAR_LONG - int UAR_ULONG - int UAR_LONGLONG - int UAR_ULONGLONG - int UAR_FLOAT - int UAR_DOUBLE - int UAR_CFLOAT - int UAR_CDOUBLE - int UAR_TYPE_SENTINEL - int UAR_HALF + int UAR_BOOL_VALUE + int UAR_BYTE_VALUE + int UAR_UBYTE_VALUE + int UAR_SHORT_VALUE + int UAR_USHORT_VALUE + int UAR_INT_VALUE + int UAR_UINT_VALUE + int UAR_LONG_VALUE + int UAR_ULONG_VALUE + int UAR_LONGLONG_VALUE + int UAR_ULONGLONG_VALUE + int UAR_FLOAT_VALUE + int UAR_DOUBLE_VALUE + int UAR_CFLOAT_VALUE + int UAR_CDOUBLE_VALUE + int UAR_TYPE_SENTINEL_VALUE + int UAR_HALF_VALUE + +cdef public api int USM_ARRAY_C_CONTIGUOUS +cdef public api int USM_ARRAY_F_CONTIGUOUS +cdef public api int USM_ARRAY_WRITABLE + +cdef public api int UAR_BOOL +cdef public api int UAR_BYTE +cdef public api int UAR_UBYTE +cdef public api int UAR_SHORT +cdef public api int UAR_USHORT +cdef public api int UAR_INT +cdef public api int UAR_UINT +cdef public api int UAR_LONG +cdef public api int UAR_ULONG +cdef public api int UAR_LONGLONG +cdef public api int UAR_ULONGLONG +cdef public api int UAR_FLOAT +cdef public api int UAR_DOUBLE +cdef public api int UAR_CFLOAT +cdef public api int UAR_CDOUBLE +cdef public api int UAR_TYPE_SENTINEL +cdef public api int UAR_HALF cdef api class usm_ndarray [object PyUSMArrayObject, type PyUSMArrayType]: diff --git a/dpnp/tensor/_usmarray.pyx b/dpnp/tensor/_usmarray.pyx index 5c3c9430a84..a50aae736fc 100644 --- a/dpnp/tensor/_usmarray.pyx +++ b/dpnp/tensor/_usmarray.pyx @@ -86,6 +86,30 @@ from ._types cimport ( ) +# Public API constants initialized from usm_ndarray_constants.h +cdef int USM_ARRAY_C_CONTIGUOUS = USM_ARRAY_C_CONTIGUOUS_VALUE +cdef int USM_ARRAY_F_CONTIGUOUS = USM_ARRAY_F_CONTIGUOUS_VALUE +cdef int USM_ARRAY_WRITABLE = USM_ARRAY_WRITABLE_VALUE + +cdef int UAR_BOOL = UAR_BOOL_VALUE +cdef int UAR_BYTE = UAR_BYTE_VALUE +cdef int UAR_UBYTE = UAR_UBYTE_VALUE +cdef int UAR_SHORT = UAR_SHORT_VALUE +cdef int UAR_USHORT = UAR_USHORT_VALUE +cdef int UAR_INT = UAR_INT_VALUE +cdef int UAR_UINT = UAR_UINT_VALUE +cdef int UAR_LONG = UAR_LONG_VALUE +cdef int UAR_ULONG = UAR_ULONG_VALUE +cdef int UAR_LONGLONG = UAR_LONGLONG_VALUE +cdef int UAR_ULONGLONG = UAR_ULONGLONG_VALUE +cdef int UAR_FLOAT = UAR_FLOAT_VALUE +cdef int UAR_DOUBLE = UAR_DOUBLE_VALUE +cdef int UAR_CFLOAT = UAR_CFLOAT_VALUE +cdef int UAR_CDOUBLE = UAR_CDOUBLE_VALUE +cdef int UAR_TYPE_SENTINEL = UAR_TYPE_SENTINEL_VALUE +cdef int UAR_HALF = UAR_HALF_VALUE + + class DLDeviceType(IntEnum): """ An :class:`enum.IntEnum` for the types of DLDevices supported by the DLPack diff --git a/dpnp/tensor/include/usm_ndarray_constants.h b/dpnp/tensor/include/usm_ndarray_constants.h index dbe1f9da8b8..ebafeca5791 100644 --- a/dpnp/tensor/include/usm_ndarray_constants.h +++ b/dpnp/tensor/include/usm_ndarray_constants.h @@ -32,31 +32,31 @@ /* Array contiguity flags */ enum { - USM_ARRAY_C_CONTIGUOUS = 1, - USM_ARRAY_F_CONTIGUOUS = 2, - USM_ARRAY_WRITABLE = 4 + USM_ARRAY_C_CONTIGUOUS_VALUE = 1, + USM_ARRAY_F_CONTIGUOUS_VALUE = 2, + USM_ARRAY_WRITABLE_VALUE = 4 }; /* These typenum values are aligned to values in NumPy */ enum { - UAR_BOOL = 0, - UAR_BYTE = 1, - UAR_UBYTE = 2, - UAR_SHORT = 3, - UAR_USHORT = 4, - UAR_INT = 5, - UAR_UINT = 6, - UAR_LONG = 7, - UAR_ULONG = 8, - UAR_LONGLONG = 9, - UAR_ULONGLONG = 10, - UAR_FLOAT = 11, - UAR_DOUBLE = 12, - UAR_CFLOAT = 14, - UAR_CDOUBLE = 15, - UAR_TYPE_SENTINEL = 17, - UAR_HALF = 23 + UAR_BOOL_VALUE = 0, + UAR_BYTE_VALUE = 1, + UAR_UBYTE_VALUE = 2, + UAR_SHORT_VALUE = 3, + UAR_USHORT_VALUE = 4, + UAR_INT_VALUE = 5, + UAR_UINT_VALUE = 6, + UAR_LONG_VALUE = 7, + UAR_ULONG_VALUE = 8, + UAR_LONGLONG_VALUE = 9, + UAR_ULONGLONG_VALUE = 10, + UAR_FLOAT_VALUE = 11, + UAR_DOUBLE_VALUE = 12, + UAR_CFLOAT_VALUE = 14, + UAR_CDOUBLE_VALUE = 15, + UAR_TYPE_SENTINEL_VALUE = 17, + UAR_HALF_VALUE = 23 }; #endif /* DPNP_TENSOR_USM_NDARRAY_CONSTANTS_H */ From a4220324ce827980f02b08d61f98a8d14da9d1a0 Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 27 May 2026 05:06:33 -0700 Subject: [PATCH 10/12] Remove duplicate tests from tesn_usm_ndarray_ctor.py --- dpnp/tests/tensor/test_usm_ndarray_ctor.py | 71 ---------------------- 1 file changed, 71 deletions(-) diff --git a/dpnp/tests/tensor/test_usm_ndarray_ctor.py b/dpnp/tests/tensor/test_usm_ndarray_ctor.py index b03a01ad370..4ef2e7492b3 100644 --- a/dpnp/tests/tensor/test_usm_ndarray_ctor.py +++ b/dpnp/tests/tensor/test_usm_ndarray_ctor.py @@ -599,77 +599,6 @@ def test_datapi_device(): X.device.print_device_info() -def _pyx_capi_int(X, pyx_capi_name, caps_name=b"int", val_restype=ctypes.c_int): - import sys - - mod = sys.modules[X.__class__.__module__] - cap = mod.__pyx_capi__.get(pyx_capi_name, None) - if cap is None: - raise ValueError( - "__pyx_capi__ does not export {} capsule".format(pyx_capi_name) - ) - # construct Python callable to invoke these functions - cap_ptr_fn = ctypes.pythonapi.PyCapsule_GetPointer - cap_ptr_fn.restype = ctypes.c_void_p - cap_ptr_fn.argtypes = [ctypes.py_object, ctypes.c_char_p] - cap_ptr = cap_ptr_fn(cap, caps_name) - val_ptr = ctypes.cast(cap_ptr, ctypes.POINTER(val_restype)) - return val_ptr.contents.value - - -def test_pyx_capi_check_constants(): - try: - X = dpt.usm_ndarray(17, dtype="i1")[1::2] - except dpctl.SyclDeviceCreationError: - pytest.skip("No SYCL devices available") - cc_flag = _pyx_capi_int(X, "USM_ARRAY_C_CONTIGUOUS") - assert cc_flag > 0 and 0 == (cc_flag & (cc_flag - 1)) - fc_flag = _pyx_capi_int(X, "USM_ARRAY_F_CONTIGUOUS") - assert fc_flag > 0 and 0 == (fc_flag & (fc_flag - 1)) - w_flag = _pyx_capi_int(X, "USM_ARRAY_WRITABLE") - assert w_flag > 0 and 0 == (w_flag & (w_flag - 1)) - - bool_typenum = _pyx_capi_int(X, "UAR_BOOL") - assert bool_typenum == dpt.dtype("bool_").num - - byte_typenum = _pyx_capi_int(X, "UAR_BYTE") - assert byte_typenum == dpt.dtype(np.byte).num - ubyte_typenum = _pyx_capi_int(X, "UAR_UBYTE") - assert ubyte_typenum == dpt.dtype(np.ubyte).num - - short_typenum = _pyx_capi_int(X, "UAR_SHORT") - assert short_typenum == dpt.dtype(np.short).num - ushort_typenum = _pyx_capi_int(X, "UAR_USHORT") - assert ushort_typenum == dpt.dtype(np.ushort).num - - int_typenum = _pyx_capi_int(X, "UAR_INT") - assert int_typenum == dpt.dtype(np.intc).num - uint_typenum = _pyx_capi_int(X, "UAR_UINT") - assert uint_typenum == dpt.dtype(np.uintc).num - - long_typenum = _pyx_capi_int(X, "UAR_LONG") - assert long_typenum == dpt.dtype("l").num - ulong_typenum = _pyx_capi_int(X, "UAR_ULONG") - assert ulong_typenum == dpt.dtype("L").num - - longlong_typenum = _pyx_capi_int(X, "UAR_LONGLONG") - assert longlong_typenum == dpt.dtype(np.longlong).num - ulonglong_typenum = _pyx_capi_int(X, "UAR_ULONGLONG") - assert ulonglong_typenum == dpt.dtype(np.ulonglong).num - - half_typenum = _pyx_capi_int(X, "UAR_HALF") - assert half_typenum == dpt.dtype(np.half).num - float_typenum = _pyx_capi_int(X, "UAR_FLOAT") - assert float_typenum == dpt.dtype(np.single).num - double_typenum = _pyx_capi_int(X, "UAR_DOUBLE") - assert double_typenum == dpt.dtype(np.double).num - - cfloat_typenum = _pyx_capi_int(X, "UAR_CFLOAT") - assert cfloat_typenum == dpt.dtype(np.csingle).num - cdouble_typenum = _pyx_capi_int(X, "UAR_CDOUBLE") - assert cdouble_typenum == dpt.dtype(np.cdouble).num - - @pytest.mark.parametrize( "shape", [(), (1,), (5,), (2, 3), (2, 3, 4), (2, 2, 2, 2, 2)] ) From 4ac3cb8ba109efb2f72257c3b2bd73a32d302b9a Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Wed, 27 May 2026 06:22:14 -0700 Subject: [PATCH 11/12] Remove unused imports in _usmarray.pyx --- dpnp/tensor/_usmarray.pyx | 10 +--------- 1 file changed, 1 insertion(+), 9 deletions(-) diff --git a/dpnp/tensor/_usmarray.pyx b/dpnp/tensor/_usmarray.pyx index a50aae736fc..3344f757cf1 100644 --- a/dpnp/tensor/_usmarray.pyx +++ b/dpnp/tensor/_usmarray.pyx @@ -55,18 +55,12 @@ from . import _flags from ._dlpack import get_build_dlpack_version from ._tensor_impl import default_device_fp_type -from ._slicing cimport ( - _is_boolean, - _is_buffer, - _is_integral, - _slice_len, -) +from ._slicing cimport _is_buffer from ._slicing import _basic_slice_meta from ._stride_utils cimport ( ERROR_INCORRECT_ORDER, - ERROR_INTERNAL, ERROR_MALLOC, ERROR_UNEXPECTED_STRIDES, _c_contig_strides, @@ -79,10 +73,8 @@ from ._stride_utils cimport ( ) from ._types cimport ( _make_typestr, - descr_to_typenum, dtype_to_typenum, type_bytesize, - typenum_from_format, ) From 6a3858e563430710f59de91c801c64ae3c53569f Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Thu, 28 May 2026 07:00:16 -0700 Subject: [PATCH 12/12] Remove redundant constant declarations from _types and _stride_utils --- dpnp/tensor/_stride_utils.pxd | 9 --------- dpnp/tensor/_stride_utils.pyx | 5 ++++- dpnp/tensor/_types.pxd | 36 ----------------------------------- dpnp/tensor/_types.pyx | 19 ------------------ 4 files changed, 4 insertions(+), 65 deletions(-) diff --git a/dpnp/tensor/_stride_utils.pxd b/dpnp/tensor/_stride_utils.pxd index 9aead47d76c..65ca61e3095 100644 --- a/dpnp/tensor/_stride_utils.pxd +++ b/dpnp/tensor/_stride_utils.pxd @@ -29,15 +29,6 @@ # distutils: language = c++ # cython: language_level=3 -cdef extern from "usm_ndarray_constants.h": - int USM_ARRAY_C_CONTIGUOUS_VALUE - int USM_ARRAY_F_CONTIGUOUS_VALUE - int USM_ARRAY_WRITABLE_VALUE - -cdef int USM_ARRAY_C_CONTIGUOUS -cdef int USM_ARRAY_F_CONTIGUOUS -cdef int USM_ARRAY_WRITABLE - cdef int ERROR_MALLOC cdef int ERROR_INTERNAL cdef int ERROR_INCORRECT_ORDER diff --git a/dpnp/tensor/_stride_utils.pyx b/dpnp/tensor/_stride_utils.pyx index 0a6a1d6e5bb..a9d1c933c45 100644 --- a/dpnp/tensor/_stride_utils.pyx +++ b/dpnp/tensor/_stride_utils.pyx @@ -34,9 +34,12 @@ from cpython.ref cimport Py_INCREF from cpython.tuple cimport PyTuple_New, PyTuple_SetItem +cdef extern from "usm_ndarray_constants.h": + int USM_ARRAY_C_CONTIGUOUS_VALUE + int USM_ARRAY_F_CONTIGUOUS_VALUE + cdef int USM_ARRAY_C_CONTIGUOUS = USM_ARRAY_C_CONTIGUOUS_VALUE cdef int USM_ARRAY_F_CONTIGUOUS = USM_ARRAY_F_CONTIGUOUS_VALUE -cdef int USM_ARRAY_WRITABLE = USM_ARRAY_WRITABLE_VALUE cdef int ERROR_MALLOC = 1 cdef int ERROR_INTERNAL = -1 diff --git a/dpnp/tensor/_types.pxd b/dpnp/tensor/_types.pxd index 6d644bc2166..a4f1ed9fbf7 100644 --- a/dpnp/tensor/_types.pxd +++ b/dpnp/tensor/_types.pxd @@ -29,42 +29,6 @@ # distutils: language = c++ # cython: language_level=3 -cdef extern from "usm_ndarray_constants.h": - int UAR_BOOL_VALUE - int UAR_BYTE_VALUE - int UAR_UBYTE_VALUE - int UAR_SHORT_VALUE - int UAR_USHORT_VALUE - int UAR_INT_VALUE - int UAR_UINT_VALUE - int UAR_LONG_VALUE - int UAR_ULONG_VALUE - int UAR_LONGLONG_VALUE - int UAR_ULONGLONG_VALUE - int UAR_FLOAT_VALUE - int UAR_DOUBLE_VALUE - int UAR_CFLOAT_VALUE - int UAR_CDOUBLE_VALUE - int UAR_TYPE_SENTINEL_VALUE - int UAR_HALF_VALUE - -cdef int UAR_BOOL -cdef int UAR_BYTE -cdef int UAR_UBYTE -cdef int UAR_SHORT -cdef int UAR_USHORT -cdef int UAR_INT -cdef int UAR_UINT -cdef int UAR_LONG -cdef int UAR_ULONG -cdef int UAR_LONGLONG -cdef int UAR_ULONGLONG -cdef int UAR_FLOAT -cdef int UAR_DOUBLE -cdef int UAR_CFLOAT -cdef int UAR_CDOUBLE -cdef int UAR_TYPE_SENTINEL -cdef int UAR_HALF cdef int type_bytesize(int typenum) diff --git a/dpnp/tensor/_types.pyx b/dpnp/tensor/_types.pyx index 5bda0a50f40..34cea76cfd0 100644 --- a/dpnp/tensor/_types.pyx +++ b/dpnp/tensor/_types.pyx @@ -32,25 +32,6 @@ import numpy as np -cdef int UAR_BOOL = UAR_BOOL_VALUE -cdef int UAR_BYTE = UAR_BYTE_VALUE -cdef int UAR_UBYTE = UAR_UBYTE_VALUE -cdef int UAR_SHORT = UAR_SHORT_VALUE -cdef int UAR_USHORT = UAR_USHORT_VALUE -cdef int UAR_INT = UAR_INT_VALUE -cdef int UAR_UINT = UAR_UINT_VALUE -cdef int UAR_LONG = UAR_LONG_VALUE -cdef int UAR_ULONG = UAR_ULONG_VALUE -cdef int UAR_LONGLONG = UAR_LONGLONG_VALUE -cdef int UAR_ULONGLONG = UAR_ULONGLONG_VALUE -cdef int UAR_FLOAT = UAR_FLOAT_VALUE -cdef int UAR_DOUBLE = UAR_DOUBLE_VALUE -cdef int UAR_CFLOAT = UAR_CFLOAT_VALUE -cdef int UAR_CDOUBLE = UAR_CDOUBLE_VALUE -cdef int UAR_TYPE_SENTINEL = UAR_TYPE_SENTINEL_VALUE -cdef int UAR_HALF = UAR_HALF_VALUE - - cdef int type_bytesize(int typenum): """ NPY_BOOL=0 : 1