Merge remote-tracking branch 'origin' into add_targets_args

Author: vlad-perevezentsev

.github/workflows/generate-coverage.yaml

@@ -91,8 +91,7 @@ jobs:
       - name: Install dpctl dependencies
         shell: bash -l {0}
         run: |
-          # TODO: unpin numpy when numpy#29167 resolved
-          pip install numpy"<2.3.0" cython setuptools"<80" pytest pytest-cov scikit-build cmake coverage[toml] versioneer[toml]==0.29
+          pip install numpy cython setuptools"<80" pytest pytest-cov scikit-build cmake coverage[toml] versioneer[toml]==0.29
 
       - name: Build dpctl with coverage
         shell: bash -l {0}

.github/workflows/openssf-scorecard.yml

@@ -69,6 +69,6 @@ jobs:
 
       # Upload the results to GitHub's code scanning dashboard.
       - name: "Upload to code-scanning"
-        uses: github/codeql-action/upload-sarif@ce28f5bb42b7a9f2c824e633a3f6ee835bab6858 # v3.29.0
+        uses: github/codeql-action/upload-sarif@d6bbdef45e766d081b84a2def353b0055f728d3e # v3.29.3
         with:
           sarif_file: results.sarif

.github/workflows/os-llvm-sycl-build.yml

@@ -107,8 +107,7 @@ jobs:
       - name: Install dpctl dependencies
         shell: bash -l {0}
         run: |
-          # TODO: unpin numpy when numpy#29167 resolved
-          pip install numpy"<2.3.0" cython setuptools"<80" pytest scikit-build cmake ninja versioneer[toml]==0.29
+          pip install numpy cython setuptools"<80" pytest scikit-build cmake ninja versioneer[toml]==0.29
 
       - name: Checkout repo
         uses: actions/checkout@v4.2.2

CHANGELOG.md

