From ae9c5489c8aaff992a11fe126faa859d3bc8608c Mon Sep 17 00:00:00 2001 From: Vladislav Perevezentsev Date: Tue, 12 May 2026 05:26:06 -0700 Subject: [PATCH 1/7] 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 2/7] 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 3/7] 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 4/7] 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 5/7] 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 6/7] 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 7/7] 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 */