@@ -14,6 +14,13 @@ and this project adheres to [Semantic Versioning](https://semver.org/spec/v2.0.0
 
 ### Maintenance
 
+## [0.20.2] - Jun. 26, 2025
+
+### Maintenance
+
+* Add Python 3.13 to package metadata [gh-2110](https://github.com/IntelPython/dpctl/pull/2110)
+* When building dpctl conda package for Python 3.13, restrict Cython to below 3.1.0, as this version and higher may cause crashes [gh-2112](https://github.com/IntelPython/dpctl/pull/2112)
+
 ## [0.20.1] - Jun. 06, 2025
 
 ### Fixed

conda-recipe/meta.yaml

@@ -41,6 +41,12 @@ requirements:
         - {{ dep }}
         {% elif dep.startswith('build>=') %}
         - {{ 'python-' ~ dep }}
+        {% elif dep.startswith('cython') %}
+        {% if dep.split(';')[1] == "python_version<'3.13'" %}
+        - {{ dep.split(';')[0] }} # [py<313]
+        {% else %}
+        - {{ dep.split(';')[0] }} # [py>=313]
+        {% endif %}
         {% else %}
         - {{ dep|replace('_','-') }}
         {% endif %}

dpctl/tensor/_copy_utils.py

@@ -41,6 +41,9 @@
 def _copy_to_numpy(ary):
     if not isinstance(ary, dpt.usm_ndarray):
         raise TypeError(f"Expected dpctl.tensor.usm_ndarray, got {type(ary)}")
+    if ary.size == 0:
+        # no data needs to be copied for zero sized array
+        return np.ndarray(ary.shape, dtype=ary.dtype)
     nb = ary.usm_data.nbytes
     q = ary.sycl_queue
     hh = dpm.MemoryUSMHost(nb, queue=q)
@@ -739,6 +742,9 @@ def astype(
             order=copy_order,
             buffer_ctor_kwargs={"queue": usm_ary.sycl_queue},
         )
+    # see #2121
+    if ary_dtype == dpt.bool:
+        usm_ary = dpt.not_equal(usm_ary, 0, order=copy_order)
     _copy_from_usm_ndarray_to_usm_ndarray(R, usm_ary)
     return R
 

dpctl/tensor/_usmarray.pyx

@@ -382,8 +382,9 @@ cdef class usm_ndarray:
         else:
             self._cleanup()
             raise ValueError("buffer='{}' was not understood.".format(buffer))
-        if (_offset + ary_min_displacement < 0 or
-           (_offset + ary_max_displacement + 1) * itemsize > _buffer.nbytes):
+        if (shape_to_elem_count(nd, shape_ptr) > 0 and
+            (_offset + ary_min_displacement < 0 or
+             (_offset + ary_max_displacement + 1) * itemsize > _buffer.nbytes)):
             self._cleanup()
             raise ValueError(("buffer='{}' can not accommodate "
                               "the requested array.").format(buffer))
@@ -1324,8 +1325,8 @@ cdef class usm_ndarray:
         allocated, or the non-partitioned parent device of the allocation
         device.
 
-        See ``DLDeviceType`` for a list of devices supported by the DLPack
-        protocol.
+        See :class:`dpctl.tensor.DLDeviceType` for a list of devices supported
+        by the DLPack protocol.
 
         Raises:
             DLPackCreationError:

dpctl/tensor/libtensor/source/sorting/py_argsort_common.hpp

@@ -126,22 +126,43 @@ py_argsort(const dpctl::tensor::usm_ndarray &src,
     bool is_dst_c_contig = dst.is_c_contiguous();
 
     if (is_src_c_contig && is_dst_c_contig) {
-        static constexpr py::ssize_t zero_offset = py::ssize_t(0);
+        if (sort_nelems > 1) {
+            static constexpr py::ssize_t zero_offset = py::ssize_t(0);
 
-        auto fn = sort_contig_fns[src_typeid][dst_typeid];
+            auto fn = sort_contig_fns[src_typeid][dst_typeid];
 
-        if (fn == nullptr) {
-            throw py::value_error("Not implemented for dtypes of input arrays");
+            if (fn == nullptr) {
+                throw py::value_error(
+                    "Not implemented for dtypes of input arrays");
+            }
+
+            sycl::event comp_ev =
+                fn(exec_q, iter_nelems, sort_nelems, src.get_data(),
+                   dst.get_data(), zero_offset, zero_offset, zero_offset,
+                   zero_offset, depends);
+
+            sycl::event keep_args_alive_ev =
+                dpctl::utils::keep_args_alive(exec_q, {src, dst}, {comp_ev});
+
+            return std::make_pair(keep_args_alive_ev, comp_ev);
         }
+        else {
+            assert(dst.get_size() == iter_nelems);
+            int dst_elemsize = dst.get_elemsize();
+            static constexpr int memset_val(0);
 
-        sycl::event comp_ev =
-            fn(exec_q, iter_nelems, sort_nelems, src.get_data(), dst.get_data(),
-               zero_offset, zero_offset, zero_offset, zero_offset, depends);
+            sycl::event fill_ev = exec_q.submit([&](sycl::handler &cgh) {
+                cgh.depends_on(depends);
 
-        sycl::event keep_args_alive_ev =
-            dpctl::utils::keep_args_alive(exec_q, {src, dst}, {comp_ev});
+                cgh.memset(reinterpret_cast<void *>(dst.get_data()), memset_val,
+                           iter_nelems * dst_elemsize);
+            });
 
-        return std::make_pair(keep_args_alive_ev, comp_ev);
+            sycl::event keep_args_alive_ev =
+                dpctl::utils::keep_args_alive(exec_q, {src, dst}, {fill_ev});
+
+            return std::make_pair(keep_args_alive_ev, fill_ev);
+        }
     }
 
     throw py::value_error(

dpctl/tensor/libtensor/source/sorting/py_sort_common.hpp

@@ -127,23 +127,38 @@ py_sort(const dpctl::tensor::usm_ndarray &src,
     bool is_dst_c_contig = dst.is_c_contiguous();
 
     if (is_src_c_contig && is_dst_c_contig) {
-        static constexpr py::ssize_t zero_offset = py::ssize_t(0);
+        if (sort_nelems > 1) {
+            static constexpr py::ssize_t zero_offset = py::ssize_t(0);
 
-        auto fn = sort_contig_fns[src_typeid];
+            auto fn = sort_contig_fns[src_typeid];
 
-        if (nullptr == fn) {
-            throw py::value_error(
-                "Not implemented for the dtype of input arrays");
-        }
+            if (nullptr == fn) {
+                throw py::value_error(
+                    "Not implemented for the dtype of input arrays");
+            }
+
+            sycl::event comp_ev =
+                fn(exec_q, iter_nelems, sort_nelems, src.get_data(),
+                   dst.get_data(), zero_offset, zero_offset, zero_offset,
+                   zero_offset, depends);
 
-        sycl::event comp_ev =
-            fn(exec_q, iter_nelems, sort_nelems, src.get_data(), dst.get_data(),
-               zero_offset, zero_offset, zero_offset, zero_offset, depends);
+            sycl::event keep_args_alive_ev =
+                dpctl::utils::keep_args_alive(exec_q, {src, dst}, {comp_ev});
 
-        sycl::event keep_args_alive_ev =
-            dpctl::utils::keep_args_alive(exec_q, {src, dst}, {comp_ev});
+            return std::make_pair(keep_args_alive_ev, comp_ev);
+        }
+        else {
+            assert(dst.get_size() == iter_nelems);
+            int src_elemsize = src.get_elemsize();
 
-        return std::make_pair(keep_args_alive_ev, comp_ev);
+            sycl::event copy_ev =
+                exec_q.copy<char>(src.get_data(), dst.get_data(),
+                                  src_elemsize * iter_nelems, depends);
+
+            return std::make_pair(
+                dpctl::utils::keep_args_alive(exec_q, {src, dst}, {copy_ev}),
+                copy_ev);
+        }
     }
 
     throw py::value_error(

dpctl/tests/elementwise/test_hyperbolic.py

@@ -14,10 +14,6 @@
 #  See the License for the specific language governing permissions and
 #  limitations under the License.
 
-import itertools
-import os
-import re
-
 import numpy as np
 import pytest
 from numpy.testing import assert_allclose
@@ -34,7 +30,6 @@
     (np.arctanh, dpt.atanh),
 ]
 _all_funcs = _hyper_funcs + _inv_hyper_funcs
-_dpt_funcs = [t[1] for t in _all_funcs]
 
 
 @pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
@@ -45,17 +40,10 @@ def test_hyper_out_type(np_call, dpt_call, dtype):
 
     a = 1 if np_call == np.arccosh else 0
 
-    X = dpt.asarray(a, dtype=dtype, sycl_queue=q)
-    expected_dtype = np_call(np.array(a, dtype=dtype)).dtype
-    expected_dtype = _map_to_device_dtype(expected_dtype, q.sycl_device)
-    assert dpt_call(X).dtype == expected_dtype
-
-    X = dpt.asarray(a, dtype=dtype, sycl_queue=q)
+    x = dpt.asarray(a, dtype=dtype, sycl_queue=q)
     expected_dtype = np_call(np.array(a, dtype=dtype)).dtype
     expected_dtype = _map_to_device_dtype(expected_dtype, q.sycl_device)
-    Y = dpt.empty_like(X, dtype=expected_dtype)
-    dpt_call(X, out=Y)
-    assert_allclose(dpt.asnumpy(dpt_call(X)), dpt.asnumpy(Y))
+    assert dpt_call(x).dtype == expected_dtype
 
 
 @pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
@@ -119,79 +107,6 @@ def test_hyper_complex_contig(np_call, dpt_call, dtype):
     )
 
 
-@pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
-@pytest.mark.parametrize("usm_type", ["device", "shared", "host"])
-def test_hyper_usm_type(np_call, dpt_call, usm_type):
-    q = get_queue_or_skip()
-
-    arg_dt = np.dtype("f4")
-    input_shape = (10, 10, 10, 10)
-    X = dpt.empty(input_shape, dtype=arg_dt, usm_type=usm_type, sycl_queue=q)
-    if np_call == np.arctanh:
-        X[..., 0::2] = -0.4
-        X[..., 1::2] = 0.3
-    elif np_call == np.arccosh:
-        X[..., 0::2] = 2.2
-        X[..., 1::2] = 5.5
-    else:
-        X[..., 0::2] = -4.4
-        X[..., 1::2] = 5.5
-
-    Y = dpt_call(X)
-    assert Y.usm_type == X.usm_type
-    assert Y.sycl_queue == X.sycl_queue
-    assert Y.flags.c_contiguous
-
-    expected_Y = np_call(dpt.asnumpy(X))
-    tol = 8 * dpt.finfo(Y.dtype).resolution
-    assert_allclose(dpt.asnumpy(Y), expected_Y, atol=tol, rtol=tol)
-
-
-@pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
-@pytest.mark.parametrize("dtype", _all_dtypes)
-def test_hyper_order(np_call, dpt_call, dtype):
-    q = get_queue_or_skip()
-    skip_if_dtype_not_supported(dtype, q)
-
-    arg_dt = np.dtype(dtype)
-    input_shape = (4, 4, 4, 4)
-    X = dpt.empty(input_shape, dtype=arg_dt, sycl_queue=q)
-    if np_call == np.arctanh:
-        X[..., 0::2] = -0.4
-        X[..., 1::2] = 0.3
-    elif np_call == np.arccosh:
-        X[..., 0::2] = 2.2
-        X[..., 1::2] = 5.5
-    else:
-        X[..., 0::2] = -4.4
-        X[..., 1::2] = 5.5
-
-    for perms in itertools.permutations(range(4)):
-        U = dpt.permute_dims(X[:, ::-1, ::-1, :], perms)
-        with np.errstate(all="ignore"):
-            expected_Y = np_call(dpt.asnumpy(U))
-        for ord in ["C", "F", "A", "K"]:
-            Y = dpt_call(U, order=ord)
-            tol = 8 * max(
-                dpt.finfo(Y.dtype).resolution,
-                np.finfo(expected_Y.dtype).resolution,
-            )
-            assert_allclose(dpt.asnumpy(Y), expected_Y, atol=tol, rtol=tol)
-
-
-@pytest.mark.parametrize("callable", _dpt_funcs)
-@pytest.mark.parametrize("dtype", _all_dtypes)
-def test_hyper_error_dtype(callable, dtype):
-    q = get_queue_or_skip()
-    skip_if_dtype_not_supported(dtype, q)
-
-    x = dpt.ones(5, dtype=dtype)
-    y = dpt.empty_like(x, dtype="int16")
-    with pytest.raises(ValueError) as excinfo:
-        callable(x, out=y)
-    assert re.match("Output array of type.*is needed", str(excinfo.value))
-
-
 @pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
 @pytest.mark.parametrize("dtype", ["f2", "f4", "f8"])
 def test_hyper_real_strided(np_call, dpt_call, dtype):
@@ -270,46 +185,3 @@ def test_hyper_real_special_cases(np_call, dpt_call, dtype):
 
     tol = 8 * dpt.finfo(dtype).resolution
     assert_allclose(dpt.asnumpy(dpt_call(yf)), Y_np, atol=tol, rtol=tol)
-
-
-@pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
-@pytest.mark.parametrize("dtype", ["c8", "c16"])
-def test_hyper_complex_special_cases_conj_property(np_call, dpt_call, dtype):
-    q = get_queue_or_skip()
-    skip_if_dtype_not_supported(dtype, q)
-
-    x = [np.nan, np.inf, -np.inf, +0.0, -0.0, +1.0, -1.0]
-    xc = [complex(*val) for val in itertools.product(x, repeat=2)]
-
-    Xc_np = np.array(xc, dtype=dtype)
-    Xc = dpt.asarray(Xc_np, dtype=dtype, sycl_queue=q)
-
-    tol = 50 * dpt.finfo(dtype).resolution
-    Y = dpt_call(Xc)
-    Yc = dpt_call(dpt.conj(Xc))
-
-    dpt.allclose(Y, dpt.conj(Yc), atol=tol, rtol=tol)
-
-
-@pytest.mark.skipif(
-    os.name != "posix", reason="Known to fail on Windows due to bug in NumPy"
-)
-@pytest.mark.parametrize("np_call, dpt_call", _all_funcs)
-@pytest.mark.parametrize("dtype", ["c8", "c16"])
-def test_hyper_complex_special_cases(np_call, dpt_call, dtype):
-    q = get_queue_or_skip()
-    skip_if_dtype_not_supported(dtype, q)
-
-    x = [np.nan, np.inf, -np.inf, +0.0, -0.0, +1.0, -1.0]
-    xc = [complex(*val) for val in itertools.product(x, repeat=2)]
-
-    Xc_np = np.array(xc, dtype=dtype)
-    Xc = dpt.asarray(Xc_np, dtype=dtype, sycl_queue=q)
-
-    with np.errstate(all="ignore"):
-        Ynp = np_call(Xc_np)
-
-    tol = 50 * dpt.finfo(dtype).resolution
-    Y = dpt_call(Xc)
-    assert_allclose(dpt.asnumpy(dpt.real(Y)), np.real(Ynp), atol=tol, rtol=tol)
-    assert_allclose(dpt.asnumpy(dpt.imag(Y)), np.imag(Ynp), atol=tol, rtol=tol)