From e825de57c080ae53eecb51b80f1a4a7ad47b5afb Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Tue, 24 Feb 2026 10:03:14 +0100 Subject: [PATCH 01/35] Bump github/codeql-action from 4.32.3 to 4.32.4 (#2780) Bumps [github/codeql-action](https://github.com/github/codeql-action) from 4.32.3 to 4.32.4. --- .github/workflows/openssf-scorecard.yml | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/.github/workflows/openssf-scorecard.yml b/.github/workflows/openssf-scorecard.yml index 8bf5e86d03ed..6ca1f6682784 100644 --- a/.github/workflows/openssf-scorecard.yml +++ b/.github/workflows/openssf-scorecard.yml @@ -72,6 +72,6 @@ jobs: # Upload the results to GitHub's code scanning dashboard. - name: "Upload to code-scanning" - uses: github/codeql-action/upload-sarif@9e907b5e64f6b83e7804b09294d44122997950d6 # v4.32.3 + uses: github/codeql-action/upload-sarif@89a39a4e59826350b863aa6b6252a07ad50cf83e # v4.32.4 with: sarif_file: results.sarif From d545555e133d2ca233aa8f01d895f8992be02723 Mon Sep 17 00:00:00 2001 From: "github-actions[bot]" <41898282+github-actions[bot]@users.noreply.github.com> Date: Tue, 24 Feb 2026 12:38:47 +0100 Subject: [PATCH 02/35] Weekly pre-commit autoupdate (#2779) This PR updates the `.pre-commit-config.yaml` using `pre-commit autoupdate`. --- .pre-commit-config.yaml | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index ace139f8d179..008adb4589d0 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -69,7 +69,7 @@ repos: - id: black exclude: "dpnp/_version.py" - repo: https://github.com/pycqa/isort - rev: 7.0.0 + rev: 8.0.0 hooks: - id: isort name: isort (python) @@ -123,11 +123,11 @@ repos: - id: pretty-format-toml args: [--autofix] - repo: https://github.com/rhysd/actionlint - rev: v1.7.10 + rev: v1.7.11 hooks: - id: actionlint - repo: https://github.com/BlankSpruce/gersemi - rev: 0.25.4 + rev: 0.26.0 hooks: - id: gersemi exclude: "dpnp/backend/cmake/Modules/" From 4ca0ffa1fffcd997609583ee67c9f7b56566c3ec Mon Sep 17 00:00:00 2001 From: "github-actions[bot]" <41898282+github-actions[bot]@users.noreply.github.com> Date: Mon, 2 Mar 2026 17:56:24 +0100 Subject: [PATCH 03/35] Weekly pre-commit autoupdate (#2788) This PR updates the `.pre-commit-config.yaml` using `pre-commit autoupdate`. --- .pre-commit-config.yaml | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 008adb4589d0..66245039ce3c 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -2,7 +2,7 @@ # See https://pre-commit.com/hooks.html for more hooks repos: - repo: https://github.com/PyCQA/bandit - rev: '1.9.3' + rev: '1.9.4' hooks: - id: bandit pass_filenames: false From 783f29813629cda5b3834ae72e1c90464d8095a0 Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Mon, 2 Mar 2026 19:01:51 +0100 Subject: [PATCH 04/35] Bump actions/download-artifact from 7.0.0 to 8.0.0 (#2790) Bumps [actions/download-artifact](https://github.com/actions/download-artifact) from 7.0.0 to 8.0.0. --- .github/workflows/check-onemath.yaml | 4 ++-- .github/workflows/conda-package.yml | 10 +++++----- 2 files changed, 7 insertions(+), 7 deletions(-) diff --git a/.github/workflows/check-onemath.yaml b/.github/workflows/check-onemath.yaml index 7b18cdfaba64..9a6e92bfc36b 100644 --- a/.github/workflows/check-onemath.yaml +++ b/.github/workflows/check-onemath.yaml @@ -87,7 +87,7 @@ jobs: fetch-depth: 0 - name: Download artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.environment-file-name }} path: ${{ env.environment-file-loc }} @@ -181,7 +181,7 @@ jobs: fetch-depth: 0 - name: Download artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.environment-file-name }} path: ${{ env.environment-file-loc }} diff --git a/.github/workflows/conda-package.yml b/.github/workflows/conda-package.yml index d2ac90621aaa..2ef8e42ca705 100644 --- a/.github/workflows/conda-package.yml +++ b/.github/workflows/conda-package.yml @@ -151,7 +151,7 @@ jobs: path: ${{ env.dpnp-repo-path }} - name: Download artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} path: ${{ env.pkg-path-in-channel }} @@ -280,7 +280,7 @@ jobs: path: ${{ env.dpnp-repo-path }} - name: Download artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} path: ${{ env.pkg-path-in-channel }} @@ -439,12 +439,12 @@ jobs: fetch-depth: ${{ env.fetch-depth }} - name: Download artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} - name: Download wheels artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Wheels Python ${{ matrix.python }} @@ -528,7 +528,7 @@ jobs: path: ${{ env.dpnp-repo-path }} - name: Download artifact - uses: actions/download-artifact@37930b1c2abaa49bbe596cd826c3c89aef350131 # v7.0.0 + uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ env.python-ver }} path: ${{ env.pkg-path-in-channel }} From fd64f3cb7d0d9cbaab62b2a162a962e3880750c2 Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Mon, 2 Mar 2026 20:27:59 +0100 Subject: [PATCH 05/35] Bump actions/upload-artifact from 6.0.0 to 7.0.0 (#2789) Bumps [actions/upload-artifact](https://github.com/actions/upload-artifact) from 6.0.0 to 7.0.0. --- .github/workflows/check-onemath.yaml | 2 +- .github/workflows/conda-package.yml | 4 ++-- .github/workflows/openssf-scorecard.yml | 2 +- 3 files changed, 4 insertions(+), 4 deletions(-) diff --git a/.github/workflows/check-onemath.yaml b/.github/workflows/check-onemath.yaml index 9a6e92bfc36b..409117c692b9 100644 --- a/.github/workflows/check-onemath.yaml +++ b/.github/workflows/check-onemath.yaml @@ -57,7 +57,7 @@ jobs: cat ${{ env.environment-file }} - name: Upload artifact - uses: actions/upload-artifact@b7c566a772e6b6bfb58ed0dc250532a479d7789f # v6.0.0 + uses: actions/upload-artifact@bbbca2ddaa5d8feaa63e36b76fdaad77386f024f # v7.0.0 with: name: ${{ env.environment-file-name }} path: ${{ env.environment-file }} diff --git a/.github/workflows/conda-package.yml b/.github/workflows/conda-package.yml index 2ef8e42ca705..a12486300aa0 100644 --- a/.github/workflows/conda-package.yml +++ b/.github/workflows/conda-package.yml @@ -102,13 +102,13 @@ jobs: MAX_BUILD_CMPL_MKL_VERSION: '2026.0a0' - name: Upload artifact - uses: actions/upload-artifact@b7c566a772e6b6bfb58ed0dc250532a479d7789f # v6.0.0 + uses: actions/upload-artifact@bbbca2ddaa5d8feaa63e36b76fdaad77386f024f # v7.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} path: ${{ env.CONDA_BLD }}${{ env.package-name }}-*.conda - name: Upload wheels artifact - uses: actions/upload-artifact@b7c566a772e6b6bfb58ed0dc250532a479d7789f # v6.0.0 + uses: actions/upload-artifact@bbbca2ddaa5d8feaa63e36b76fdaad77386f024f # v7.0.0 with: name: ${{ env.package-name }} ${{ runner.os }} Wheels Python ${{ matrix.python }} path: ${{ env.WHEELS_OUTPUT_FOLDER }}${{ env.package-name }}-*.whl diff --git a/.github/workflows/openssf-scorecard.yml b/.github/workflows/openssf-scorecard.yml index 6ca1f6682784..8b4cc3b93f64 100644 --- a/.github/workflows/openssf-scorecard.yml +++ b/.github/workflows/openssf-scorecard.yml @@ -64,7 +64,7 @@ jobs: # Upload the results as artifacts (optional). Commenting out will disable uploads of run results in SARIF # format to the repository Actions tab. - name: "Upload artifact" - uses: actions/upload-artifact@b7c566a772e6b6bfb58ed0dc250532a479d7789f # v6.0.0 + uses: actions/upload-artifact@bbbca2ddaa5d8feaa63e36b76fdaad77386f024f # v7.0.0 with: name: SARIF file path: results.sarif From 7803d3a783e0ea9a00ed6e2aeb31eeab1132e5b2 Mon Sep 17 00:00:00 2001 From: Anton <100830759+antonwolfy@users.noreply.github.com> Date: Tue, 3 Mar 2026 13:21:00 +0100 Subject: [PATCH 06/35] Mute expecting runtime warning raised in tests (#2792) The PR updates marks observing `RuntimeWarning` as expected to suppress them. Also the PR update DLPack tests to avoid raising the warning in SYCL queue relating tests. --- dpnp/tests/test_mathematical.py | 1 + dpnp/tests/test_special.py | 3 +++ dpnp/tests/test_sycl_queue.py | 4 ++-- 3 files changed, 6 insertions(+), 2 deletions(-) diff --git a/dpnp/tests/test_mathematical.py b/dpnp/tests/test_mathematical.py index 760c1a0ceb2e..e1f32bbd7931 100644 --- a/dpnp/tests/test_mathematical.py +++ b/dpnp/tests/test_mathematical.py @@ -1733,6 +1733,7 @@ def test_nan_infs_complex(self): class TestSpacing: + @pytest.mark.filterwarnings("ignore::RuntimeWarning") @pytest.mark.parametrize("sign", [1, -1]) @pytest.mark.parametrize("dt", get_float_dtypes()) def test_basic(self, sign, dt): diff --git a/dpnp/tests/test_special.py b/dpnp/tests/test_special.py index 44b2d38e6919..1ebb64d8da7f 100644 --- a/dpnp/tests/test_special.py +++ b/dpnp/tests/test_special.py @@ -94,6 +94,9 @@ def _check_variant_func( # calling numpy testing func, because it's more verbose assert_allclose(x.asnumpy(), y.asnumpy(), rtol=rtol, atol=atol) + @pytest.mark.usefixtures( + "suppress_overflow_encountered_in_cast_numpy_warnings" + ) def test_erfc(self, inverse): self._check_variant_func( inverse, diff --git a/dpnp/tests/test_sycl_queue.py b/dpnp/tests/test_sycl_queue.py index d1853579036a..699cd81c96f6 100644 --- a/dpnp/tests/test_sycl_queue.py +++ b/dpnp/tests/test_sycl_queue.py @@ -1086,7 +1086,7 @@ def test_array_creation_from_dpctl(copy, device): @pytest.mark.parametrize("arr_dtype", get_all_dtypes(no_float16=True)) @pytest.mark.parametrize("shape", [tuple(), (2,), (3, 0, 1), (2, 2, 2)]) def test_from_dlpack(arr_dtype, shape, device): - X = dpnp.empty(shape=shape, dtype=arr_dtype, device=device) + X = dpnp.ones(shape=shape, dtype=arr_dtype, device=device) Y = dpnp.from_dlpack(X) assert_array_equal(X, Y) assert X.__dlpack_device__() == Y.__dlpack_device__() @@ -1101,7 +1101,7 @@ def test_from_dlpack(arr_dtype, shape, device): @pytest.mark.parametrize("device", valid_dev, ids=dev_ids) @pytest.mark.parametrize("arr_dtype", get_all_dtypes(no_float16=True)) def test_from_dlpack_with_dpt(arr_dtype, device): - X = dpctl.tensor.empty((64,), dtype=arr_dtype, device=device) + X = dpt.ones((64,), dtype=arr_dtype, device=device) Y = dpnp.from_dlpack(X) assert_array_equal(X, Y) assert isinstance(Y, dpnp.dpnp_array.dpnp_array) From 2c864972d7ce576c82b8c8a2393d38c1c0b384f3 Mon Sep 17 00:00:00 2001 From: Abhishek Bagusetty <59661409+abagusetty@users.noreply.github.com> Date: Wed, 4 Mar 2026 13:20:24 -0600 Subject: [PATCH 07/35] Add `scipy.linalg.lu()` decomposition support (#2787) This PR adds `dpnp.scipy.linalg.lu()` with support for all three output modes: default `(P, L, U)`, `permute_l=True (PL, U)`, and `p_indices=True` `(p, L, U)`, including batched inputs. Fixes: https://github.com/IntelPython/dpnp/issues/2786 --- CHANGELOG.md | 1 + dpnp/scipy/linalg/__init__.py | 3 +- dpnp/scipy/linalg/_decomp_lu.py | 153 +++++- dpnp/scipy/linalg/_utils.py | 255 +++++++++- dpnp/tests/test_linalg.py | 450 ++++++++++++++++++ dpnp/tests/test_sycl_queue.py | 12 + dpnp/tests/test_usm_type.py | 12 + .../linalg_tests/test_decomp_lu.py | 15 +- 8 files changed, 886 insertions(+), 15 deletions(-) diff --git a/CHANGELOG.md b/CHANGELOG.md index f177be311f84..61cde1ddfefc 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -26,6 +26,7 @@ Also, that release drops support for Python 3.9, making Python 3.10 the minimum * Added implementation of `dpnp.ndarray.__bytes__` method [#2671](https://github.com/IntelPython/dpnp/pull/2671) * Added implementation of `dpnp.divmod` [#2674](https://github.com/IntelPython/dpnp/pull/2674) * Added implementation of `dpnp.isin` function [#2595](https://github.com/IntelPython/dpnp/pull/2595) +* Added implementation of `dpnp.scipy.linalg.lu` (SciPy-compatible) [#2787](https://github.com/IntelPython/dpnp/pull/2787) ### Changed diff --git a/dpnp/scipy/linalg/__init__.py b/dpnp/scipy/linalg/__init__.py index 3afc08a6fdb9..81eadd890fa9 100644 --- a/dpnp/scipy/linalg/__init__.py +++ b/dpnp/scipy/linalg/__init__.py @@ -35,9 +35,10 @@ """ -from ._decomp_lu import lu_factor, lu_solve +from ._decomp_lu import lu, lu_factor, lu_solve __all__ = [ + "lu", "lu_factor", "lu_solve", ] diff --git a/dpnp/scipy/linalg/_decomp_lu.py b/dpnp/scipy/linalg/_decomp_lu.py index 292d7fffe4b4..823b2fccc230 100644 --- a/dpnp/scipy/linalg/_decomp_lu.py +++ b/dpnp/scipy/linalg/_decomp_lu.py @@ -46,11 +46,154 @@ ) from ._utils import ( + dpnp_lu, dpnp_lu_factor, dpnp_lu_solve, ) +def lu( + a, permute_l=False, overwrite_a=False, check_finite=True, p_indices=False +): + """ + Compute LU decomposition of a matrix with partial pivoting. + + The decomposition satisfies:: + + A = P @ L @ U + + where `P` is a permutation matrix, `L` is lower triangular with unit + diagonal elements, and `U` is upper triangular. If `permute_l` is set to + ``True`` then `L` is returned already permuted and hence satisfying + ``A = L @ U``. + + For full documentation refer to :obj:`scipy.linalg.lu`. + + Parameters + ---------- + a : (..., M, N) {dpnp.ndarray, usm_ndarray} + Input array to decompose. + permute_l : bool, optional + Perform the multiplication ``P @ L`` (Default: do not permute). + + Default: ``False``. + overwrite_a : {None, bool}, optional + Whether to overwrite data in `a` (may increase performance). + + Default: ``False``. + check_finite : {None, bool}, optional + Whether to check that the input matrix contains only finite numbers. + Disabling may give a performance gain, but may result in problems + (crashes, non-termination) if the inputs do contain infinities or NaNs. + + Default: ``True``. + p_indices : bool, optional + If ``True`` the permutation information is returned as row indices + instead of a permutation matrix. + + Default: ``False``. + + Returns + ------- + **(If ``permute_l`` is ``False``)** + + p : (..., M, M) dpnp.ndarray or (..., M) dpnp.ndarray + If `p_indices` is ``False`` (default), the permutation matrix. + The permutation matrix always has a real dtype (``float32`` or + ``float64``) even when `a` is complex, since it only contains + 0s and 1s. + If `p_indices` is ``True``, a 1-D (or batched) array of row + permutation indices such that ``A = L[p] @ U``. + l : (..., M, K) dpnp.ndarray + Lower triangular or trapezoidal matrix with unit diagonal. + ``K = min(M, N)``. + u : (..., K, N) dpnp.ndarray + Upper triangular or trapezoidal matrix. + + **(If ``permute_l`` is ``True``)** + + pl : (..., M, K) dpnp.ndarray + Permuted ``L`` matrix: ``pl = P @ L``. + ``K = min(M, N)``. + u : (..., K, N) dpnp.ndarray + Upper triangular or trapezoidal matrix. + + Notes + ----- + Permutation matrices are costly since they are nothing but row reorder of + ``L`` and hence indices are strongly recommended to be used instead if the + permutation is required. The relation in the 2D case then becomes simply + ``A = L[P, :] @ U``. In higher dimensions, it is better to use `permute_l` + to avoid complicated indexing tricks. + + In the 2D case, if one has the indices however, for some reason, the + permutation matrix is still needed then it can be constructed by + ``dpnp.eye(M)[P, :]``. + + Warning + ------- + This function synchronizes in order to validate array elements + when ``check_finite=True``, and also synchronizes to compute the + permutation from LAPACK pivot indices. + + See Also + -------- + :obj:`dpnp.scipy.linalg.lu_factor` : LU factorize a matrix + (compact representation). + :obj:`dpnp.scipy.linalg.lu_solve` : Solve an equation system using + the LU factorization of a matrix. + + Examples + -------- + >>> import dpnp as np + >>> A = np.array([[2, 5, 8, 7], [5, 2, 2, 8], + ... [7, 5, 6, 6], [5, 4, 4, 8]]) + >>> p, l, u = np.scipy.linalg.lu(A) + >>> np.allclose(A, p @ l @ u) + array(True) + + Retrieve the permutation as row indices with ``p_indices=True``: + + >>> p, l, u = np.scipy.linalg.lu(A, p_indices=True) + >>> p + array([1, 3, 0, 2]) + >>> np.allclose(A, l[p] @ u) + array(True) + + Return the permuted ``L`` directly with ``permute_l=True``: + + >>> pl, u = np.scipy.linalg.lu(A, permute_l=True) + >>> np.allclose(A, pl @ u) + array(True) + + Non-square matrices are supported: + + >>> B = np.array([[1, 2, 3], [4, 5, 6]]) + >>> p, l, u = np.scipy.linalg.lu(B) + >>> np.allclose(B, p @ l @ u) + array(True) + + Batched input: + + >>> C = np.random.randn(3, 2, 4, 4) + >>> p, l, u = np.scipy.linalg.lu(C) + >>> np.allclose(C, p @ l @ u) + array(True) + + """ + + dpnp.check_supported_arrays_type(a) + assert_stacked_2d(a) + + return dpnp_lu( + a, + overwrite_a=overwrite_a, + check_finite=check_finite, + p_indices=p_indices, + permute_l=permute_l, + ) + + def lu_factor(a, overwrite_a=False, check_finite=True): """ Compute the pivoted LU decomposition of `a` matrix. @@ -180,13 +323,13 @@ def lu_solve(lu_and_piv, b, trans=0, overwrite_b=False, check_finite=True): """ - lu, piv = lu_and_piv - dpnp.check_supported_arrays_type(lu, piv, b) - assert_stacked_2d(lu) - assert_stacked_square(lu) + lu_matrix, piv = lu_and_piv + dpnp.check_supported_arrays_type(lu_matrix, piv, b) + assert_stacked_2d(lu_matrix) + assert_stacked_square(lu_matrix) return dpnp_lu_solve( - lu, + lu_matrix, piv, b, trans=trans, diff --git a/dpnp/scipy/linalg/_utils.py b/dpnp/scipy/linalg/_utils.py index f00db6fdfb92..d083f1c2c0a2 100644 --- a/dpnp/scipy/linalg/_utils.py +++ b/dpnp/scipy/linalg/_utils.py @@ -49,7 +49,7 @@ import dpnp import dpnp.backend.extensions.lapack._lapack_impl as li from dpnp.dpnp_utils import get_usm_allocations -from dpnp.linalg.dpnp_utils_linalg import _common_type +from dpnp.linalg.dpnp_utils_linalg import _common_type, _real_type def _align_lu_solve_broadcast(lu, b): @@ -83,6 +83,48 @@ def _align_lu_solve_broadcast(lu, b): return lu, b +def _apply_permutation_to_rows(mat, perm_indices): + """ + Apply a permutation to the rows (axis=-2) of a matrix. + + Returns ``out`` such that + ``out[..., i, :] = mat[..., perm_indices[..., i], :]``. + + For 2-D inputs this is equivalent to ``mat[perm_indices]`` (a single + device gather). For batched inputs :func:`dpnp.take_along_axis` is + used so the operation stays entirely on the device. + + Parameters + ---------- + mat : dpnp.ndarray, shape (..., M, N) + Matrix whose rows are to be permuted. + perm_indices : dpnp.ndarray, shape (..., M) + Permutation indices (dtype int64). + + Returns + ------- + out : dpnp.ndarray, shape (..., M, N) + Row-permuted matrix. + """ + + if perm_indices.ndim == 1: + # 2-D case: simple fancy indexing, single kernel launch. + return mat[perm_indices] + + # Batched case: ensure *mat* has the same batch dimensions as + # *perm_indices*. This is needed, for example, when permuting + # a shared identity matrix across a batch. + target_shape = perm_indices.shape[:-1] + mat.shape[-2:] + if mat.shape != target_shape: + mat = dpnp.broadcast_to(mat, target_shape) + + # Expand (..., M) → (..., M, 1), then broadcast to the full shape + # of *mat* so take_along_axis can gather along axis -2. + idx = dpnp.expand_dims(perm_indices, axis=-1) + idx = dpnp.broadcast_to(idx, target_shape).copy() + return dpnp.take_along_axis(mat, idx, axis=-2) + + def _batched_lu_factor_scipy(a, res_type): # pylint: disable=too-many-locals """SciPy-compatible LU factorization for batched inputs.""" @@ -338,6 +380,71 @@ def _map_trans_to_mkl(trans): raise ValueError("`trans` must be 0 (N), 1 (T), or 2 (C)") +def _pivots_to_permutation(piv, m): + """ + Convert 0-based LAPACK pivot indices (sequential row swaps) + to a permutation array. + + The returned permutation ``perm`` satisfies ``A[perm] = L @ U`` + (i.e. the forward row-permutation produced by LAPACK). + + The computation is performed entirely on the device. A host-side + Python loop of ``K = min(M, N)`` iterations drives the sequential + swap logic, but each iteration only launches device kernels + (:func:`dpnp.take_along_axis` for gather, + :func:`dpnp.put_along_axis` for scatter); **no data is transferred + between host and device**. + + .. note:: + + A future custom SYCL kernel could fuse all ``K`` swap steps + into a single launch to eliminate per-step kernel overhead. + + Parameters + ---------- + piv : dpnp.ndarray, shape (..., K) + 0-based pivot indices as returned by :obj:`dpnp_lu_factor`. + m : int + Number of rows of the original matrix. + + Returns + ------- + perm : dpnp.ndarray, shape (..., M), dtype int64 + Permutation indices. + """ + + batch_shape = piv.shape[:-1] + k = piv.shape[-1] + + # Initialise the identity permutation on the device. + perm = dpnp.broadcast_to( + dpnp.arange( + m, + dtype=dpnp.int64, + usm_type=piv.usm_type, + sycl_queue=piv.sycl_queue, + ), + (*batch_shape, m), + ).copy() + + # Apply sequential row swaps entirely on the device. + # Each iteration launches a small number of device kernels (gather + + # slice-assign + scatter) but never transfers data to the host. + for i in range(k): + # Pivot target for step *i*: shape (..., 1) + j = piv[..., i : i + 1] + + # Gather the two values to be swapped. + val_i = perm[..., i : i + 1].copy() # slice (free) + val_j = dpnp.take_along_axis(perm, j, axis=-1) # gather + + # Perform the swap. + perm[..., i : i + 1] = val_j # slice assign + dpnp.put_along_axis(perm, j, val_i, axis=-1) # scatter + + return perm + + def dpnp_lu_factor(a, overwrite_a=False, check_finite=True): """ dpnp_lu_factor(a, overwrite_a=False, check_finite=True) @@ -432,6 +539,152 @@ def dpnp_lu_factor(a, overwrite_a=False, check_finite=True): return (a_h, ipiv_h) +def _assemble_lu_output( + low, + up, + inv_perm, + permute_l, + p_indices, + m, + real_type, + a_usm_type, + a_sycl_queue, +): + """Select and build the correct dpnp_lu return value.""" + if permute_l: + return _apply_permutation_to_rows(low, inv_perm), up + if p_indices: + return inv_perm, low, up + eye_m = dpnp.eye( + m, dtype=real_type, usm_type=a_usm_type, sycl_queue=a_sycl_queue + ) + return ( + _apply_permutation_to_rows(eye_m, inv_perm), + low, + up, + ) # perm_matrix, L, U + + +def dpnp_lu( + a, + overwrite_a=False, + check_finite=True, + p_indices=False, + permute_l=False, +): + """ + dpnp_lu(a, overwrite_a=False, check_finite=True, p_indices=False, + permute_l=False) + + Compute pivoted LU decomposition and return separate P, L, U matrices + (SciPy-compatible behavior). + + This function mimics the behavior of `scipy.linalg.lu` including + support for `permute_l`, `p_indices`, `overwrite_a`, and `check_finite`. + + """ + + a_sycl_queue = a.sycl_queue + a_usm_type = a.usm_type + m, n = a.shape[-2:] + k = min(m, n) + batch_shape = a.shape[:-2] + + res_type = _common_type(a) + + # The permutation matrix P uses a real dtype (SciPy convention): + # P only contains 0s and 1s, so complex storage would be wasteful. + real_type = _real_type(res_type) + + # ---- Fast path: scalar (1x1) matrices ---- + # For 1x1 input, P = I, L = I, U = A. This avoids invoking LAPACK + # entirely (matches SciPy's scalar fast path). + if m == 1 and n == 1: + if check_finite: + if not dpnp.isfinite(a).all(): + raise ValueError("array must not contain infs or NaNs") + + low = dpnp.ones_like(a, dtype=res_type) + up = dpnp.astype(a, res_type, copy=not overwrite_a) + inv_perm = dpnp.zeros_like(a, shape=(*batch_shape, 1), dtype=dpnp.int64) + + return _assemble_lu_output( + low, + up, + inv_perm, + permute_l, + p_indices, + m, + real_type, + a_usm_type, + a_sycl_queue, + ) + + # ---- Fast path: empty arrays ---- + if a.size == 0: + low = dpnp.empty_like(a, shape=(*batch_shape, m, k), dtype=res_type) + up = dpnp.empty_like(a, shape=(*batch_shape, k, n), dtype=res_type) + inv_perm = dpnp.empty_like(a, shape=(*batch_shape, m), dtype=dpnp.int64) + return _assemble_lu_output( + low, + up, + inv_perm, + permute_l, + p_indices, + m, + real_type, + a_usm_type, + a_sycl_queue, + ) + + # ---- General case: LAPACK factorization ---- + lu_compact, piv = dpnp_lu_factor( + a, overwrite_a=overwrite_a, check_finite=check_finite + ) + + # ---- Extract L: lower-triangular with unit diagonal ---- + # L has shape (..., M, K). + low = dpnp.tril(lu_compact[..., :, :k], k=-1) + low += dpnp.eye( + m, + k, + dtype=lu_compact.dtype, + usm_type=a_usm_type, + sycl_queue=a_sycl_queue, + ) + + # ---- Extract U: upper-triangular ---- + # U has shape (..., K, N). + up = dpnp.triu(lu_compact[..., :k, :]) + + # ---- Convert pivot indices → row permutation ---- + # ``perm`` (forward): A[perm] = L @ U. + # This is the only step that requires a host transfer because the + # sequential swap semantics of LAPACK pivots cannot be parallelised. + # Only the small pivot array (min(M, N) elements per slice) is + # transferred; all subsequent work stays on the device. + perm = _pivots_to_permutation(piv, m) + + # ``inv_perm`` (inverse): A = L[inv_perm] @ U. + # This is SciPy's ``p_indices`` convention. + # ``dpnp.argsort`` is an efficient on-device O(M log M) operation + # that avoids a second host round-trip. + inv_perm = dpnp.argsort(perm, axis=-1).astype(dpnp.int64) + + # ---- Assemble output (SciPy convention) ---- + return _assemble_lu_output( + low, + up, + inv_perm, + permute_l, + p_indices, + m, + real_type, + a_usm_type, + a_sycl_queue, + ) + + def dpnp_lu_solve(lu, piv, b, trans=0, overwrite_b=False, check_finite=True): """ dpnp_lu_solve(lu, piv, b, trans=0, overwrite_b=False, check_finite=True) diff --git a/dpnp/tests/test_linalg.py b/dpnp/tests/test_linalg.py index 31d99d71ce49..7d8018fa83a2 100644 --- a/dpnp/tests/test_linalg.py +++ b/dpnp/tests/test_linalg.py @@ -2605,6 +2605,456 @@ def test_invalid_shapes(self, a_shape, b_shape): dpnp.scipy.linalg.lu_solve((lu, piv), b, check_finite=False) +class TestLu: + @staticmethod + def _make_nonsingular_np(shape, dtype, order): + A = generate_random_numpy_array(shape, dtype, order) + m, n = shape + k = min(m, n) + for i in range(k): + off = numpy.sum(numpy.abs(A[i, :n])) - numpy.abs(A[i, i]) + A[i, i] = A.dtype.type(off + 1.0) + return A + + @pytest.mark.parametrize( + "shape", + [(1, 1), (2, 2), (3, 3), (1, 5), (5, 1), (2, 5), (5, 2)], + ) + @pytest.mark.parametrize("order", ["C", "F"]) + @pytest.mark.parametrize( + "dtype", get_all_dtypes(no_none=True, no_bool=True) + ) + def test_lu_default(self, shape, order, dtype): + a_np = self._make_nonsingular_np(shape, dtype, order) + a_dp = dpnp.array(a_np, order=order) + + P, L, U = dpnp.scipy.linalg.lu(a_dp) + + m, n = shape + k = min(m, n) + assert P.shape == (m, m) + assert L.shape == (m, k) + assert U.shape == (k, n) + + A_cast = a_dp.astype(L.dtype, copy=False) + A_rec = P @ L @ U + assert dpnp.allclose(A_rec, A_cast, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize( + "shape", + [(1, 1), (2, 2), (3, 3), (1, 5), (5, 1), (2, 5), (5, 2)], + ) + @pytest.mark.parametrize("order", ["C", "F"]) + @pytest.mark.parametrize( + "dtype", get_all_dtypes(no_none=True, no_bool=True) + ) + def test_lu_permute_l(self, shape, order, dtype): + a_np = self._make_nonsingular_np(shape, dtype, order) + a_dp = dpnp.array(a_np, order=order) + + PL, U = dpnp.scipy.linalg.lu(a_dp, permute_l=True) + + m, n = shape + k = min(m, n) + assert PL.shape == (m, k) + assert U.shape == (k, n) + + A_cast = a_dp.astype(PL.dtype, copy=False) + A_rec = PL @ U + assert dpnp.allclose(A_rec, A_cast, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize( + "shape", + [(1, 1), (2, 2), (3, 3), (1, 5), (5, 1), (2, 5), (5, 2)], + ) + @pytest.mark.parametrize("order", ["C", "F"]) + @pytest.mark.parametrize( + "dtype", get_all_dtypes(no_none=True, no_bool=True) + ) + def test_lu_p_indices(self, shape, order, dtype): + a_np = self._make_nonsingular_np(shape, dtype, order) + a_dp = dpnp.array(a_np, order=order) + + p, L, U = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + + m, n = shape + k = min(m, n) + assert p.shape == (m,) + assert L.shape == (m, k) + assert U.shape == (k, n) + assert dpnp.issubdtype(p.dtype, dpnp.integer) + + A_rec = L[p] @ U + A_cast = a_dp.astype(L.dtype, copy=False) + assert dpnp.allclose(A_rec, A_cast, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize("in_dtype", get_float_complex_dtypes()) + def test_p_matrix_dtype(self, in_dtype): + expected_p_dtype = numpy.dtype(in_dtype).char.lower() + + a_np = self._make_nonsingular_np((4, 4), in_dtype, "F") + a_dp = dpnp.array(a_np, order="F") + P, L, U = dpnp.scipy.linalg.lu(a_dp) + + assert P.dtype == expected_p_dtype + assert dpnp.issubdtype(P.dtype, dpnp.floating) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_p_indices_dtype(self, dtype): + a_np = self._make_nonsingular_np((4, 4), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + p, _, _ = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + assert dpnp.issubdtype(p.dtype, dpnp.integer) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_l_structure(self, dtype): + a_np = self._make_nonsingular_np((5, 5), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + _, L, _ = dpnp.scipy.linalg.lu(a_dp) + L_np = dpnp.asnumpy(L) + + # unit diagonal + diag_abs = numpy.abs(numpy.diag(L_np)) + assert_allclose(diag_abs, numpy.ones(5, dtype=diag_abs.dtype)) + # lower triangular + assert_allclose(numpy.triu(L_np, 1), numpy.zeros_like(L_np)) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_u_upper_triangular(self, dtype): + a_np = self._make_nonsingular_np((5, 5), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + _, _, U = dpnp.scipy.linalg.lu(a_dp) + U_np = dpnp.asnumpy(U) + assert_allclose(numpy.tril(U_np, -1), numpy.zeros_like(U_np)) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_p_is_permutation(self, dtype): + a_np = self._make_nonsingular_np((5, 5), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + P, _, _ = dpnp.scipy.linalg.lu(a_dp) + P_np = dpnp.asnumpy(P) + + assert_allclose(P_np.sum(axis=0), numpy.ones(5, dtype=P_np.dtype)) + assert_allclose(P_np.sum(axis=1), numpy.ones(5, dtype=P_np.dtype)) + assert_allclose( + P_np.T @ P_np, numpy.eye(5, dtype=P_np.dtype), atol=1e-15 + ) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_modes_consistency(self, dtype): + a_np = self._make_nonsingular_np((5, 5), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + + P, L, U = dpnp.scipy.linalg.lu(a_dp) + PL, U2 = dpnp.scipy.linalg.lu(a_dp, permute_l=True) + p, L3, U3 = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + + A_cast = a_dp.astype(L.dtype, copy=False) + A1 = P @ L @ U + A2 = PL @ U2 + p_np = dpnp.asnumpy(p) + A3_np = dpnp.asnumpy(L3)[p_np] @ dpnp.asnumpy(U3) + + assert dpnp.allclose(A1, A_cast, rtol=1e-6, atol=1e-6) + assert dpnp.allclose(A2, A_cast, rtol=1e-6, atol=1e-6) + assert_allclose(A3_np, dpnp.asnumpy(A_cast), rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_p_times_l_equals_pl(self, dtype): + a_np = self._make_nonsingular_np((5, 5), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + P, L, _ = dpnp.scipy.linalg.lu(a_dp) + PL, _ = dpnp.scipy.linalg.lu(a_dp, permute_l=True) + assert dpnp.allclose(P @ L, PL, rtol=1e-12, atol=1e-12) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_p_indices_to_matrix(self, dtype): + a_np = self._make_nonsingular_np((5, 5), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + P, _, _ = dpnp.scipy.linalg.lu(a_dp) + p, _, _ = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + P_from_idx = dpnp.eye(5, dtype=P.dtype)[p] + assert dpnp.allclose(P_from_idx, P, rtol=1e-15, atol=1e-15) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_overwrite_a_false(self, dtype): + a_dp = dpnp.array([[4, 3], [6, 3]], dtype=dtype, order="F") + a_dp_orig = a_dp.copy() + dpnp.scipy.linalg.lu(a_dp, overwrite_a=False) + assert dpnp.allclose(a_dp, a_dp_orig) + + @pytest.mark.parametrize("shape", [(0, 0), (0, 2), (2, 0)]) + def test_empty_inputs(self, shape): + a_dp = dpnp.empty(shape, dtype=dpnp.default_float_type(), order="F") + P, L, U = dpnp.scipy.linalg.lu(a_dp) + m, n = shape + k = min(m, n) + assert P.shape == (m, m) + assert L.shape == (m, k) + assert U.shape == (k, n) + + @pytest.mark.parametrize("shape", [(0, 0), (0, 2), (2, 0)]) + def test_empty_permute_l(self, shape): + a_dp = dpnp.empty(shape, dtype=dpnp.default_float_type(), order="F") + PL, U = dpnp.scipy.linalg.lu(a_dp, permute_l=True) + m, n = shape + k = min(m, n) + assert PL.shape == (m, k) + assert U.shape == (k, n) + + @pytest.mark.parametrize("shape", [(0, 0), (0, 2), (2, 0)]) + def test_empty_p_indices(self, shape): + a_dp = dpnp.empty(shape, dtype=dpnp.default_float_type(), order="F") + p, L, U = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + m, n = shape + k = min(m, n) + assert p.shape == (m,) + assert L.shape == (m, k) + assert U.shape == (k, n) + + @pytest.mark.parametrize( + "sl", + [ + (slice(None, None, 2), slice(None, None, 2)), + (slice(None, None, -1), slice(None, None, -1)), + ], + ) + def test_strided(self, sl): + base = self._make_nonsingular_np((7, 7), dpnp.default_float_type(), "F") + a_np = base[sl] + a_dp = dpnp.array(a_np) + + P, L, U = dpnp.scipy.linalg.lu(a_dp) + A_rec = P @ L @ U + assert dpnp.allclose(A_rec, a_dp, rtol=1e-6, atol=1e-6) + + @pytest.mark.filterwarnings("ignore::RuntimeWarning") + def test_singular_matrix(self): + a_np = numpy.array([[1.0, 2.0], [2.0, 4.0]]) + a_dp = dpnp.array(a_np) + P, L, U = dpnp.scipy.linalg.lu(a_dp) + A_rec = dpnp.asnumpy(P @ L @ U) + assert_allclose(A_rec, a_np, atol=1e-12) + + def test_identity_matrix(self): + n = 4 + I_dp = dpnp.eye(n, dtype=dpnp.default_float_type()) + P, L, U = dpnp.scipy.linalg.lu(I_dp) + I_np = numpy.eye(n) + assert_allclose(dpnp.asnumpy(P), I_np, atol=1e-15) + assert_allclose(dpnp.asnumpy(L), I_np, atol=1e-15) + assert_allclose(dpnp.asnumpy(U), I_np, atol=1e-15) + + def test_1d_input_raises(self): + a_dp = dpnp.array([1.0, 2.0, 3.0]) + with pytest.raises(ValueError): + dpnp.scipy.linalg.lu(a_dp) + + @pytest.mark.parametrize("bad", [numpy.inf, -numpy.inf, numpy.nan]) + def test_check_finite_raises(self, bad): + a_dp = dpnp.array([[1.0, 2.0], [3.0, bad]], order="F") + assert_raises(ValueError, dpnp.scipy.linalg.lu, a_dp, check_finite=True) + + @pytest.mark.parametrize("bad", [numpy.inf, -numpy.inf, numpy.nan]) + def test_check_finite_raises_scalar(self, bad): + # Covers the 1x1 scalar fast path in dpnp_lu + a_dp = dpnp.array([[bad]]) + assert_raises(ValueError, dpnp.scipy.linalg.lu, a_dp, check_finite=True) + + def test_check_finite_disabled(self): + a_dp = dpnp.array([[1.0, numpy.nan], [3.0, 4.0]]) + result = dpnp.scipy.linalg.lu(a_dp, check_finite=False) + assert len(result) == 3 + + +class TestLuBatched: + @staticmethod + def _make_nonsingular_nd_np(shape, dtype, order): + A = generate_random_numpy_array(shape, dtype, order) + m, n = shape[-2], shape[-1] + k = min(m, n) + A3 = A.reshape((-1, m, n)) + for B in A3: + for i in range(k): + off = numpy.sum(numpy.abs(B[i, :n])) - numpy.abs(B[i, i]) + B[i, i] = A.dtype.type(off + 1.0) + A = A3.reshape(shape) + A = numpy.array(A, order=order) + return A + + @staticmethod + def _reconstruct_p_indices(p, L, U): + """Reconstruct A from (p, L, U) for batched p_indices mode.""" + idx = dpnp.expand_dims(p, axis=-1) + idx = dpnp.broadcast_to(idx, L.shape).copy() + PL = dpnp.take_along_axis(L, idx, axis=-2) + return PL @ U + + @pytest.mark.parametrize( + "shape", + [(2, 2, 2), (3, 4, 4), (2, 3, 5, 2), (4, 1, 3)], + ids=["(2,2,2)", "(3,4,4)", "(2,3,5,2)", "(4,1,3)"], + ) + @pytest.mark.parametrize("order", ["C", "F"]) + @pytest.mark.parametrize("dtype", get_all_dtypes(no_bool=True)) + def test_lu_default_batched(self, shape, order, dtype): + a_np = self._make_nonsingular_nd_np(shape, dtype, order) + a_dp = dpnp.array(a_np, order=order) + + P, L, U = dpnp.scipy.linalg.lu(a_dp) + + m, n = shape[-2], shape[-1] + k = min(m, n) + assert P.shape == (*shape[:-2], m, m) + assert L.shape == (*shape[:-2], m, k) + assert U.shape == (*shape[:-2], k, n) + + A_cast = a_dp.astype(L.dtype, copy=False) + A_rec = P @ L @ U + assert dpnp.allclose(A_rec, A_cast, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize( + "shape", + [(2, 2, 2), (3, 4, 4), (2, 3, 5, 2), (4, 1, 3)], + ids=["(2,2,2)", "(3,4,4)", "(2,3,5,2)", "(4,1,3)"], + ) + @pytest.mark.parametrize("order", ["C", "F"]) + @pytest.mark.parametrize("dtype", get_all_dtypes(no_bool=True)) + def test_lu_permute_l_batched(self, shape, order, dtype): + a_np = self._make_nonsingular_nd_np(shape, dtype, order) + a_dp = dpnp.array(a_np, order=order) + + PL, U = dpnp.scipy.linalg.lu(a_dp, permute_l=True) + + m, n = shape[-2], shape[-1] + k = min(m, n) + assert PL.shape == (*shape[:-2], m, k) + assert U.shape == (*shape[:-2], k, n) + + A_cast = a_dp.astype(PL.dtype, copy=False) + A_rec = PL @ U + assert dpnp.allclose(A_rec, A_cast, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize( + "shape", + [(2, 2, 2), (3, 4, 4), (2, 3, 5, 2), (4, 1, 3)], + ids=["(2,2,2)", "(3,4,4)", "(2,3,5,2)", "(4,1,3)"], + ) + @pytest.mark.parametrize("order", ["C", "F"]) + @pytest.mark.parametrize("dtype", get_all_dtypes(no_bool=True)) + def test_lu_p_indices_batched(self, shape, order, dtype): + a_np = self._make_nonsingular_nd_np(shape, dtype, order) + a_dp = dpnp.array(a_np, order=order) + + p, L, U = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + + m, n = shape[-2], shape[-1] + k = min(m, n) + assert p.shape == (*shape[:-2], m) + assert L.shape == (*shape[:-2], m, k) + assert U.shape == (*shape[:-2], k, n) + assert dpnp.issubdtype(p.dtype, dpnp.integer) + + A_cast = a_dp.astype(L.dtype, copy=False) + A_rec = self._reconstruct_p_indices(p, L, U) + assert dpnp.allclose(A_rec, A_cast, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + @pytest.mark.parametrize("order", ["C", "F"]) + def test_overwrite_a(self, dtype, order): + a_np = self._make_nonsingular_nd_np((3, 2, 2), dtype, order) + a_dp = dpnp.array(a_np, order=order) + a_dp_orig = a_dp.copy() + + dpnp.scipy.linalg.lu(a_dp, overwrite_a=False) + assert dpnp.allclose(a_dp, a_dp_orig) + + @pytest.mark.parametrize("dtype", get_float_complex_dtypes()) + def test_modes_consistency_batched(self, dtype): + a_np = self._make_nonsingular_nd_np((3, 4, 4), dtype, "F") + a_dp = dpnp.array(a_np, order="F") + + P, L, U = dpnp.scipy.linalg.lu(a_dp) + PL, U2 = dpnp.scipy.linalg.lu(a_dp, permute_l=True) + p, L3, U3 = dpnp.scipy.linalg.lu(a_dp, p_indices=True) + + A1 = P @ L @ U + A2 = PL @ U2 + A3 = self._reconstruct_p_indices(p, L3, U3) + + A_cast2 = a_dp.astype(L.dtype, copy=False) + assert dpnp.allclose(A1, A_cast2, rtol=1e-6, atol=1e-6) + assert dpnp.allclose(A2, A_cast2, rtol=1e-6, atol=1e-6) + assert dpnp.allclose(A3, A_cast2, rtol=1e-6, atol=1e-6) + + @pytest.mark.parametrize( + "shape", [(0, 2, 2), (2, 0, 2), (2, 2, 0), (0, 0, 0)] + ) + def test_empty_inputs(self, shape): + a = dpnp.empty(shape, dtype=dpnp.default_float_type(), order="F") + + P, L, U = dpnp.scipy.linalg.lu(a) + m, n = shape[-2:] + k = min(m, n) + assert P.shape == (*shape[:-2], m, m) + assert L.shape == (*shape[:-2], m, k) + assert U.shape == (*shape[:-2], k, n) + + @pytest.mark.parametrize( + "shape", [(0, 2, 2), (2, 0, 2), (2, 2, 0), (0, 0, 0)] + ) + def test_empty_permute_l(self, shape): + a = dpnp.empty(shape, dtype=dpnp.default_float_type(), order="F") + + PL, U = dpnp.scipy.linalg.lu(a, permute_l=True) + m, n = shape[-2:] + k = min(m, n) + assert PL.shape == (*shape[:-2], m, k) + assert U.shape == (*shape[:-2], k, n) + + @pytest.mark.parametrize( + "shape", [(0, 2, 2), (2, 0, 2), (2, 2, 0), (0, 0, 0)] + ) + def test_empty_p_indices(self, shape): + a = dpnp.empty(shape, dtype=dpnp.default_float_type(), order="F") + + p, L, U = dpnp.scipy.linalg.lu(a, p_indices=True) + m, n = shape[-2:] + k = min(m, n) + assert p.shape == (*shape[:-2], m) + assert L.shape == (*shape[:-2], m, k) + assert U.shape == (*shape[:-2], k, n) + + def test_strided(self): + a_np = self._make_nonsingular_nd_np( + (5, 3, 3), dpnp.default_float_type(), "F" + ) + a_dp = dpnp.array(a_np, order="F") + a_stride = a_dp[::2] + + P, L, U = dpnp.scipy.linalg.lu(a_stride) + for i in range(a_stride.shape[0]): + A_rec = dpnp.asnumpy(P[i] @ L[i] @ U[i]) + A_orig = dpnp.asnumpy(a_stride[i].astype(L.dtype, copy=False)) + assert_allclose(A_rec, A_orig, rtol=1e-6, atol=1e-6) + + @pytest.mark.filterwarnings("ignore::RuntimeWarning") + def test_singular_matrix(self): + a = dpnp.zeros((3, 2, 2), dtype=dpnp.default_float_type()) + a[0] = dpnp.array([[1.0, 2.0], [2.0, 4.0]]) + a[1] = dpnp.eye(2) + a[2] = dpnp.array([[1.0, 1.0], [1.0, 1.0]]) + + P, L, U = dpnp.scipy.linalg.lu(a) + A_rec = P @ L @ U + assert dpnp.allclose(A_rec, a, rtol=1e-6, atol=1e-6) + + def test_check_finite_raises(self): + a = dpnp.ones((2, 3, 3), dtype=dpnp.default_float_type(), order="F") + a[1, 0, 0] = dpnp.nan + assert_raises(ValueError, dpnp.scipy.linalg.lu, a, check_finite=True) + + class TestMatrixPower: @pytest.mark.parametrize("dtype", get_all_dtypes()) @pytest.mark.parametrize( diff --git a/dpnp/tests/test_sycl_queue.py b/dpnp/tests/test_sycl_queue.py index 699cd81c96f6..560f235bb56f 100644 --- a/dpnp/tests/test_sycl_queue.py +++ b/dpnp/tests/test_sycl_queue.py @@ -1666,6 +1666,18 @@ def test_lu_factor(self, data, device): param_queue = param.sycl_queue assert_sycl_queue_equal(param_queue, a.sycl_queue) + @pytest.mark.parametrize( + "data", + [[[1.0, 2.0], [3.0, 5.0]], [[]], [[[1.0, 2.0], [3.0, 5.0]]], [[[]]]], + ) + def test_lu(self, data, device): + a = dpnp.array(data, device=device) + result = dpnp.scipy.linalg.lu(a) + + for param in result: + param_queue = param.sycl_queue + assert_sycl_queue_equal(param_queue, a.sycl_queue) + @pytest.mark.parametrize( "a_data, b_data", [ diff --git a/dpnp/tests/test_usm_type.py b/dpnp/tests/test_usm_type.py index 4fc0f2b958fa..b73eb67d51ee 100644 --- a/dpnp/tests/test_usm_type.py +++ b/dpnp/tests/test_usm_type.py @@ -1527,6 +1527,18 @@ def test_lstsq(self, m, n, nrhs, usm_type, usm_type_other): [usm_type, usm_type_other] ) + @pytest.mark.parametrize( + "data", + [[[1.0, 2.0], [3.0, 5.0]], [[]], [[[1.0, 2.0], [3.0, 5.0]]], [[[]]]], + ) + def test_lu(self, data, usm_type): + a = dpnp.array(data, usm_type=usm_type) + result = dpnp.scipy.linalg.lu(a) + + assert a.usm_type == usm_type + for param in result: + assert param.usm_type == a.usm_type + @pytest.mark.parametrize( "data", [[[1.0, 2.0], [3.0, 5.0]], [[]], [[[1.0, 2.0], [3.0, 5.0]]], [[[]]]], diff --git a/dpnp/tests/third_party/cupyx/scipy_tests/linalg_tests/test_decomp_lu.py b/dpnp/tests/third_party/cupyx/scipy_tests/linalg_tests/test_decomp_lu.py index fb51c3e39244..440521419652 100644 --- a/dpnp/tests/third_party/cupyx/scipy_tests/linalg_tests/test_decomp_lu.py +++ b/dpnp/tests/third_party/cupyx/scipy_tests/linalg_tests/test_decomp_lu.py @@ -124,7 +124,6 @@ def test_lu_factor_reconstruction_singular(self, dtype): ) @testing.fix_random() @testing.with_requires("scipy") -@pytest.mark.skip("lu() is not supported yet") class TestLU(unittest.TestCase): @testing.for_dtypes("fdFD") @@ -132,7 +131,7 @@ def test_lu(self, dtype): a_cpu = testing.shaped_random(self.shape, numpy, dtype=dtype) a_gpu = cupy.asarray(a_cpu) result_cpu = scipy.linalg.lu(a_cpu, permute_l=self.permute_l) - result_gpu = cupy.linalg.lu(a_gpu, permute_l=self.permute_l) + result_gpu = cupy.scipy.linalg.lu(a_gpu, permute_l=self.permute_l) assert len(result_cpu) == len(result_gpu) if not self.permute_l: # check permutation matrix @@ -140,22 +139,22 @@ def test_lu(self, dtype): result_gpu = list(result_gpu) P_cpu = result_cpu.pop(0) P_gpu = result_gpu.pop(0) - cupy.testing.assert_array_equal(P_gpu, P_cpu) - cupy.testing.assert_allclose(result_gpu[0], result_cpu[0], atol=1e-5) - cupy.testing.assert_allclose(result_gpu[1], result_cpu[1], atol=1e-5) + testing.assert_array_equal(P_gpu, P_cpu) + testing.assert_allclose(result_gpu[0], result_cpu[0], atol=1e-5) + testing.assert_allclose(result_gpu[1], result_cpu[1], atol=1e-5) @testing.for_dtypes("fdFD") def test_lu_reconstruction(self, dtype): m, n = self.shape A = testing.shaped_random(self.shape, cupy, dtype=dtype) if self.permute_l: - PL, U = cupy.linalg.lu(A, permute_l=self.permute_l) + PL, U = cupy.scipy.linalg.lu(A, permute_l=self.permute_l) PLU = PL @ U else: - P, L, U = cupy.linalg.lu(A, permute_l=self.permute_l) + P, L, U = cupy.scipy.linalg.lu(A, permute_l=self.permute_l) PLU = P @ L @ U # check that reconstruction is close to original - cupy.testing.assert_allclose(PLU, A, atol=1e-5) + testing.assert_allclose(PLU, A, atol=1e-5) @testing.parameterize( From a2825a1f6297da6c0cce05b3fa4bd867fa5ac133 Mon Sep 17 00:00:00 2001 From: Anton <100830759+antonwolfy@users.noreply.github.com> Date: Thu, 5 Mar 2026 13:08:22 +0100 Subject: [PATCH 08/35] Import exceptions directly from `dpnp.exceptions` module (#2798) This PR updates the code to consistently use the `dpnp.exceptions` module as the single source of the exceptions where it's applicable. --- dpnp/dpnp_algo/dpnp_fill.py | 3 ++- dpnp/dpnp_array.py | 2 +- dpnp/dpnp_iface_indexing.py | 5 +++-- dpnp/dpnp_iface_logic.py | 3 ++- dpnp/dpnp_iface_manipulation.py | 2 +- dpnp/dpnp_iface_mathematical.py | 3 ++- dpnp/dpnp_utils/dpnp_utils_einsum.py | 2 +- dpnp/dpnp_utils/dpnp_utils_linearalgebra.py | 3 +-- dpnp/dpnp_utils/dpnp_utils_statistics.py | 2 +- dpnp/fft/dpnp_utils_fft.py | 2 +- dpnp/tests/test_array_api_info.py | 3 ++- dpnp/tests/test_arraymanipulation.py | 2 +- dpnp/tests/test_counting.py | 2 +- dpnp/tests/test_fft.py | 2 +- dpnp/tests/test_fill.py | 2 +- dpnp/tests/test_flipping.py | 2 +- dpnp/tests/test_indexing.py | 3 +-- dpnp/tests/test_linalg.py | 3 +-- dpnp/tests/test_logic.py | 2 +- dpnp/tests/test_manipulation.py | 2 +- dpnp/tests/test_mathematical.py | 7 ++----- dpnp/tests/test_nanfunctions.py | 2 +- dpnp/tests/test_product.py | 3 +-- dpnp/tests/test_sort.py | 2 +- dpnp/tests/test_sycl_queue.py | 2 +- dpnp/tests/third_party/cupy/core_tests/test_ndarray.py | 4 +--- dpnp/tests/third_party/cupy/lib_tests/test_shape_base.py | 2 +- .../tests/third_party/cupy/manipulation_tests/test_dims.py | 2 +- .../tests/third_party/cupy/manipulation_tests/test_join.py | 5 +++-- .../third_party/cupy/manipulation_tests/test_transpose.py | 2 +- dpnp/tests/third_party/cupy/math_tests/test_sumprod.py | 2 +- dpnp/tests/third_party/cupy/sorting_tests/test_sort.py | 2 +- .../third_party/cupy/statistics_tests/test_meanvar.py | 2 +- dpnp/tests/third_party/cupy/testing/_loops.py | 2 +- 34 files changed, 43 insertions(+), 46 deletions(-) diff --git a/dpnp/dpnp_algo/dpnp_fill.py b/dpnp/dpnp_algo/dpnp_fill.py index 112ea3af0fdb..c3bfa8fa2e80 100644 --- a/dpnp/dpnp_algo/dpnp_fill.py +++ b/dpnp/dpnp_algo/dpnp_fill.py @@ -38,6 +38,7 @@ ) import dpnp +from dpnp.exceptions import ExecutionPlacementError def dpnp_fill(arr, val): @@ -50,7 +51,7 @@ def dpnp_fill(arr, val): if val.shape != (): raise ValueError("`val` must be a scalar or 0D-array") if dpu.get_execution_queue((exec_q, val.sycl_queue)) is None: - raise dpu.ExecutionPlacementError( + raise ExecutionPlacementError( "Input arrays have incompatible queues." ) a_val = dpt.astype(val, arr.dtype) diff --git a/dpnp/dpnp_array.py b/dpnp/dpnp_array.py index bb864d4444a9..dad67fc1b584 100644 --- a/dpnp/dpnp_array.py +++ b/dpnp/dpnp_array.py @@ -39,11 +39,11 @@ import dpctl.tensor as dpt import dpctl.tensor._type_utils as dtu -from dpctl.tensor._numpy_helper import AxisError import dpnp from . import memory as dpm +from .exceptions import AxisError def _get_unwrapped_index_key(key): diff --git a/dpnp/dpnp_iface_indexing.py b/dpnp/dpnp_iface_indexing.py index 7718412701e8..db70f1fd2384 100644 --- a/dpnp/dpnp_iface_indexing.py +++ b/dpnp/dpnp_iface_indexing.py @@ -63,6 +63,7 @@ ) from .dpnp_array import dpnp_array from .dpnp_utils import call_origin, get_usm_allocations +from .exceptions import ExecutionPlacementError def _ravel_multi_index_checks(multi_index, dims, order): @@ -129,7 +130,7 @@ def _choose_run(inds, chcs, q, usm_type, out=None, mode=0): ) if dpu.get_execution_queue((q, out.sycl_queue)) is None: - raise dpu.ExecutionPlacementError( + raise ExecutionPlacementError( "Input and output allocation queues are not compatible" ) @@ -291,7 +292,7 @@ def _take_index(x, inds, axis, q, usm_type, out=None, mode=0): ) if dpu.get_execution_queue((q, out.sycl_queue)) is None: - raise dpu.ExecutionPlacementError( + raise ExecutionPlacementError( "Input and output allocation queues are not compatible" ) diff --git a/dpnp/dpnp_iface_logic.py b/dpnp/dpnp_iface_logic.py index 1834f25a0485..6464bd49af1b 100644 --- a/dpnp/dpnp_iface_logic.py +++ b/dpnp/dpnp_iface_logic.py @@ -55,6 +55,7 @@ from .dpnp_array import dpnp_array from .dpnp_utils import get_usm_allocations +from .exceptions import ExecutionPlacementError def _isclose_scalar_tol(a, b, rtol, atol, equal_nan): @@ -1267,7 +1268,7 @@ def isin( ) is None ): - raise dpu.ExecutionPlacementError( + raise ExecutionPlacementError( "Input arrays have incompatible allocation queues" ) usm_element = dpnp.get_usm_ndarray(element) diff --git a/dpnp/dpnp_iface_manipulation.py b/dpnp/dpnp_iface_manipulation.py index dd872485a602..ff7ac85666a1 100644 --- a/dpnp/dpnp_iface_manipulation.py +++ b/dpnp/dpnp_iface_manipulation.py @@ -48,7 +48,6 @@ import dpctl.tensor as dpt import numpy from dpctl.tensor._numpy_helper import ( - AxisError, normalize_axis_index, normalize_axis_tuple, ) @@ -60,6 +59,7 @@ # pylint: disable=no-name-in-module from .dpnp_utils import get_usm_allocations from .dpnp_utils.dpnp_utils_pad import dpnp_pad +from .exceptions import AxisError class InsertDeleteParams(NamedTuple): diff --git a/dpnp/dpnp_iface_mathematical.py b/dpnp/dpnp_iface_mathematical.py index e339c24d384c..366a3363404a 100644 --- a/dpnp/dpnp_iface_mathematical.py +++ b/dpnp/dpnp_iface_mathematical.py @@ -84,6 +84,7 @@ from .dpnp_utils import get_usm_allocations from .dpnp_utils.dpnp_utils_linearalgebra import dpnp_cross from .dpnp_utils.dpnp_utils_reduction import dpnp_wrap_reduction_call +from .exceptions import ExecutionPlacementError def _get_max_min(dtype): @@ -273,7 +274,7 @@ def _process_ediff1d_args(arg, arg_name, ary_dtype, ary_sycl_queue, usm_type): usm_type = dpu.get_coerced_usm_type([usm_type, arg.usm_type]) # check that arrays have the same allocation queue if dpu.get_execution_queue([ary_sycl_queue, arg.sycl_queue]) is None: - raise dpu.ExecutionPlacementError( + raise ExecutionPlacementError( f"ary and {arg_name} must be allocated on the same SYCL queue" ) diff --git a/dpnp/dpnp_utils/dpnp_utils_einsum.py b/dpnp/dpnp_utils/dpnp_utils_einsum.py index 284268e2868b..4a1a58635989 100644 --- a/dpnp/dpnp_utils/dpnp_utils_einsum.py +++ b/dpnp/dpnp_utils/dpnp_utils_einsum.py @@ -33,11 +33,11 @@ import dpctl import numpy -from dpctl.utils import ExecutionPlacementError import dpnp from dpnp.dpnp_array import dpnp_array from dpnp.dpnp_utils import get_usm_allocations, map_dtype_to_device +from dpnp.exceptions import ExecutionPlacementError _einsum_symbols = "abcdefghijklmnopqrstuvwxyzABCDEFGHIJKLMNOPQRSTUVWXYZ" diff --git a/dpnp/dpnp_utils/dpnp_utils_linearalgebra.py b/dpnp/dpnp_utils/dpnp_utils_linearalgebra.py index 191b8aa65d13..d2a1cdfbac46 100644 --- a/dpnp/dpnp_utils/dpnp_utils_linearalgebra.py +++ b/dpnp/dpnp_utils/dpnp_utils_linearalgebra.py @@ -32,16 +32,15 @@ import dpctl.utils as dpu import numpy from dpctl.tensor._numpy_helper import ( - AxisError, normalize_axis_index, normalize_axis_tuple, ) -from dpctl.utils import ExecutionPlacementError import dpnp import dpnp.backend.extensions.blas._blas_impl as bi from dpnp.dpnp_array import dpnp_array from dpnp.dpnp_utils import get_usm_allocations +from dpnp.exceptions import AxisError, ExecutionPlacementError __all__ = [ "dpnp_cross", diff --git a/dpnp/dpnp_utils/dpnp_utils_statistics.py b/dpnp/dpnp_utils/dpnp_utils_statistics.py index 3a3bc04a31af..c8414b661851 100644 --- a/dpnp/dpnp_utils/dpnp_utils_statistics.py +++ b/dpnp/dpnp_utils/dpnp_utils_statistics.py @@ -31,10 +31,10 @@ import dpctl import dpctl.tensor as dpt from dpctl.tensor._numpy_helper import normalize_axis_tuple -from dpctl.utils import ExecutionPlacementError import dpnp from dpnp.dpnp_array import dpnp_array +from dpnp.exceptions import ExecutionPlacementError __all__ = ["dpnp_cov", "dpnp_median"] diff --git a/dpnp/fft/dpnp_utils_fft.py b/dpnp/fft/dpnp_utils_fft.py index 709494e6255e..28032b9d3be2 100644 --- a/dpnp/fft/dpnp_utils_fft.py +++ b/dpnp/fft/dpnp_utils_fft.py @@ -49,10 +49,10 @@ normalize_axis_index, normalize_axis_tuple, ) -from dpctl.utils import ExecutionPlacementError import dpnp import dpnp.backend.extensions.fft._fft_impl as fi +from dpnp.exceptions import ExecutionPlacementError from ..dpnp_array import dpnp_array from ..dpnp_utils import map_dtype_to_device diff --git a/dpnp/tests/test_array_api_info.py b/dpnp/tests/test_array_api_info.py index b310192ffc59..0e2fe7dc5a04 100644 --- a/dpnp/tests/test_array_api_info.py +++ b/dpnp/tests/test_array_api_info.py @@ -1,9 +1,10 @@ import numpy import pytest -from dpctl import SyclDeviceCreationError, get_devices, select_default_device +from dpctl import get_devices, select_default_device from dpctl.tensor._tensor_impl import default_device_complex_type import dpnp +from dpnp.exceptions import SyclDeviceCreationError from dpnp.tests.helper import ( has_support_aspect64, is_win_platform, diff --git a/dpnp/tests/test_arraymanipulation.py b/dpnp/tests/test_arraymanipulation.py index ba83ee94d8b0..fe74368a8c81 100644 --- a/dpnp/tests/test_arraymanipulation.py +++ b/dpnp/tests/test_arraymanipulation.py @@ -3,10 +3,10 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError from numpy.testing import assert_array_equal, assert_equal, assert_raises import dpnp +from dpnp.exceptions import AxisError from .helper import get_all_dtypes, get_float_complex_dtypes from .third_party.cupy import testing diff --git a/dpnp/tests/test_counting.py b/dpnp/tests/test_counting.py index 762abd58b687..821471068fd1 100644 --- a/dpnp/tests/test_counting.py +++ b/dpnp/tests/test_counting.py @@ -1,6 +1,5 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError from numpy.testing import ( assert_allclose, assert_equal, @@ -8,6 +7,7 @@ ) import dpnp +from dpnp.exceptions import AxisError from .helper import ( get_all_dtypes, diff --git a/dpnp/tests/test_fft.py b/dpnp/tests/test_fft.py index 226420057748..b10bf1b46016 100644 --- a/dpnp/tests/test_fft.py +++ b/dpnp/tests/test_fft.py @@ -2,11 +2,11 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.utils import ExecutionPlacementError from numpy.testing import assert_raises import dpnp from dpnp.dpnp_utils import map_dtype_to_device +from dpnp.exceptions import ExecutionPlacementError from .helper import ( assert_dtype_allclose, diff --git a/dpnp/tests/test_fill.py b/dpnp/tests/test_fill.py index 3102de395d93..db53ab976cba 100644 --- a/dpnp/tests/test_fill.py +++ b/dpnp/tests/test_fill.py @@ -1,9 +1,9 @@ import dpctl import pytest -from dpctl.utils import ExecutionPlacementError from numpy.testing import assert_array_equal import dpnp +from dpnp.exceptions import ExecutionPlacementError @pytest.mark.parametrize( diff --git a/dpnp/tests/test_flipping.py b/dpnp/tests/test_flipping.py index cc84242f4557..f48db162f002 100644 --- a/dpnp/tests/test_flipping.py +++ b/dpnp/tests/test_flipping.py @@ -2,12 +2,12 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError from numpy.testing import ( assert_equal, ) import dpnp +from dpnp.exceptions import AxisError from .helper import ( get_all_dtypes, diff --git a/dpnp/tests/test_indexing.py b/dpnp/tests/test_indexing.py index 9a55efe138b7..b6cae0733d40 100644 --- a/dpnp/tests/test_indexing.py +++ b/dpnp/tests/test_indexing.py @@ -4,9 +4,7 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError from dpctl.tensor._type_utils import _to_device_supported_dtype -from dpctl.utils import ExecutionPlacementError from numpy.testing import ( assert_, assert_array_equal, @@ -17,6 +15,7 @@ import dpnp from dpnp.dpnp_array import dpnp_array +from dpnp.exceptions import AxisError, ExecutionPlacementError from .helper import ( get_abs_array, diff --git a/dpnp/tests/test_linalg.py b/dpnp/tests/test_linalg.py index 7d8018fa83a2..170a2a7b5a13 100644 --- a/dpnp/tests/test_linalg.py +++ b/dpnp/tests/test_linalg.py @@ -4,8 +4,6 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError -from dpctl.utils import ExecutionPlacementError from numpy.testing import ( assert_allclose, assert_array_equal, @@ -15,6 +13,7 @@ ) import dpnp +from dpnp.exceptions import AxisError, ExecutionPlacementError from .helper import ( assert_dtype_allclose, diff --git a/dpnp/tests/test_logic.py b/dpnp/tests/test_logic.py index cae51e6777ef..e68ba8162442 100644 --- a/dpnp/tests/test_logic.py +++ b/dpnp/tests/test_logic.py @@ -1,7 +1,6 @@ import dpctl import numpy import pytest -from dpctl.utils import ExecutionPlacementError from numpy.testing import ( assert_allclose, assert_array_equal, @@ -10,6 +9,7 @@ ) import dpnp +from dpnp.exceptions import ExecutionPlacementError from .helper import ( generate_random_numpy_array, diff --git a/dpnp/tests/test_manipulation.py b/dpnp/tests/test_manipulation.py index 8ddba08dbb92..c35050afaa86 100644 --- a/dpnp/tests/test_manipulation.py +++ b/dpnp/tests/test_manipulation.py @@ -3,7 +3,6 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError from numpy.testing import ( assert_array_equal, assert_equal, @@ -11,6 +10,7 @@ ) import dpnp +from dpnp.exceptions import AxisError from .helper import ( assert_dtype_allclose, diff --git a/dpnp/tests/test_mathematical.py b/dpnp/tests/test_mathematical.py index e1f32bbd7931..4bac0e0cc314 100644 --- a/dpnp/tests/test_mathematical.py +++ b/dpnp/tests/test_mathematical.py @@ -2,11 +2,7 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.tensor._numpy_helper import ( - AxisError, - normalize_axis_index, -) -from dpctl.utils import ExecutionPlacementError +from dpctl.tensor._numpy_helper import normalize_axis_index from numpy.testing import ( assert_allclose, assert_array_equal, @@ -18,6 +14,7 @@ import dpnp from dpnp.dpnp_array import dpnp_array from dpnp.dpnp_utils import map_dtype_to_device +from dpnp.exceptions import AxisError, ExecutionPlacementError from .helper import ( LTS_VERSION, diff --git a/dpnp/tests/test_nanfunctions.py b/dpnp/tests/test_nanfunctions.py index d92cee045a72..48520015d354 100644 --- a/dpnp/tests/test_nanfunctions.py +++ b/dpnp/tests/test_nanfunctions.py @@ -2,7 +2,6 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.utils import ExecutionPlacementError from numpy.testing import ( assert_allclose, assert_almost_equal, @@ -13,6 +12,7 @@ ) import dpnp +from dpnp.exceptions import ExecutionPlacementError from .helper import ( assert_dtype_allclose, diff --git a/dpnp/tests/test_product.py b/dpnp/tests/test_product.py index afe767a5e5d9..3ac324b055e8 100644 --- a/dpnp/tests/test_product.py +++ b/dpnp/tests/test_product.py @@ -1,12 +1,11 @@ import dpctl import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError -from dpctl.utils import ExecutionPlacementError from numpy.testing import assert_allclose, assert_array_equal, assert_raises import dpnp from dpnp.dpnp_utils import map_dtype_to_device +from dpnp.exceptions import AxisError, ExecutionPlacementError from .helper import ( assert_dtype_allclose, diff --git a/dpnp/tests/test_sort.py b/dpnp/tests/test_sort.py index 5e883c575f85..27a2afe79b6a 100644 --- a/dpnp/tests/test_sort.py +++ b/dpnp/tests/test_sort.py @@ -1,9 +1,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError from numpy.testing import assert_array_equal, assert_equal, assert_raises import dpnp +from dpnp.exceptions import AxisError from .helper import ( assert_dtype_allclose, diff --git a/dpnp/tests/test_sycl_queue.py b/dpnp/tests/test_sycl_queue.py index 560f235bb56f..b0f746720af8 100644 --- a/dpnp/tests/test_sycl_queue.py +++ b/dpnp/tests/test_sycl_queue.py @@ -5,13 +5,13 @@ import dpctl.tensor as dpt import numpy import pytest -from dpctl.utils import ExecutionPlacementError from numpy.testing import assert_array_equal, assert_raises import dpnp import dpnp.linalg from dpnp.dpnp_array import dpnp_array from dpnp.dpnp_utils import get_usm_allocations +from dpnp.exceptions import ExecutionPlacementError from .helper import ( generate_random_numpy_array, diff --git a/dpnp/tests/third_party/cupy/core_tests/test_ndarray.py b/dpnp/tests/third_party/cupy/core_tests/test_ndarray.py index 95d753c90473..ac6073a3098e 100644 --- a/dpnp/tests/third_party/cupy/core_tests/test_ndarray.py +++ b/dpnp/tests/third_party/cupy/core_tests/test_ndarray.py @@ -6,12 +6,12 @@ import dpctl import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError # from cupy_backends.cuda.api import driver # from cupy_backends.cuda.api import runtime # from cupy_backends.cuda import stream as stream_module import dpnp as cupy +from dpnp.exceptions import AxisError # from cupy import _util # from cupy import _core @@ -19,8 +19,6 @@ # from cupy import get_array_module from dpnp.tests.third_party.cupy import testing -# from cupy.exceptions import AxisError - def get_array_module(*args): for arg in args: diff --git a/dpnp/tests/third_party/cupy/lib_tests/test_shape_base.py b/dpnp/tests/third_party/cupy/lib_tests/test_shape_base.py index c241824fa81d..d6a163906ce9 100644 --- a/dpnp/tests/third_party/cupy/lib_tests/test_shape_base.py +++ b/dpnp/tests/third_party/cupy/lib_tests/test_shape_base.py @@ -2,9 +2,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.helper import has_support_aspect64 from dpnp.tests.third_party.cupy import testing diff --git a/dpnp/tests/third_party/cupy/manipulation_tests/test_dims.py b/dpnp/tests/third_party/cupy/manipulation_tests/test_dims.py index 7355d07e1d9b..ae0f6ce18b47 100644 --- a/dpnp/tests/third_party/cupy/manipulation_tests/test_dims.py +++ b/dpnp/tests/third_party/cupy/manipulation_tests/test_dims.py @@ -2,9 +2,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.third_party.cupy import testing diff --git a/dpnp/tests/third_party/cupy/manipulation_tests/test_join.py b/dpnp/tests/third_party/cupy/manipulation_tests/test_join.py index 0695de034e0c..838bb3646c1e 100644 --- a/dpnp/tests/third_party/cupy/manipulation_tests/test_join.py +++ b/dpnp/tests/third_party/cupy/manipulation_tests/test_join.py @@ -2,11 +2,12 @@ import pytest if numpy.lib.NumpyVersion(numpy.__version__) >= "2.0.0b1": - from numpy.exceptions import AxisError, ComplexWarning + from numpy.exceptions import ComplexWarning else: - from numpy import AxisError, ComplexWarning + from numpy import ComplexWarning import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.helper import has_support_aspect64 from dpnp.tests.third_party.cupy import testing diff --git a/dpnp/tests/third_party/cupy/manipulation_tests/test_transpose.py b/dpnp/tests/third_party/cupy/manipulation_tests/test_transpose.py index 7e7a62dce52a..0a3555fe7798 100644 --- a/dpnp/tests/third_party/cupy/manipulation_tests/test_transpose.py +++ b/dpnp/tests/third_party/cupy/manipulation_tests/test_transpose.py @@ -2,9 +2,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.third_party.cupy import testing diff --git a/dpnp/tests/third_party/cupy/math_tests/test_sumprod.py b/dpnp/tests/third_party/cupy/math_tests/test_sumprod.py index b8f98456a13a..b1c1e569ae2f 100644 --- a/dpnp/tests/third_party/cupy/math_tests/test_sumprod.py +++ b/dpnp/tests/third_party/cupy/math_tests/test_sumprod.py @@ -2,9 +2,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.helper import ( has_support_aspect16, has_support_aspect64, diff --git a/dpnp/tests/third_party/cupy/sorting_tests/test_sort.py b/dpnp/tests/third_party/cupy/sorting_tests/test_sort.py index 7e0eade13254..ba64ef949cb0 100644 --- a/dpnp/tests/third_party/cupy/sorting_tests/test_sort.py +++ b/dpnp/tests/third_party/cupy/sorting_tests/test_sort.py @@ -4,9 +4,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.helper import has_support_aspect64 from dpnp.tests.third_party.cupy import testing diff --git a/dpnp/tests/third_party/cupy/statistics_tests/test_meanvar.py b/dpnp/tests/third_party/cupy/statistics_tests/test_meanvar.py index bf5d37df2fba..2eda8849e819 100644 --- a/dpnp/tests/third_party/cupy/statistics_tests/test_meanvar.py +++ b/dpnp/tests/third_party/cupy/statistics_tests/test_meanvar.py @@ -2,9 +2,9 @@ import numpy import pytest -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests.helper import has_support_aspect16, has_support_aspect64 from dpnp.tests.third_party.cupy import testing diff --git a/dpnp/tests/third_party/cupy/testing/_loops.py b/dpnp/tests/third_party/cupy/testing/_loops.py index 63cd09147c4b..026c451e71e3 100644 --- a/dpnp/tests/third_party/cupy/testing/_loops.py +++ b/dpnp/tests/third_party/cupy/testing/_loops.py @@ -10,9 +10,9 @@ import numpy import pytest from dpctl import select_default_device -from dpctl.tensor._numpy_helper import AxisError import dpnp as cupy +from dpnp.exceptions import AxisError from dpnp.tests import config from dpnp.tests.third_party.cupy.testing import _array, _parameterized from dpnp.tests.third_party.cupy.testing._pytest_impl import is_available From 4b644661997e8df576027c02cd2f62cfb1e6f4d1 Mon Sep 17 00:00:00 2001 From: Anton <100830759+antonwolfy@users.noreply.github.com> Date: Fri, 6 Mar 2026 15:54:40 +0100 Subject: [PATCH 09/35] Update docstrings for `dpnp.scipy.linalg` LU functions (#2802) This PR implements few improvements for LU function in `dpnp.scipy.linalg` namespace: * resolved issue with `Returns` section formatting in `lu` function * changed to proper `func:` role for the functions cross-references in `See also` section * used `Warnings` section name instead of `Warning` * changed `{None, bool}` to `bool` for overwrite_a, check_finite, and overwrite_b parameters --- dpnp/dpnp_iface_histograms.py | 20 ++++++------ dpnp/dpnp_iface_manipulation.py | 8 ++--- dpnp/dpnp_iface_nanfunctions.py | 16 +++++----- dpnp/scipy/linalg/_decomp_lu.py | 56 +++++++++++++++------------------ 4 files changed, 48 insertions(+), 52 deletions(-) diff --git a/dpnp/dpnp_iface_histograms.py b/dpnp/dpnp_iface_histograms.py index 8f3363e79fe0..0a2f18fe3644 100644 --- a/dpnp/dpnp_iface_histograms.py +++ b/dpnp/dpnp_iface_histograms.py @@ -306,8 +306,8 @@ def bincount(x, weights=None, minlength=0): For full documentation refer to :obj:`numpy.bincount`. - Warning - ------- + Warnings + -------- This function synchronizes in order to calculate binning edges. This may harm performance in some applications. @@ -504,8 +504,8 @@ def histogram(a, bins=10, range=None, density=None, weights=None): For full documentation refer to :obj:`numpy.histogram`. - Warning - ------- + Warnings + -------- This function may synchronize in order to check a monotonically increasing array of bin edges. This may harm performance in some applications. @@ -675,8 +675,8 @@ def histogram_bin_edges(a, bins=10, range=None, weights=None): For full documentation refer to :obj:`numpy.histogram_bin_edges`. - Warning - ------- + Warnings + -------- This function may synchronize in order to check a monotonically increasing array of bin edges. This may harm performance in some applications. @@ -767,8 +767,8 @@ def histogram2d(x, y, bins=10, range=None, density=None, weights=None): For full documentation refer to :obj:`numpy.histogram2d`. - Warning - ------- + Warnings + -------- This function may synchronize in order to check a monotonically increasing array of bin edges. This may harm performance in some applications. @@ -1100,8 +1100,8 @@ def histogramdd(sample, bins=10, range=None, density=None, weights=None): For full documentation refer to :obj:`numpy.histogramdd`. - Warning - ------- + Warnings + -------- This function may synchronize in order to check a monotonically increasing array of bin edges. This may harm performance in some applications. diff --git a/dpnp/dpnp_iface_manipulation.py b/dpnp/dpnp_iface_manipulation.py index ff7ac85666a1..0594a406ac5a 100644 --- a/dpnp/dpnp_iface_manipulation.py +++ b/dpnp/dpnp_iface_manipulation.py @@ -829,8 +829,8 @@ def asfarray(a, dtype=None, *, device=None, usm_type=None, sycl_queue=None): out : dpnp.ndarray The input `a` as a float ndarray. - Warning - ------- + Warnings + -------- This function is deprecated in favor of :obj:`dpnp.asarray` and will be removed in a future release. @@ -3099,8 +3099,8 @@ def resize(a, new_shape): be used. In most other cases either indexing (to reduce the size) or padding (to increase the size) may be a more appropriate solution. - Warning - ------- + Warnings + -------- This functionality does **not** consider axes separately, i.e. it does not apply interpolation/extrapolation. It fills the return array with the required number of elements, iterating diff --git a/dpnp/dpnp_iface_nanfunctions.py b/dpnp/dpnp_iface_nanfunctions.py index b8abad2a2088..a5fb750cf586 100644 --- a/dpnp/dpnp_iface_nanfunctions.py +++ b/dpnp/dpnp_iface_nanfunctions.py @@ -122,15 +122,15 @@ def nanargmax(a, axis=None, out=None, *, keepdims=False): For full documentation refer to :obj:`numpy.nanargmax`. - Warning - ------- + Warnings + -------- This function synchronizes in order to test for all-NaN slices in the array. This may harm performance in some applications. To avoid synchronization, the user is recommended to filter NaNs themselves and use `dpnp.argmax` on the filtered array. - Warning - ------- + Warnings + -------- The results cannot be trusted if a slice contains only NaNs and -Infs. @@ -206,15 +206,15 @@ def nanargmin(a, axis=None, out=None, *, keepdims=False): For full documentation refer to :obj:`numpy.nanargmin`. - Warning - ------- + Warnings + -------- This function synchronizes in order to test for all-NaN slices in the array. This may harm performance in some applications. To avoid synchronization, the user is recommended to filter NaNs themselves and use `dpnp.argmax` on the filtered array. - Warning - ------- + Warnings + -------- The results cannot be trusted if a slice contains only NaNs and -Infs. diff --git a/dpnp/scipy/linalg/_decomp_lu.py b/dpnp/scipy/linalg/_decomp_lu.py index 823b2fccc230..f96d56b0e423 100644 --- a/dpnp/scipy/linalg/_decomp_lu.py +++ b/dpnp/scipy/linalg/_decomp_lu.py @@ -77,11 +77,11 @@ def lu( Perform the multiplication ``P @ L`` (Default: do not permute). Default: ``False``. - overwrite_a : {None, bool}, optional + overwrite_a : bool, optional Whether to overwrite data in `a` (may increase performance). Default: ``False``. - check_finite : {None, bool}, optional + check_finite : bool, optional Whether to check that the input matrix contains only finite numbers. Disabling may give a performance gain, but may result in problems (crashes, non-termination) if the inputs do contain infinities or NaNs. @@ -95,23 +95,19 @@ def lu( Returns ------- - **(If ``permute_l`` is ``False``)** + The tuple ``(p, l, u)`` is returned if ``permute_l`` is ``False`` + (default), else the tuple ``(pl, u)`` is returned, where: p : (..., M, M) dpnp.ndarray or (..., M) dpnp.ndarray - If `p_indices` is ``False`` (default), the permutation matrix. - The permutation matrix always has a real dtype (``float32`` or - ``float64``) even when `a` is complex, since it only contains - 0s and 1s. + Permutation matrix or permutation indices. + If `p_indices` is ``False`` (default), a permutation matrix. + The permutation matrix always has a real-valued floating-point dtype + even when `a` is complex, since it only contains 0s and 1s. If `p_indices` is ``True``, a 1-D (or batched) array of row permutation indices such that ``A = L[p] @ U``. l : (..., M, K) dpnp.ndarray Lower triangular or trapezoidal matrix with unit diagonal. ``K = min(M, N)``. - u : (..., K, N) dpnp.ndarray - Upper triangular or trapezoidal matrix. - - **(If ``permute_l`` is ``True``)** - pl : (..., M, K) dpnp.ndarray Permuted ``L`` matrix: ``pl = P @ L``. ``K = min(M, N)``. @@ -130,18 +126,18 @@ def lu( permutation matrix is still needed then it can be constructed by ``dpnp.eye(M)[P, :]``. - Warning - ------- + Warnings + -------- This function synchronizes in order to validate array elements when ``check_finite=True``, and also synchronizes to compute the permutation from LAPACK pivot indices. See Also -------- - :obj:`dpnp.scipy.linalg.lu_factor` : LU factorize a matrix - (compact representation). - :obj:`dpnp.scipy.linalg.lu_solve` : Solve an equation system using - the LU factorization of a matrix. + :func:`dpnp.scipy.linalg.lu_factor` : LU factorize a matrix + (compact representation). + :func:`dpnp.scipy.linalg.lu_solve` : Solve an equation system using + the LU factorization of a matrix. Examples -------- @@ -211,11 +207,11 @@ def lu_factor(a, overwrite_a=False, check_finite=True): ---------- a : (..., M, N) {dpnp.ndarray, usm_ndarray} Input array to decompose. - overwrite_a : {None, bool}, optional + overwrite_a : bool, optional Whether to overwrite data in `a` (may increase performance). Default: ``False``. - check_finite : {None, bool}, optional + check_finite : bool, optional Whether to check that the input matrix contains only finite numbers. Disabling may give a performance gain, but may result in problems (crashes, non-termination) if the inputs do contain infinities or NaNs. @@ -233,15 +229,15 @@ def lu_factor(a, overwrite_a=False, check_finite=True): row i of matrix was interchanged with row piv[i]. Where ``K = min(M, N)``. - Warning - ------- + Warnings + -------- This function synchronizes in order to validate array elements when ``check_finite=True``. See Also -------- - :obj:`dpnp.scipy.linalg.lu_solve` : Solve an equation system using - the LU factorization of `a` matrix. + :func:`dpnp.scipy.linalg.lu_solve` : Solve an equation system using + the LU factorization of `a` matrix. Examples -------- @@ -273,7 +269,7 @@ def lu_solve(lu_and_piv, b, trans=0, overwrite_b=False, check_finite=True): lu, piv : {tuple of dpnp.ndarrays or usm_ndarrays} LU factorization of matrix `a` (..., M, M) together with pivot indices. b : {(M,), (..., M, K)} {dpnp.ndarray, usm_ndarray} - Right-hand side + Right-hand side. trans : {0, 1, 2} , optional Type of system to solve: @@ -286,11 +282,11 @@ def lu_solve(lu_and_piv, b, trans=0, overwrite_b=False, check_finite=True): ===== ================= Default: ``0``. - overwrite_b : {None, bool}, optional + overwrite_b : bool, optional Whether to overwrite data in `b` (may increase performance). Default: ``False``. - check_finite : {None, bool}, optional + check_finite : bool, optional Whether to check that the input matrix contains only finite numbers. Disabling may give a performance gain, but may result in problems (crashes, non-termination) if the inputs do contain infinities or NaNs. @@ -302,14 +298,14 @@ def lu_solve(lu_and_piv, b, trans=0, overwrite_b=False, check_finite=True): x : {(M,), (..., M, K)} dpnp.ndarray Solution to the system - Warning - ------- + Warnings + -------- This function synchronizes in order to validate array elements when ``check_finite=True``. See Also -------- - :obj:`dpnp.scipy.linalg.lu_factor` : LU factorize a matrix. + :func:`dpnp.scipy.linalg.lu_factor` : LU factorize a matrix. Examples -------- From 7fae3a6ea29c4fe231d01cf78ed09b3a8c956004 Mon Sep 17 00:00:00 2001 From: "github-actions[bot]" <41898282+github-actions[bot]@users.noreply.github.com> Date: Fri, 13 Mar 2026 17:31:49 +0100 Subject: [PATCH 10/35] Weekly pre-commit autoupdate (#2808) This PR updates the `.pre-commit-config.yaml` using `pre-commit autoupdate`. --- .pre-commit-config.yaml | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 66245039ce3c..d08ddc36c1ab 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -57,19 +57,19 @@ repos: hooks: - id: pyupgrade - repo: https://github.com/codespell-project/codespell - rev: v2.4.1 + rev: v2.4.2 hooks: - id: codespell args: ["-L", "abd"] # ignore "abd" used in einsum tests additional_dependencies: - tomli - repo: https://github.com/psf/black - rev: 26.1.0 + rev: 26.3.0 hooks: - id: black exclude: "dpnp/_version.py" - repo: https://github.com/pycqa/isort - rev: 8.0.0 + rev: 8.0.1 hooks: - id: isort name: isort (python) From c5f212a2a88232c865481ee688ef4929f107d922 Mon Sep 17 00:00:00 2001 From: Anton <100830759+antonwolfy@users.noreply.github.com> Date: Wed, 18 Mar 2026 12:00:27 +0100 Subject: [PATCH 11/35] Mute expecting runtime warning raised in the test (#2822) This PR follows up #2792 and marks one more `RuntimeWarning` warning as expected due to possible overflow. --- dpnp/tests/test_special.py | 3 +++ 1 file changed, 3 insertions(+) diff --git a/dpnp/tests/test_special.py b/dpnp/tests/test_special.py index 1ebb64d8da7f..075bef5aeca3 100644 --- a/dpnp/tests/test_special.py +++ b/dpnp/tests/test_special.py @@ -106,6 +106,9 @@ def test_erfc(self, inverse): atol=self.tol, ) + @pytest.mark.usefixtures( + "suppress_overflow_encountered_in_cast_numpy_warnings" + ) def test_erfcx(self, inverse): self._check_variant_func( inverse, From 7d9765b772599f123ca247f12b9d3bcd386ea33c Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Wed, 18 Mar 2026 13:11:11 +0100 Subject: [PATCH 12/35] Bump styfle/cancel-workflow-action from 0.13.0 to 0.13.1 (#2820) Bumps [styfle/cancel-workflow-action](https://github.com/styfle/cancel-workflow-action) from 0.13.0 to 0.13.1. --- .github/workflows/build-sphinx.yml | 2 +- .github/workflows/check-onemath.yaml | 2 +- .github/workflows/conda-package.yml | 2 +- .github/workflows/cron-run-tests.yaml | 2 +- .github/workflows/generate_coverage.yaml | 2 +- 5 files changed, 5 insertions(+), 5 deletions(-) diff --git a/.github/workflows/build-sphinx.yml b/.github/workflows/build-sphinx.yml index 0745ca1ca9dc..60530556efcc 100644 --- a/.github/workflows/build-sphinx.yml +++ b/.github/workflows/build-sphinx.yml @@ -47,7 +47,7 @@ jobs: steps: - name: Cancel Previous Runs - uses: styfle/cancel-workflow-action@3155a141048f8f89c06b4cdae32e7853e97536bc # 0.13.0 + uses: styfle/cancel-workflow-action@d07a454dad7609a92316b57b23c9ccfd4f59af66 # 0.13.1 with: access_token: ${{ github.token }} diff --git a/.github/workflows/check-onemath.yaml b/.github/workflows/check-onemath.yaml index 409117c692b9..9296ba2cc903 100644 --- a/.github/workflows/check-onemath.yaml +++ b/.github/workflows/check-onemath.yaml @@ -34,7 +34,7 @@ jobs: steps: - name: Cancel Previous Runs - uses: styfle/cancel-workflow-action@3155a141048f8f89c06b4cdae32e7853e97536bc # 0.13.0 + uses: styfle/cancel-workflow-action@d07a454dad7609a92316b57b23c9ccfd4f59af66 # 0.13.1 with: access_token: ${{ github.token }} diff --git a/.github/workflows/conda-package.yml b/.github/workflows/conda-package.yml index a12486300aa0..17ee76b6567c 100644 --- a/.github/workflows/conda-package.yml +++ b/.github/workflows/conda-package.yml @@ -49,7 +49,7 @@ jobs: steps: - name: Cancel Previous Runs - uses: styfle/cancel-workflow-action@3155a141048f8f89c06b4cdae32e7853e97536bc # 0.13.0 + uses: styfle/cancel-workflow-action@d07a454dad7609a92316b57b23c9ccfd4f59af66 # 0.13.1 with: access_token: ${{ github.token }} diff --git a/.github/workflows/cron-run-tests.yaml b/.github/workflows/cron-run-tests.yaml index f8e8394c6713..ea4fd4f14fc3 100644 --- a/.github/workflows/cron-run-tests.yaml +++ b/.github/workflows/cron-run-tests.yaml @@ -43,7 +43,7 @@ jobs: steps: - name: Cancel Previous Runs - uses: styfle/cancel-workflow-action@3155a141048f8f89c06b4cdae32e7853e97536bc # 0.13.0 + uses: styfle/cancel-workflow-action@d07a454dad7609a92316b57b23c9ccfd4f59af66 # 0.13.1 with: access_token: ${{ github.token }} diff --git a/.github/workflows/generate_coverage.yaml b/.github/workflows/generate_coverage.yaml index 2cbe97ab0242..bfc3c7357a3e 100644 --- a/.github/workflows/generate_coverage.yaml +++ b/.github/workflows/generate_coverage.yaml @@ -33,7 +33,7 @@ jobs: steps: - name: Cancel Previous Runs - uses: styfle/cancel-workflow-action@3155a141048f8f89c06b4cdae32e7853e97536bc # 0.13.0 + uses: styfle/cancel-workflow-action@d07a454dad7609a92316b57b23c9ccfd4f59af66 # 0.13.1 with: access_token: ${{ github.token }} From 19efa58b4797e28aacaa5715be1b5ef93bb02e2e Mon Sep 17 00:00:00 2001 From: "github-actions[bot]" <41898282+github-actions[bot]@users.noreply.github.com> Date: Wed, 18 Mar 2026 14:32:04 +0100 Subject: [PATCH 13/35] Weekly pre-commit autoupdate (#2818) This PR updates the `.pre-commit-config.yaml` using `pre-commit autoupdate`. --- .pre-commit-config.yaml | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index d08ddc36c1ab..92b81fe95852 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -64,7 +64,7 @@ repos: additional_dependencies: - tomli - repo: https://github.com/psf/black - rev: 26.3.0 + rev: 26.3.1 hooks: - id: black exclude: "dpnp/_version.py" @@ -94,7 +94,7 @@ repos: - id: clang-format args: ["-i"] - repo: https://github.com/gitleaks/gitleaks - rev: v8.30.0 + rev: v8.30.1 hooks: - id: gitleaks - repo: https://github.com/jumanjihouse/pre-commit-hooks @@ -127,7 +127,7 @@ repos: hooks: - id: actionlint - repo: https://github.com/BlankSpruce/gersemi - rev: 0.26.0 + rev: 0.26.1 hooks: - id: gersemi exclude: "dpnp/backend/cmake/Modules/" From 3a6af45d0130b0b8c7eb7dbd41a3d7e3fd18a4d5 Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Wed, 18 Mar 2026 16:25:24 +0100 Subject: [PATCH 14/35] Bump actions/download-artifact from 8.0.0 to 8.0.1 (#2819) Bumps [actions/download-artifact](https://github.com/actions/download-artifact) from 8.0.0 to 8.0.1. --- .github/workflows/check-onemath.yaml | 4 ++-- .github/workflows/conda-package.yml | 10 +++++----- 2 files changed, 7 insertions(+), 7 deletions(-) diff --git a/.github/workflows/check-onemath.yaml b/.github/workflows/check-onemath.yaml index 9296ba2cc903..acbfcac96890 100644 --- a/.github/workflows/check-onemath.yaml +++ b/.github/workflows/check-onemath.yaml @@ -87,7 +87,7 @@ jobs: fetch-depth: 0 - name: Download artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.environment-file-name }} path: ${{ env.environment-file-loc }} @@ -181,7 +181,7 @@ jobs: fetch-depth: 0 - name: Download artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.environment-file-name }} path: ${{ env.environment-file-loc }} diff --git a/.github/workflows/conda-package.yml b/.github/workflows/conda-package.yml index 17ee76b6567c..f10a372edf23 100644 --- a/.github/workflows/conda-package.yml +++ b/.github/workflows/conda-package.yml @@ -151,7 +151,7 @@ jobs: path: ${{ env.dpnp-repo-path }} - name: Download artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} path: ${{ env.pkg-path-in-channel }} @@ -280,7 +280,7 @@ jobs: path: ${{ env.dpnp-repo-path }} - name: Download artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} path: ${{ env.pkg-path-in-channel }} @@ -439,12 +439,12 @@ jobs: fetch-depth: ${{ env.fetch-depth }} - name: Download artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ matrix.python }} - name: Download wheels artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.package-name }} ${{ runner.os }} Wheels Python ${{ matrix.python }} @@ -528,7 +528,7 @@ jobs: path: ${{ env.dpnp-repo-path }} - name: Download artifact - uses: actions/download-artifact@70fc10c6e5e1ce46ad2ea6f2b72d43f7d47b13c3 # v8.0.0 + uses: actions/download-artifact@3e5f45b2cfb9172054b4087a40e8e0b5a5461e7c # v8.0.1 with: name: ${{ env.package-name }} ${{ runner.os }} Python ${{ env.python-ver }} path: ${{ env.pkg-path-in-channel }} From 506a7756bc69839d3e7d0d9b9c914cb4d9684121 Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Wed, 18 Mar 2026 17:30:54 +0100 Subject: [PATCH 15/35] Bump github/codeql-action from 4.32.4 to 4.32.6 (#2809) Bumps [github/codeql-action](https://github.com/github/codeql-action) from 4.32.4 to 4.32.6. --- .github/workflows/openssf-scorecard.yml | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/.github/workflows/openssf-scorecard.yml b/.github/workflows/openssf-scorecard.yml index 8b4cc3b93f64..5d7e0677281e 100644 --- a/.github/workflows/openssf-scorecard.yml +++ b/.github/workflows/openssf-scorecard.yml @@ -72,6 +72,6 @@ jobs: # Upload the results to GitHub's code scanning dashboard. - name: "Upload to code-scanning" - uses: github/codeql-action/upload-sarif@89a39a4e59826350b863aa6b6252a07ad50cf83e # v4.32.4 + uses: github/codeql-action/upload-sarif@0d579ffd059c29b07949a3cce3983f0780820c98 # v4.32.6 with: sarif_file: results.sarif From cba3d51374714064122210bf6aba9631dfd06e53 Mon Sep 17 00:00:00 2001 From: "dependabot[bot]" <49699333+dependabot[bot]@users.noreply.github.com> Date: Wed, 18 Mar 2026 19:07:48 +0100 Subject: [PATCH 16/35] Bump mshick/add-pr-comment from 2.8.2 to 3.9.0 (#2821) Bumps [mshick/add-pr-comment](https://github.com/mshick/add-pr-comment) from 2.8.2 to 3.9.0. --- .github/workflows/build-sphinx.yml | 4 ++-- .github/workflows/conda-package.yml | 2 +- 2 files changed, 3 insertions(+), 3 deletions(-) diff --git a/.github/workflows/build-sphinx.yml b/.github/workflows/build-sphinx.yml index 60530556efcc..87a7311b95e4 100644 --- a/.github/workflows/build-sphinx.yml +++ b/.github/workflows/build-sphinx.yml @@ -224,7 +224,7 @@ jobs: if: env.GH_EVENT_OPEN_PR_UPSTREAM == 'true' env: PR_NUM: ${{ github.event.number }} - uses: mshick/add-pr-comment@b8f338c590a895d50bcbfa6c5859251edc8952fc # v2.8.2 + uses: mshick/add-pr-comment@ffd016c7e151d97d69d21a843022fd4cd5b96fe5 # v3.9.0.8.3.9.0 with: message-id: url_to_docs message: | @@ -268,7 +268,7 @@ jobs: git push tokened_docs gh-pages - name: Modify the comment with URL to official documentation - uses: mshick/add-pr-comment@b8f338c590a895d50bcbfa6c5859251edc8952fc # v2.8.2 + uses: mshick/add-pr-comment@ffd016c7e151d97d69d21a843022fd4cd5b96fe5 # v3.9.0.8.3.9.0 with: message-id: url_to_docs find: | diff --git a/.github/workflows/conda-package.yml b/.github/workflows/conda-package.yml index f10a372edf23..c894c530a20e 100644 --- a/.github/workflows/conda-package.yml +++ b/.github/workflows/conda-package.yml @@ -654,7 +654,7 @@ jobs: - name: Post result to PR if: ${{ github.event.pull_request && !github.event.pull_request.head.repo.fork }} - uses: mshick/add-pr-comment@b8f338c590a895d50bcbfa6c5859251edc8952fc # v2.8.2 + uses: mshick/add-pr-comment@ffd016c7e151d97d69d21a843022fd4cd5b96fe5 # v3.9.0.8.3.9.0 with: message-id: array_api_results message: | From 21d5e1ace58640d8ab4b2831c832d28e6ccfc448 Mon Sep 17 00:00:00 2001 From: Anton <100830759+antonwolfy@users.noreply.github.com> Date: Wed, 18 Mar 2026 20:34:25 +0100 Subject: [PATCH 17/35] Upgrade `clang-format` to v22 and improve `pylint` pre-commit configuration (#2813) ## Summary - Migrate clang-format from `pocc/pre-commit-hooks` to `pre-commit/mirrors-clang-format` v22.1.0 - Update `.clang-format` config to maintain consistent code style across versions - Configure pylint to properly handle C extension modules (numpy, dpctl) - Remove manual clang-format-12 installation from GitHub workflow - Reformat all C++ files with clang-format v22 - Add reformatting commit to `.git-blame-ignore-revs` ## Changes ### Pre-commit configuration - Switch to `pre-commit/mirrors-clang-format` (v22.1.0) for better version pinning and consistency across environments - Add `--disable=c-extension-no-member` to pylint to avoid false positives on C extension objects ### Code style - Update `.clang-format`: Set `AfterControlStatement: Never` to preserve existing brace placement style - Reformat all C++ source files with clang-format v22 (net -159 lines due to template parameter formatting improvements) ### Pylint configuration - Add `extension-pkg-allow-list = ["numpy"]` to help pylint understand numpy better - Add `generated-members` patterns to handle dynamically created attributes on numpy types ### CI/CD - Remove obsolete `apt-get install clang-format-12` step from pre-commit workflow --- .clang-format | 2 +- .git-blame-ignore-revs | 3 + .github/workflows/pre-commit.yml | 7 -- .pre-commit-config.yaml | 7 +- benchmarks/asv.conf.json | 12 +- dpnp/backend/extensions/blas/dot_common.hpp | 7 +- dpnp/backend/extensions/blas/gemm.cpp | 3 +- dpnp/backend/extensions/blas/gemm_batch.cpp | 3 +- dpnp/backend/extensions/blas/gemv.cpp | 3 +- dpnp/backend/extensions/blas/syrk.cpp | 3 +- dpnp/backend/extensions/common/ext/common.hpp | 6 +- .../extensions/common/ext/dispatch_table.hpp | 9 +- .../elementwise_functions/common.hpp | 107 +++++++++--------- .../elementwise_functions.hpp | 49 +++----- .../simplify_iteration_space.cpp | 3 +- dpnp/backend/extensions/fft/common.hpp | 8 +- dpnp/backend/extensions/fft/out_of_place.tpp | 5 +- .../extensions/lapack/evd_batch_common.hpp | 3 +- .../backend/extensions/lapack/geqrf_batch.cpp | 14 +-- dpnp/backend/extensions/lapack/gesv.cpp | 19 ++-- dpnp/backend/extensions/lapack/gesv_batch.cpp | 11 +- .../extensions/lapack/gesv_common_utils.hpp | 6 +- dpnp/backend/extensions/lapack/gesvd.cpp | 7 +- .../backend/extensions/lapack/gesvd_batch.cpp | 10 +- .../extensions/lapack/gesvd_common_utils.hpp | 6 +- dpnp/backend/extensions/lapack/getrf.cpp | 16 +-- dpnp/backend/extensions/lapack/getrs.cpp | 3 +- .../backend/extensions/lapack/getrs_batch.cpp | 3 +- dpnp/backend/extensions/lapack/heevd.cpp | 4 +- .../backend/extensions/lapack/heevd_batch.cpp | 4 +- .../extensions/lapack/linalg_exceptions.hpp | 5 +- .../backend/extensions/lapack/orgqr_batch.cpp | 18 +-- dpnp/backend/extensions/lapack/syevd.cpp | 4 +- .../backend/extensions/lapack/syevd_batch.cpp | 4 +- .../backend/extensions/lapack/ungqr_batch.cpp | 18 +-- .../extensions/statistics/bincount.cpp | 5 +- .../statistics/histogram_common.hpp | 55 ++------- .../extensions/statistics/histogramdd.cpp | 5 +- .../statistics/sliding_window1d.hpp | 68 +++-------- .../ufunc/elementwise_functions/erf_funcs.cpp | 3 +- .../elementwise_functions/interpolate.cpp | 4 +- .../ufunc/elementwise_functions/populate.hpp | 22 ++-- dpnp/backend/extensions/ufunc/ufunc_py.cpp | 5 +- dpnp/backend/extensions/vm/common.hpp | 10 +- dpnp/backend/kernels/dpnp_krnl_random.cpp | 6 +- .../kernels/elementwise_functions/degrees.hpp | 5 +- .../kernels/elementwise_functions/divmod.hpp | 5 +- .../kernels/elementwise_functions/fabs.hpp | 5 +- .../kernels/elementwise_functions/fmax.hpp | 6 +- .../kernels/elementwise_functions/fmin.hpp | 6 +- .../kernels/elementwise_functions/isclose.hpp | 6 +- .../elementwise_functions/nan_to_num.hpp | 3 +- .../kernels/elementwise_functions/radians.hpp | 5 +- dpnp/backend/src/dpnp_fptr.hpp | 3 +- dpnp/backend/src/queue_sycl.cpp | 3 +- dpnp/backend/src/queue_sycl.hpp | 10 +- dpnp/backend/tests/test_random.cpp | 5 +- pyproject.toml | 4 + 58 files changed, 241 insertions(+), 400 deletions(-) diff --git a/.clang-format b/.clang-format index 622a5bf67634..4304e0ed40d4 100644 --- a/.clang-format +++ b/.clang-format @@ -16,7 +16,7 @@ BinPackParameters: false BraceWrapping: AfterCaseLabel: true AfterClass: true - AfterControlStatement: MultiLine + AfterControlStatement: Never AfterEnum: true AfterFunction: true AfterNamespace: true diff --git a/.git-blame-ignore-revs b/.git-blame-ignore-revs index 841f009ace89..e4fe0bcc4b2f 100644 --- a/.git-blame-ignore-revs +++ b/.git-blame-ignore-revs @@ -17,3 +17,6 @@ c106d91b866f4acd30226b68519b12a73a881490 # Add pygrep-hooks to pre-commit config e62718415aa3660da5f607e352c991a063a54219 + +# Bump clang-format from 12.0.1 to 22.1.0 version +c2d65bd451a7d8e5b6319147da95e9dabf7a382b diff --git a/.github/workflows/pre-commit.yml b/.github/workflows/pre-commit.yml index d5f9f22146fc..d8f59405ce89 100644 --- a/.github/workflows/pre-commit.yml +++ b/.github/workflows/pre-commit.yml @@ -15,13 +15,6 @@ jobs: timeout-minutes: 10 steps: - - name: Set up clang-format - run: | - sudo apt-get install -y clang-format-12 - sudo unlink /usr/bin/clang-format - sudo ln -s /usr/bin/clang-format-12 /usr/bin/clang-format - clang-format --version - - name: Set up pip packages uses: BSFishy/pip-action@8f2d471d809dc20b6ada98c91910b6ae6243f318 # v1 with: diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 92b81fe95852..57ec9e2a2a8e 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -88,8 +88,8 @@ repos: additional_dependencies: - flake8-docstrings==1.7.0 - flake8-bugbear==24.12.12 -- repo: https://github.com/pocc/pre-commit-hooks - rev: v1.3.5 +- repo: https://github.com/pre-commit/mirrors-clang-format + rev: v22.1.0 hooks: - id: clang-format args: ["-i"] @@ -114,7 +114,8 @@ repos: "-sn", # Don't display the score "--disable=import-error", "--disable=redefined-builtin", - "--disable=unused-wildcard-import" + "--disable=unused-wildcard-import", + "--disable=c-extension-no-member" ] files: '^dpnp/(dpnp_iface.*|fft|linalg|scipy|dpnp_array)' - repo: https://github.com/macisamuele/language-formatters-pre-commit-hooks diff --git a/benchmarks/asv.conf.json b/benchmarks/asv.conf.json index c5e5663e21fb..3d0e7f88d55f 100644 --- a/benchmarks/asv.conf.json +++ b/benchmarks/asv.conf.json @@ -15,7 +15,9 @@ // List of branches to benchmark. If not provided, defaults to "master" // (for git) or "tip" (for mercurial). - "branches": ["HEAD"], + "branches": [ + "HEAD" + ], // The DVCS being used. If not set, it will be automatically // determined from "repo" by looking at the protocol in the URL @@ -35,7 +37,9 @@ // The Pythons you'd like to test against. If not provided, defaults // to the current version of Python used to run `asv`. - "pythons": ["3.7"], + "pythons": [ + "3.7" + ], // The matrix of dependencies to test. Each key is the name of a // package (in PyPI) and the values are version numbers. An empty @@ -53,7 +57,6 @@ // environments in. If not provided, defaults to "env" "env_dir": "env", - // The directory (relative to the current directory) that raw benchmark // results are stored in. If not provided, defaults to "results". "results_dir": "results", @@ -79,7 +82,8 @@ // skipped for the matching benchmark. // // "regressions_first_commits": { - // "some_benchmark": "352cdf", // Consider regressions only after this commit + // "some_benchmark": "352cdf", // Consider regressions only after this + // commit // "another_benchmark": null, // Skip regression detection altogether // } } diff --git a/dpnp/backend/extensions/blas/dot_common.hpp b/dpnp/backend/extensions/blas/dot_common.hpp index 1672e7217cba..383804ff1718 100644 --- a/dpnp/backend/extensions/blas/dot_common.hpp +++ b/dpnp/backend/extensions/blas/dot_common.hpp @@ -97,8 +97,7 @@ std::pair if (!dpctl::utils::queues_are_compatible( exec_q, - {vectorX.get_queue(), vectorY.get_queue(), result.get_queue()})) - { + {vectorX.get_queue(), vectorY.get_queue(), result.get_queue()})) { throw py::value_error( "USM allocations are not compatible with the execution queue."); } @@ -120,8 +119,8 @@ std::pair const int vectorY_typenum = vectorY.get_typenum(); const int result_typenum = result.get_typenum(); - if (result_typenum != vectorX_typenum || result_typenum != vectorY_typenum) - { + if (result_typenum != vectorX_typenum || + result_typenum != vectorY_typenum) { throw py::value_error("Given arrays must be of the same type."); } diff --git a/dpnp/backend/extensions/blas/gemm.cpp b/dpnp/backend/extensions/blas/gemm.cpp index 48c1ae98ead4..86f751baf2e0 100644 --- a/dpnp/backend/extensions/blas/gemm.cpp +++ b/dpnp/backend/extensions/blas/gemm.cpp @@ -181,8 +181,7 @@ std::tuple if (!dpctl::utils::queues_are_compatible( exec_q, - {matrixA.get_queue(), matrixB.get_queue(), resultC.get_queue()})) - { + {matrixA.get_queue(), matrixB.get_queue(), resultC.get_queue()})) { throw py::value_error( "USM allocations are not compatible with the execution queue."); } diff --git a/dpnp/backend/extensions/blas/gemm_batch.cpp b/dpnp/backend/extensions/blas/gemm_batch.cpp index a6cd7ac4e130..d02b035922c0 100644 --- a/dpnp/backend/extensions/blas/gemm_batch.cpp +++ b/dpnp/backend/extensions/blas/gemm_batch.cpp @@ -237,8 +237,7 @@ std::tuple if (!dpctl::utils::queues_are_compatible( exec_q, - {matrixA.get_queue(), matrixB.get_queue(), resultC.get_queue()})) - { + {matrixA.get_queue(), matrixB.get_queue(), resultC.get_queue()})) { throw py::value_error( "USM allocations are not compatible with the execution queue."); } diff --git a/dpnp/backend/extensions/blas/gemv.cpp b/dpnp/backend/extensions/blas/gemv.cpp index a9c5414ef8c7..0b6ae78bc76e 100644 --- a/dpnp/backend/extensions/blas/gemv.cpp +++ b/dpnp/backend/extensions/blas/gemv.cpp @@ -169,8 +169,7 @@ std::pair if (!dpctl::utils::queues_are_compatible( exec_q, - {matrixA.get_queue(), vectorX.get_queue(), vectorY.get_queue()})) - { + {matrixA.get_queue(), vectorX.get_queue(), vectorY.get_queue()})) { throw py::value_error( "USM allocations are not compatible with the execution queue."); } diff --git a/dpnp/backend/extensions/blas/syrk.cpp b/dpnp/backend/extensions/blas/syrk.cpp index 8b0ebce3d888..9668e72b57f6 100644 --- a/dpnp/backend/extensions/blas/syrk.cpp +++ b/dpnp/backend/extensions/blas/syrk.cpp @@ -248,8 +248,7 @@ std::pair } if (!dpctl::utils::queues_are_compatible( - exec_q, {matrixA.get_queue(), resultC.get_queue()})) - { + exec_q, {matrixA.get_queue(), resultC.get_queue()})) { throw py::value_error( "USM allocations are not compatible with the execution queue."); } diff --git a/dpnp/backend/extensions/common/ext/common.hpp b/dpnp/backend/extensions/common/ext/common.hpp index d626b56ea00c..f0ce1722bfb1 100644 --- a/dpnp/backend/extensions/common/ext/common.hpp +++ b/dpnp/backend/extensions/common/ext/common.hpp @@ -213,8 +213,7 @@ sycl::nd_range<1> pybind11::dtype dtype_from_typenum(int dst_typenum); template - typename factoryT, + template typename factoryT, int _num_types = type_dispatch::num_types> inline void init_dispatch_vector(dispatchT dispatch_vector[]) { @@ -223,8 +222,7 @@ inline void init_dispatch_vector(dispatchT dispatch_vector[]) } template - typename factoryT, + template typename factoryT, int _num_types = type_dispatch::num_types> inline void init_dispatch_table(dispatchT dispatch_table[][_num_types]) { diff --git a/dpnp/backend/extensions/common/ext/dispatch_table.hpp b/dpnp/backend/extensions/common/ext/dispatch_table.hpp index 4cfe1bd57250..6655f054f355 100644 --- a/dpnp/backend/extensions/common/ext/dispatch_table.hpp +++ b/dpnp/backend/extensions/common/ext/dispatch_table.hpp @@ -99,8 +99,7 @@ using SupportedDTypeList2 = std::vector; template - typename Func> + template typename Func> struct TableBuilder { template @@ -125,8 +124,7 @@ struct TableBuilder template - typename Func> + template typename Func> struct TableBuilder2 { template @@ -232,8 +230,7 @@ class DispatchTable2 } template - typename Func> + template typename Func> void populate_dispatch_table() { using TBulder = typename TableBuilder2::type; diff --git a/dpnp/backend/extensions/elementwise_functions/common.hpp b/dpnp/backend/extensions/elementwise_functions/common.hpp index df2b3afe53b9..f3b15c8d6774 100644 --- a/dpnp/backend/extensions/elementwise_functions/common.hpp +++ b/dpnp/backend/extensions/elementwise_functions/common.hpp @@ -131,8 +131,7 @@ struct UnaryTwoOutputsContigFunctor else if constexpr (enable_sg_loadstore && UnaryTwoOutputsOpT::supports_sg_loadstore::value && UnaryTwoOutputsOpT::supports_vec::value && - (vec_sz > 1)) - { + (vec_sz > 1)) { auto sg = ndit.get_sub_group(); const std::uint16_t sgSize = sg.get_max_local_range()[0]; @@ -171,8 +170,7 @@ struct UnaryTwoOutputsContigFunctor } else if constexpr (enable_sg_loadstore && UnaryTwoOutputsOpT::supports_sg_loadstore::value && - std::is_same_v) - { + std::is_same_v) { // default: use scalar-value function auto sg = ndit.get_sub_group(); @@ -214,8 +212,7 @@ struct UnaryTwoOutputsContigFunctor } } else if constexpr (enable_sg_loadstore && - UnaryTwoOutputsOpT::supports_sg_loadstore::value) - { + UnaryTwoOutputsOpT::supports_sg_loadstore::value) { // default: use scalar-value function auto sg = ndit.get_sub_group(); @@ -359,8 +356,7 @@ struct BinaryTwoOutputsContigFunctor if constexpr (enable_sg_loadstore && BinaryOperatorT::supports_sg_loadstore::value && - BinaryOperatorT::supports_vec::value && (vec_sz > 1)) - { + BinaryOperatorT::supports_vec::value && (vec_sz > 1)) { auto sg = ndit.get_sub_group(); std::uint16_t sgSize = sg.get_max_local_range()[0]; @@ -405,8 +401,7 @@ struct BinaryTwoOutputsContigFunctor } } else if constexpr (enable_sg_loadstore && - BinaryOperatorT::supports_sg_loadstore::value) - { + BinaryOperatorT::supports_sg_loadstore::value) { auto sg = ndit.get_sub_group(); const std::uint16_t sgSize = sg.get_max_local_range()[0]; @@ -528,21 +523,18 @@ struct BinaryTwoOutputsStridedFunctor * dpctl::tensor::kernels::elementwise_common namespace. */ template - class UnaryTwoOutputsType, + template class UnaryTwoOutputsType, template - class UnaryTwoOutputsContigFunctorT, + bool enable> class UnaryTwoOutputsContigFunctorT, template - class kernel_name, + std::uint8_t nv> class kernel_name, std::uint8_t vec_sz = 4u, std::uint8_t n_vecs = 2u> sycl::event @@ -576,8 +568,7 @@ sycl::event if (is_aligned(arg_p) && is_aligned(res1_p) && - is_aligned(res2_p)) - { + is_aligned(res2_p)) { static constexpr bool enable_sg_loadstore = true; using KernelName = BaseKernelName; using Impl = @@ -613,12 +604,15 @@ sycl::event * dpctl::tensor::kernels::elementwise_common namespace. */ template - class UnaryTwoOutputsType, - template - class UnaryTwoOutputsStridedFunctorT, - template - class kernel_name> + template class UnaryTwoOutputsType, + template class UnaryTwoOutputsStridedFunctorT, + template class kernel_name> sycl::event unary_two_outputs_strided_impl( sycl::queue &exec_q, std::size_t nelems, @@ -665,27 +659,25 @@ sycl::event unary_two_outputs_strided_impl( * @note It extends binary_contig_impl from * dpctl::tensor::kernels::elementwise_common namespace. */ -template - class BinaryTwoOutputsType, - template - class BinaryTwoOutputsContigFunctorT, - template - class kernel_name, - std::uint8_t vec_sz = 4u, - std::uint8_t n_vecs = 2u> +template < + typename argTy1, + typename argTy2, + template class BinaryTwoOutputsType, + template class BinaryTwoOutputsContigFunctorT, + template class kernel_name, + std::uint8_t vec_sz = 4u, + std::uint8_t n_vecs = 2u> sycl::event binary_two_outputs_contig_impl(sycl::queue &exec_q, std::size_t nelems, @@ -726,8 +718,7 @@ sycl::event if (is_aligned(arg1_tp) && is_aligned(arg2_tp) && is_aligned(res1_tp) && - is_aligned(res2_tp)) - { + is_aligned(res2_tp)) { static constexpr bool enable_sg_loadstore = true; using KernelName = BaseKernelName; using Impl = BinaryTwoOutputsContigFunctorT - class BinaryTwoOutputsType, - template - class BinaryTwoOutputsStridedFunctorT, - template - class kernel_name> +template class BinaryTwoOutputsType, + template class BinaryTwoOutputsStridedFunctorT, + template class kernel_name> sycl::event binary_two_outputs_strided_impl( sycl::queue &exec_q, std::size_t nelems, diff --git a/dpnp/backend/extensions/elementwise_functions/elementwise_functions.hpp b/dpnp/backend/extensions/elementwise_functions/elementwise_functions.hpp index c996ac07df02..6a29c9a33c5a 100644 --- a/dpnp/backend/extensions/elementwise_functions/elementwise_functions.hpp +++ b/dpnp/backend/extensions/elementwise_functions/elementwise_functions.hpp @@ -309,8 +309,7 @@ std::pair // check that types are supported if (dst1_typeid != func_output_typeids.first || - dst2_typeid != func_output_typeids.second) - { + dst2_typeid != func_output_typeids.second) { throw py::value_error( "One of destination arrays has unexpected elemental data type."); } @@ -362,8 +361,7 @@ std::pair dpctl::tensor::overlap::SameLogicalTensors(); if ((overlap(src, dst1) && !same_logical_tensors(src, dst1)) || (overlap(src, dst2) && !same_logical_tensors(src, dst2)) || - (overlap(dst1, dst2) && !same_logical_tensors(dst1, dst2))) - { + (overlap(dst1, dst2) && !same_logical_tensors(dst1, dst2))) { throw py::value_error("Arrays index overlapping segments of memory"); } @@ -430,8 +428,7 @@ std::pair simplified_dst2_strides, src_offset, dst1_offset, dst2_offset); if (nd == 1 && simplified_src_strides[0] == 1 && - simplified_dst1_strides[0] == 1 && simplified_dst2_strides[0] == 1) - { + simplified_dst1_strides[0] == 1 && simplified_dst2_strides[0] == 1) { // Special case of contiguous data auto contig_fn = contig_dispatch_vector[src_typeid]; @@ -625,8 +622,7 @@ std::pair py_binary_ufunc( auto const &same_logical_tensors = dpctl::tensor::overlap::SameLogicalTensors(); if ((overlap(src1, dst) && !same_logical_tensors(src1, dst)) || - (overlap(src2, dst) && !same_logical_tensors(src2, dst))) - { + (overlap(src2, dst) && !same_logical_tensors(src2, dst))) { throw py::value_error("Arrays index overlapping segments of memory"); } // check memory overlap @@ -693,8 +689,7 @@ std::pair py_binary_ufunc( if ((nd == 1) && isEqual(simplified_src1_strides, unit_stride) && isEqual(simplified_src2_strides, unit_stride) && - isEqual(simplified_dst_strides, unit_stride)) - { + isEqual(simplified_dst_strides, unit_stride)) { auto contig_fn = contig_dispatch_table[src1_typeid][src2_typeid]; if (contig_fn != nullptr) { @@ -716,8 +711,7 @@ std::pair py_binary_ufunc( // special case of C-contiguous matrix and a row if (isEqual(simplified_src2_strides, zero_one_strides) && isEqual(simplified_src1_strides, {simplified_shape[1], one}) && - isEqual(simplified_dst_strides, {simplified_shape[1], one})) - { + isEqual(simplified_dst_strides, {simplified_shape[1], one})) { auto matrix_row_broadcast_fn = contig_matrix_row_broadcast_dispatch_table[src1_typeid] [src2_typeid]; @@ -731,8 +725,7 @@ std::pair py_binary_ufunc( is_aligned( src2_data + src2_offset * src2_itemsize) && is_aligned( - dst_data + dst_offset * dst_itemsize)) - { + dst_data + dst_offset * dst_itemsize)) { std::size_t n0 = simplified_shape[0]; std::size_t n1 = simplified_shape[1]; sycl::event comp_ev = matrix_row_broadcast_fn( @@ -749,8 +742,7 @@ std::pair py_binary_ufunc( } if (isEqual(simplified_src1_strides, one_zero_strides) && isEqual(simplified_src2_strides, {one, simplified_shape[0]}) && - isEqual(simplified_dst_strides, {one, simplified_shape[0]})) - { + isEqual(simplified_dst_strides, {one, simplified_shape[0]})) { auto row_matrix_broadcast_fn = contig_row_matrix_broadcast_dispatch_table[src1_typeid] [src2_typeid]; @@ -765,8 +757,7 @@ std::pair py_binary_ufunc( is_aligned( src2_data + src2_offset * src2_itemsize) && is_aligned( - dst_data + dst_offset * dst_itemsize)) - { + dst_data + dst_offset * dst_itemsize)) { std::size_t n0 = simplified_shape[1]; std::size_t n1 = simplified_shape[0]; sycl::event comp_ev = row_matrix_broadcast_fn( @@ -839,8 +830,7 @@ py::object py_binary_ufunc_result_type(const py::dtype &input1_dtype, } if (src1_typeid < 0 || src1_typeid >= td_ns::num_types || src2_typeid < 0 || - src2_typeid >= td_ns::num_types) - { + src2_typeid >= td_ns::num_types) { throw std::runtime_error("binary output type lookup failed"); } int dst_typeid = output_types_table[src1_typeid][src2_typeid]; @@ -898,8 +888,8 @@ std::pair } // check that queues are compatible - if (!dpctl::utils::queues_are_compatible(exec_q, {src1, src2, dst1, dst2})) - { + if (!dpctl::utils::queues_are_compatible(exec_q, + {src1, src2, dst1, dst2})) { throw py::value_error( "Execution queue is not compatible with allocation queues"); } @@ -955,8 +945,7 @@ std::pair (overlap(src1, dst2) && !same_logical_tensors(src1, dst2)) || (overlap(src2, dst1) && !same_logical_tensors(src2, dst1)) || (overlap(src2, dst2) && !same_logical_tensors(src2, dst2)) || - (overlap(dst1, dst2))) - { + (overlap(dst1, dst2))) { throw py::value_error("Arrays index overlapping segments of memory"); } @@ -1031,8 +1020,7 @@ std::pair if ((nd == 1) && isEqual(simplified_src1_strides, unit_stride) && isEqual(simplified_src2_strides, unit_stride) && isEqual(simplified_dst1_strides, unit_stride) && - isEqual(simplified_dst2_strides, unit_stride)) - { + isEqual(simplified_dst2_strides, unit_stride)) { auto contig_fn = contig_dispatch_table[src1_typeid][src2_typeid]; if (contig_fn != nullptr) { @@ -1107,8 +1095,7 @@ std::pair py_binary_two_outputs_ufunc_result_type( } if (src1_typeid < 0 || src1_typeid >= td_ns::num_types || src2_typeid < 0 || - src2_typeid >= td_ns::num_types) - { + src2_typeid >= td_ns::num_types) { throw std::runtime_error("binary output type lookup failed"); } std::pair dst_typeids = @@ -1263,8 +1250,7 @@ std::pair std::initializer_list{1}; if ((nd == 1) && isEqual(simplified_rhs_strides, unit_stride) && - isEqual(simplified_lhs_strides, unit_stride)) - { + isEqual(simplified_lhs_strides, unit_stride)) { auto contig_fn = contig_dispatch_table[rhs_typeid][lhs_typeid]; if (contig_fn != nullptr) { @@ -1283,8 +1269,7 @@ std::pair static constexpr py::ssize_t one{1}; // special case of C-contiguous matrix and a row if (isEqual(simplified_rhs_strides, one_zero_strides) && - isEqual(simplified_lhs_strides, {one, simplified_shape[0]})) - { + isEqual(simplified_lhs_strides, {one, simplified_shape[0]})) { auto row_matrix_broadcast_fn = contig_row_matrix_broadcast_dispatch_table[rhs_typeid] [lhs_typeid]; diff --git a/dpnp/backend/extensions/elementwise_functions/simplify_iteration_space.cpp b/dpnp/backend/extensions/elementwise_functions/simplify_iteration_space.cpp index e34cb74fcb0a..c60602ccb01d 100644 --- a/dpnp/backend/extensions/elementwise_functions/simplify_iteration_space.cpp +++ b/dpnp/backend/extensions/elementwise_functions/simplify_iteration_space.cpp @@ -292,8 +292,7 @@ void simplify_iteration_space_4( simplified_dst_strides.reserve(nd); if ((src1_strides[0] < 0) && (src2_strides[0] < 0) && - (src3_strides[0] < 0) && (dst_strides[0] < 0)) - { + (src3_strides[0] < 0) && (dst_strides[0] < 0)) { simplified_src1_strides.push_back(-src1_strides[0]); simplified_src2_strides.push_back(-src2_strides[0]); simplified_src3_strides.push_back(-src3_strides[0]); diff --git a/dpnp/backend/extensions/fft/common.hpp b/dpnp/backend/extensions/fft/common.hpp index f76da9721316..44f0b43f8597 100644 --- a/dpnp/backend/extensions/fft/common.hpp +++ b/dpnp/backend/extensions/fft/common.hpp @@ -56,8 +56,7 @@ class DescriptorWrapper { mkl_dft::precision fft_prec = get_precision(); if (fft_prec == mkl_dft::precision::DOUBLE && - !q.get_device().has(sycl::aspect::fp64)) - { + !q.get_device().has(sycl::aspect::fp64)) { throw py::value_error("Descriptor is double precision but the " "device does not support double precision."); } @@ -66,10 +65,7 @@ class DescriptorWrapper queue_ptr_ = std::make_unique(q); } - descr_type &get_descriptor() - { - return descr_; - } + descr_type &get_descriptor() { return descr_; } const sycl::queue &get_queue() const { diff --git a/dpnp/backend/extensions/fft/out_of_place.tpp b/dpnp/backend/extensions/fft/out_of_place.tpp index 290408dc60bc..ed5cd37df7f1 100644 --- a/dpnp/backend/extensions/fft/out_of_place.tpp +++ b/dpnp/backend/extensions/fft/out_of_place.tpp @@ -82,9 +82,8 @@ std::pair } sycl::queue exec_q = descr.get_queue(); - if (!dpctl::utils::queues_are_compatible(exec_q, - {in.get_queue(), out.get_queue()})) - { + if (!dpctl::utils::queues_are_compatible( + exec_q, {in.get_queue(), out.get_queue()})) { throw py::value_error("USM allocations are not compatible with the " "execution queue of the descriptor."); } diff --git a/dpnp/backend/extensions/lapack/evd_batch_common.hpp b/dpnp/backend/extensions/lapack/evd_batch_common.hpp index e1debdc35934..d2edffcf520a 100644 --- a/dpnp/backend/extensions/lapack/evd_batch_common.hpp +++ b/dpnp/backend/extensions/lapack/evd_batch_common.hpp @@ -75,8 +75,7 @@ std::pair expected_eig_vecs_nd, expected_eig_vals_nd); if (eig_vecs_shape[2] != eig_vals_shape[0] || - eig_vecs_shape[0] != eig_vals_shape[1]) - { + eig_vecs_shape[0] != eig_vals_shape[1]) { throw py::value_error( "The shape of 'eig_vals' must be (batch_size, n), " "where batch_size = " + diff --git a/dpnp/backend/extensions/lapack/geqrf_batch.cpp b/dpnp/backend/extensions/lapack/geqrf_batch.cpp index e0821e23e440..033c3db01b10 100644 --- a/dpnp/backend/extensions/lapack/geqrf_batch.cpp +++ b/dpnp/backend/extensions/lapack/geqrf_batch.cpp @@ -98,13 +98,13 @@ static sycl::event geqrf_batch_impl(sycl::queue &exec_q, geqrf_batch_event = mkl_lapack::geqrf_batch( exec_q, - m, // The number of rows in each matrix in the batch; (0 ≤ m). - // It must be a non-negative integer. - n, // The number of columns in each matrix in the batch; (0 ≤ n). - // It must be a non-negative integer. - a, // Pointer to the batch of matrices, each of size (m x n). - lda, // The leading dimension of each matrix in the batch. - // For row major layout, lda ≥ max(1, m). + m, // The number of rows in each matrix in the batch; (0 ≤ m). + // It must be a non-negative integer. + n, // The number of columns in each matrix in the batch; (0 ≤ n). + // It must be a non-negative integer. + a, // Pointer to the batch of matrices, each of size (m x n). + lda, // The leading dimension of each matrix in the batch. + // For row major layout, lda ≥ max(1, m). stride_a, // Stride between consecutive matrices in the batch. tau, // Pointer to the array of scalar factors of the elementary // reflectors for each matrix in the batch. diff --git a/dpnp/backend/extensions/lapack/gesv.cpp b/dpnp/backend/extensions/lapack/gesv.cpp index 0569fab2c350..bec24db585a6 100644 --- a/dpnp/backend/extensions/lapack/gesv.cpp +++ b/dpnp/backend/extensions/lapack/gesv.cpp @@ -114,14 +114,14 @@ static sycl::event gesv_impl(sycl::queue &exec_q, try { getrf_event = mkl_lapack::getrf( exec_q, - n, // The order of the square matrix A (0 ≤ n). - // It must be a non-negative integer. - n, // The number of columns in the square matrix A (0 ≤ n). - // It must be a non-negative integer. - a, // Pointer to the square matrix A (n x n). - lda, // The leading dimension of matrix A. - // It must be at least max(1, n). - ipiv, // Pointer to the output array of pivot indices. + n, // The order of the square matrix A (0 ≤ n). + // It must be a non-negative integer. + n, // The number of columns in the square matrix A (0 ≤ n). + // It must be a non-negative integer. + a, // Pointer to the square matrix A (n x n). + lda, // The leading dimension of matrix A. + // It must be at least max(1, n). + ipiv, // Pointer to the output array of pivot indices. scratchpad, // Pointer to scratchpad memory to be used by MKL // routine for storing intermediate results. scratchpad_size, depends); @@ -242,8 +242,7 @@ std::pair // Ensure `batch_size`, `n` and 'nrhs' are non-zero, otherwise return empty // events if (helper::check_zeros_shape(coeff_matrix_nd, coeff_matrix_shape) || - helper::check_zeros_shape(dependent_vals_nd, dependent_vals_shape)) - { + helper::check_zeros_shape(dependent_vals_nd, dependent_vals_shape)) { // nothing to do return std::make_pair(sycl::event(), sycl::event()); } diff --git a/dpnp/backend/extensions/lapack/gesv_batch.cpp b/dpnp/backend/extensions/lapack/gesv_batch.cpp index ce02f8517eb5..893279245344 100644 --- a/dpnp/backend/extensions/lapack/gesv_batch.cpp +++ b/dpnp/backend/extensions/lapack/gesv_batch.cpp @@ -258,10 +258,10 @@ static sycl::event gesv_batch_impl(sycl::queue &exec_q, try { gesv_event = mkl_lapack::gesv( exec_q, - n, // The order of the square matrix A - // and the number of rows in matrix B (0 ≤ n). - nrhs, // The number of right-hand sides, - // i.e., the number of columns in matrix B (0 ≤ nrhs). + n, // The order of the square matrix A + // and the number of rows in matrix B (0 ≤ n). + nrhs, // The number of right-hand sides, + // i.e., the number of columns in matrix B (0 ≤ nrhs). a_batch, // Pointer to the square coefficient matrix A (n x n). lda, // The leading dimension of a, must be at least max(1, n). current_ipiv, // The pivot indices that define the permutation @@ -341,8 +341,7 @@ std::pair // Ensure `batch_size`, `n` and 'nrhs' are non-zero, otherwise return empty // events if (helper::check_zeros_shape(coeff_matrix_nd, coeff_matrix_shape) || - helper::check_zeros_shape(dependent_vals_nd, dependent_vals_shape)) - { + helper::check_zeros_shape(dependent_vals_nd, dependent_vals_shape)) { // nothing to do return std::make_pair(sycl::event(), sycl::event()); } diff --git a/dpnp/backend/extensions/lapack/gesv_common_utils.hpp b/dpnp/backend/extensions/lapack/gesv_common_utils.hpp index d86d7e29413e..62f1e9589a0b 100644 --- a/dpnp/backend/extensions/lapack/gesv_common_utils.hpp +++ b/dpnp/backend/extensions/lapack/gesv_common_utils.hpp @@ -64,8 +64,7 @@ inline void common_gesv_checks(sycl::queue &exec_q, } if (dependent_vals_nd < min_dependent_vals_ndim || - dependent_vals_nd > max_dependent_vals_ndim) - { + dependent_vals_nd > max_dependent_vals_ndim) { throw py::value_error("The dependent values array has ndim=" + std::to_string(dependent_vals_nd) + ", but a " + std::to_string(min_dependent_vals_ndim) + @@ -95,8 +94,7 @@ inline void common_gesv_checks(sycl::queue &exec_q, // check compatibility of execution queue and allocation queue if (!dpctl::utils::queues_are_compatible(exec_q, - {coeff_matrix, dependent_vals})) - { + {coeff_matrix, dependent_vals})) { throw py::value_error( "Execution queue is not compatible with allocation queues."); } diff --git a/dpnp/backend/extensions/lapack/gesvd.cpp b/dpnp/backend/extensions/lapack/gesvd.cpp index d46179ac3b9a..e347837e3cfe 100644 --- a/dpnp/backend/extensions/lapack/gesvd.cpp +++ b/dpnp/backend/extensions/lapack/gesvd.cpp @@ -171,8 +171,7 @@ std::pair // Ensure `m` and 'n' are non-zero, otherwise return empty // events if (gesvd_utils::check_zeros_shape_gesvd(a_array, out_s, out_u, out_vt, - jobu_val, jobvt_val)) - { + jobu_val, jobvt_val)) { // nothing to do return std::make_pair(sycl::event(), sycl::event()); } @@ -223,8 +222,8 @@ struct GesvdContigFactory { fnT get() { - if constexpr (types::GesvdTypePairSupportFactory::is_defined) - { + if constexpr (types::GesvdTypePairSupportFactory::is_defined) { return gesvd_impl; } else { diff --git a/dpnp/backend/extensions/lapack/gesvd_batch.cpp b/dpnp/backend/extensions/lapack/gesvd_batch.cpp index eb9903ba6e1e..868facc200e2 100644 --- a/dpnp/backend/extensions/lapack/gesvd_batch.cpp +++ b/dpnp/backend/extensions/lapack/gesvd_batch.cpp @@ -102,8 +102,7 @@ static sycl::event gesvd_batch_impl(sycl::queue &exec_q, std::int64_t vt_size = 0; if (jobu == oneapi::mkl::jobsvd::somevec || - jobu == oneapi::mkl::jobsvd::vectorsina) - { + jobu == oneapi::mkl::jobsvd::vectorsina) { u_size = m * k; vt_size = k * n; } @@ -238,8 +237,7 @@ std::pair // Ensure `batch_size`, `m` and 'n' are non-zero, otherwise return empty // events if (gesvd_utils::check_zeros_shape_gesvd(a_array, out_s, out_u, out_vt, - jobu_val, jobvt_val)) - { + jobu_val, jobvt_val)) { // nothing to do return std::make_pair(sycl::event(), sycl::event()); } @@ -293,8 +291,8 @@ struct GesvdBatchContigFactory { fnT get() { - if constexpr (types::GesvdTypePairSupportFactory::is_defined) - { + if constexpr (types::GesvdTypePairSupportFactory::is_defined) { return gesvd_batch_impl; } else { diff --git a/dpnp/backend/extensions/lapack/gesvd_common_utils.hpp b/dpnp/backend/extensions/lapack/gesvd_common_utils.hpp index ce2d9c1eb474..1cd2c8ac4997 100644 --- a/dpnp/backend/extensions/lapack/gesvd_common_utils.hpp +++ b/dpnp/backend/extensions/lapack/gesvd_common_utils.hpp @@ -122,8 +122,7 @@ inline void common_gesvd_checks(sycl::queue &exec_q, // check compatibility of execution queue and allocation queue if (!dpctl::utils::queues_are_compatible(exec_q, - {a_array, out_s, out_u, out_vt})) - { + {a_array, out_s, out_u, out_vt})) { throw py::value_error( "Execution queue is not compatible with allocation queues."); } @@ -131,8 +130,7 @@ inline void common_gesvd_checks(sycl::queue &exec_q, auto const &overlap = dpctl::tensor::overlap::MemoryOverlap(); if (overlap(a_array, out_s) || overlap(a_array, out_u) || overlap(a_array, out_vt) || overlap(out_s, out_u) || - overlap(out_s, out_vt) || overlap(out_u, out_vt)) - { + overlap(out_s, out_vt) || overlap(out_u, out_vt)) { throw py::value_error("Arrays have overlapping segments of memory"); } diff --git a/dpnp/backend/extensions/lapack/getrf.cpp b/dpnp/backend/extensions/lapack/getrf.cpp index abf20aff643a..870ccc8e811a 100644 --- a/dpnp/backend/extensions/lapack/getrf.cpp +++ b/dpnp/backend/extensions/lapack/getrf.cpp @@ -91,14 +91,14 @@ static sycl::event getrf_impl(sycl::queue &exec_q, getrf_event = mkl_lapack::getrf( exec_q, - m, // The number of rows in the input matrix A (0 ≤ m). - // It must be a non-negative integer. - n, // The number of columns in the input matrix A (0 ≤ n). - // It must be a non-negative integer. - a, // Pointer to the input matrix A (m x n). - lda, // The leading dimension of matrix A. - // It must be at least max(1, m). - ipiv, // Pointer to the output array of pivot indices. + m, // The number of rows in the input matrix A (0 ≤ m). + // It must be a non-negative integer. + n, // The number of columns in the input matrix A (0 ≤ n). + // It must be a non-negative integer. + a, // Pointer to the input matrix A (m x n). + lda, // The leading dimension of matrix A. + // It must be at least max(1, m). + ipiv, // Pointer to the output array of pivot indices. scratchpad, // Pointer to scratchpad memory to be used by MKL // routine for storing intermediate results. scratchpad_size, depends); diff --git a/dpnp/backend/extensions/lapack/getrs.cpp b/dpnp/backend/extensions/lapack/getrs.cpp index 8108afd97003..94e1a1027898 100644 --- a/dpnp/backend/extensions/lapack/getrs.cpp +++ b/dpnp/backend/extensions/lapack/getrs.cpp @@ -208,8 +208,7 @@ std::pair // check compatibility of execution queue and allocation queue if (!dpctl::utils::queues_are_compatible(exec_q, - {a_array, b_array, ipiv_array})) - { + {a_array, b_array, ipiv_array})) { throw py::value_error( "Execution queue is not compatible with allocation queues"); } diff --git a/dpnp/backend/extensions/lapack/getrs_batch.cpp b/dpnp/backend/extensions/lapack/getrs_batch.cpp index 9fc6ce1a5dfc..f4fb446c328d 100644 --- a/dpnp/backend/extensions/lapack/getrs_batch.cpp +++ b/dpnp/backend/extensions/lapack/getrs_batch.cpp @@ -253,8 +253,7 @@ std::pair // check compatibility of execution queue and allocation queue if (!dpctl::utils::queues_are_compatible(exec_q, - {a_array, b_array, ipiv_array})) - { + {a_array, b_array, ipiv_array})) { throw py::value_error( "Execution queue is not compatible with allocation queues"); } diff --git a/dpnp/backend/extensions/lapack/heevd.cpp b/dpnp/backend/extensions/lapack/heevd.cpp index 5990e5344a17..96d6a03e9b8e 100644 --- a/dpnp/backend/extensions/lapack/heevd.cpp +++ b/dpnp/backend/extensions/lapack/heevd.cpp @@ -124,8 +124,8 @@ struct HeevdContigFactory { fnT get() { - if constexpr (types::HeevdTypePairSupportFactory::is_defined) - { + if constexpr (types::HeevdTypePairSupportFactory::is_defined) { return heevd_impl; } else { diff --git a/dpnp/backend/extensions/lapack/heevd_batch.cpp b/dpnp/backend/extensions/lapack/heevd_batch.cpp index e1c1a96bc320..e8614498bd41 100644 --- a/dpnp/backend/extensions/lapack/heevd_batch.cpp +++ b/dpnp/backend/extensions/lapack/heevd_batch.cpp @@ -161,8 +161,8 @@ struct HeevdBatchContigFactory { fnT get() { - if constexpr (types::HeevdTypePairSupportFactory::is_defined) - { + if constexpr (types::HeevdTypePairSupportFactory::is_defined) { return heevd_batch_impl; } else { diff --git a/dpnp/backend/extensions/lapack/linalg_exceptions.hpp b/dpnp/backend/extensions/lapack/linalg_exceptions.hpp index d087adfbd2b6..c823d1995a4e 100644 --- a/dpnp/backend/extensions/lapack/linalg_exceptions.hpp +++ b/dpnp/backend/extensions/lapack/linalg_exceptions.hpp @@ -37,10 +37,7 @@ class LinAlgError : public std::exception public: explicit LinAlgError(const char *message) : msg_(message) {} - const char *what() const noexcept override - { - return msg_.c_str(); - } + const char *what() const noexcept override { return msg_.c_str(); } private: std::string msg_; diff --git a/dpnp/backend/extensions/lapack/orgqr_batch.cpp b/dpnp/backend/extensions/lapack/orgqr_batch.cpp index ef1c85b91f4a..a29fe9b342fc 100644 --- a/dpnp/backend/extensions/lapack/orgqr_batch.cpp +++ b/dpnp/backend/extensions/lapack/orgqr_batch.cpp @@ -100,15 +100,15 @@ static sycl::event orgqr_batch_impl(sycl::queue &exec_q, orgqr_batch_event = mkl_lapack::orgqr_batch( exec_q, - m, // The number of rows in each matrix in the batch; (0 ≤ m). - // It must be a non-negative integer. - n, // The number of columns in each matrix in the batch; (0 ≤ n). - // It must be a non-negative integer. - k, // The number of elementary reflectors - // whose product defines the matrices Qi; (0 ≤ k ≤ n). - a, // Pointer to the batch of matrices, each of size (m x n). - lda, // The leading dimension of each matrix in the batch. - // For row major layout, lda ≥ max(1, m). + m, // The number of rows in each matrix in the batch; (0 ≤ m). + // It must be a non-negative integer. + n, // The number of columns in each matrix in the batch; (0 ≤ n). + // It must be a non-negative integer. + k, // The number of elementary reflectors + // whose product defines the matrices Qi; (0 ≤ k ≤ n). + a, // Pointer to the batch of matrices, each of size (m x n). + lda, // The leading dimension of each matrix in the batch. + // For row major layout, lda ≥ max(1, m). stride_a, // Stride between consecutive matrices in the batch. tau, // Pointer to the array of scalar factors of the elementary // reflectors for each matrix in the batch. diff --git a/dpnp/backend/extensions/lapack/syevd.cpp b/dpnp/backend/extensions/lapack/syevd.cpp index af69cf9e6b7e..3ecd386299ac 100644 --- a/dpnp/backend/extensions/lapack/syevd.cpp +++ b/dpnp/backend/extensions/lapack/syevd.cpp @@ -124,8 +124,8 @@ struct SyevdContigFactory { fnT get() { - if constexpr (types::SyevdTypePairSupportFactory::is_defined) - { + if constexpr (types::SyevdTypePairSupportFactory::is_defined) { return syevd_impl; } else { diff --git a/dpnp/backend/extensions/lapack/syevd_batch.cpp b/dpnp/backend/extensions/lapack/syevd_batch.cpp index 0c326e5d79bb..13237d27a35c 100644 --- a/dpnp/backend/extensions/lapack/syevd_batch.cpp +++ b/dpnp/backend/extensions/lapack/syevd_batch.cpp @@ -161,8 +161,8 @@ struct SyevdBatchContigFactory { fnT get() { - if constexpr (types::SyevdTypePairSupportFactory::is_defined) - { + if constexpr (types::SyevdTypePairSupportFactory::is_defined) { return syevd_batch_impl; } else { diff --git a/dpnp/backend/extensions/lapack/ungqr_batch.cpp b/dpnp/backend/extensions/lapack/ungqr_batch.cpp index 7c890d968b0a..04de27cb257c 100644 --- a/dpnp/backend/extensions/lapack/ungqr_batch.cpp +++ b/dpnp/backend/extensions/lapack/ungqr_batch.cpp @@ -100,15 +100,15 @@ static sycl::event ungqr_batch_impl(sycl::queue &exec_q, ungqr_batch_event = mkl_lapack::ungqr_batch( exec_q, - m, // The number of rows in each matrix in the batch; (0 ≤ m). - // It must be a non-negative integer. - n, // The number of columns in each matrix in the batch; (0 ≤ n). - // It must be a non-negative integer. - k, // The number of elementary reflectors - // whose product defines the matrices Qi; (0 ≤ k ≤ n). - a, // Pointer to the batch of matrices, each of size (m x n). - lda, // The leading dimension of each matrix in the batch. - // For row major layout, lda ≥ max(1, m). + m, // The number of rows in each matrix in the batch; (0 ≤ m). + // It must be a non-negative integer. + n, // The number of columns in each matrix in the batch; (0 ≤ n). + // It must be a non-negative integer. + k, // The number of elementary reflectors + // whose product defines the matrices Qi; (0 ≤ k ≤ n). + a, // Pointer to the batch of matrices, each of size (m x n). + lda, // The leading dimension of each matrix in the batch. + // For row major layout, lda ≥ max(1, m). stride_a, // Stride between consecutive matrices in the batch. tau, // Pointer to the array of scalar factors of the elementary // reflectors for each matrix in the batch. diff --git a/dpnp/backend/extensions/statistics/bincount.cpp b/dpnp/backend/extensions/statistics/bincount.cpp index ba258cd55447..9bfe5c2a2449 100644 --- a/dpnp/backend/extensions/statistics/bincount.cpp +++ b/dpnp/backend/extensions/statistics/bincount.cpp @@ -59,10 +59,7 @@ struct BincountEdges { } - boundsT get_bounds() const - { - return {min, max}; - } + boundsT get_bounds() const { return {min, max}; } template size_t get_bin(const sycl::nd_item<_Dims> &, diff --git a/dpnp/backend/extensions/statistics/histogram_common.hpp b/dpnp/backend/extensions/statistics/histogram_common.hpp index 539b42475fbf..02fc66f26610 100644 --- a/dpnp/backend/extensions/statistics/histogram_common.hpp +++ b/dpnp/backend/extensions/statistics/histogram_common.hpp @@ -64,10 +64,7 @@ struct CachedData local_data = LocalData(shape, cgh); } - T *get_ptr() const - { - return &local_data[0]; - } + T *get_ptr() const { return &local_data[0]; } template void init(const sycl::nd_item<_Dims> &item) const @@ -83,15 +80,9 @@ struct CachedData } } - size_t size() const - { - return local_data.size(); - } + size_t size() const { return local_data.size(); } - T &operator[](const sycl::id &id) const - { - return local_data[id]; - } + T &operator[](const sycl::id &id) const { return local_data[id]; } template > T &operator[](const size_t id) const @@ -119,25 +110,16 @@ struct UncachedData _shape = shape; } - T *get_ptr() const - { - return global_data; - } + T *get_ptr() const { return global_data; } template void init(const sycl::nd_item<_Dims> &) const { } - size_t size() const - { - return _shape.size(); - } + size_t size() const { return _shape.size(); } - T &operator[](const sycl::id &id) const - { - return global_data[id]; - } + T &operator[](const sycl::id &id) const { return global_data[id]; } template > T &operator[](const size_t id) const @@ -235,10 +217,7 @@ struct HistWithLocalCopies } } - uint32_t size() const - { - return local_hist.size(); - } + uint32_t size() const { return local_hist.size(); } private: LocalHist local_hist; @@ -251,10 +230,7 @@ struct HistGlobalMemory static constexpr bool const sync_after_init = false; static constexpr bool const sync_before_finalize = false; - HistGlobalMemory(T *global_data) - { - global_hist = global_data; - } + HistGlobalMemory(T *global_data) { global_hist = global_data; } template void init(const sycl::nd_item<_Dims> &) const @@ -280,24 +256,15 @@ struct HistGlobalMemory template struct NoWeights { - constexpr T get(size_t) const - { - return 1; - } + constexpr T get(size_t) const { return 1; } }; template struct Weights { - Weights(T *weights) - { - data = weights; - } + Weights(T *weights) { data = weights; } - T get(size_t id) const - { - return data[id]; - } + T get(size_t id) const { return data[id]; } private: T *data = nullptr; diff --git a/dpnp/backend/extensions/statistics/histogramdd.cpp b/dpnp/backend/extensions/statistics/histogramdd.cpp index a5ed4a8c7d1c..bd2177073333 100644 --- a/dpnp/backend/extensions/statistics/histogramdd.cpp +++ b/dpnp/backend/extensions/statistics/histogramdd.cpp @@ -90,10 +90,7 @@ struct EdgesDd } } - boundsT get_bounds() const - { - return {&min[0], &max[0]}; - } + boundsT get_bounds() const { return {&min[0], &max[0]}; } auto get_bin_for_dim(const EdgesT &val, const EdgesT *edges_data, diff --git a/dpnp/backend/extensions/statistics/sliding_window1d.hpp b/dpnp/backend/extensions/statistics/sliding_window1d.hpp index c5a5bac111dd..f33a23609666 100644 --- a/dpnp/backend/extensions/statistics/sliding_window1d.hpp +++ b/dpnp/backend/extensions/statistics/sliding_window1d.hpp @@ -129,30 +129,15 @@ class _RegistryDataStorage return sycl::shift_group_right(sbgroup, data[y], x); } - constexpr SizeT size_y() const - { - return _size; - } + constexpr SizeT size_y() const { return _size; } - SizeT size_x() const - { - return sbgroup.get_max_local_range()[0]; - } + SizeT size_x() const { return sbgroup.get_max_local_range()[0]; } - SizeT total_size() const - { - return size_x() * size_y(); - } + SizeT total_size() const { return size_x() * size_y(); } - ncT *ptr() - { - return data; - } + ncT *ptr() { return data; } - SizeT x() const - { - return sbgroup.get_local_linear_id(); - } + SizeT x() const { return sbgroup.get_local_linear_id(); } protected: const sycl::sub_group sbgroup; @@ -277,8 +262,7 @@ struct RegistryData : public _RegistryDataStorage T *load(const T *const data, const bool &mask, const T &default_v) { - return load( - data, [mask](auto &&) { return mask; }, default_v); + return load(data, [mask](auto &&) { return mask; }, default_v); } T *load(const T *const data) @@ -349,10 +333,7 @@ struct RegistryData : public _RegistryDataStorage return store(data, [mask](auto &&) { return mask; }); } - T *store(T *const data) - { - return store(data, true); - } + T *store(T *const data) { return store(data, true); } }; template @@ -379,10 +360,7 @@ struct RegistryWindow : public RegistryData } } - void advance_left(const T &fill_value) - { - advance_left(1, fill_value); - } + void advance_left(const T &fill_value) { advance_left(1, fill_value); } void advance_left() { @@ -400,25 +378,13 @@ class Span Span(T *const data, const SizeT size) : data_(data), size_(size) {} - T *begin() const - { - return data(); - } + T *begin() const { return data(); } - T *end() const - { - return data() + size(); - } + T *end() const { return data() + size(); } - SizeT size() const - { - return size_; - } + SizeT size() const { return size_; } - T *data() const - { - return data_; - } + T *data() const { return data_; } protected: T *const data_; @@ -443,15 +409,9 @@ class PaddedSpan : public Span { } - T *padded_begin() const - { - return this->begin() - pad(); - } + T *padded_begin() const { return this->begin() - pad(); } - SizeT pad() const - { - return pad_; - } + SizeT pad() const { return pad_; } protected: const SizeT pad_; diff --git a/dpnp/backend/extensions/ufunc/elementwise_functions/erf_funcs.cpp b/dpnp/backend/extensions/ufunc/elementwise_functions/erf_funcs.cpp index 5254e50d3faf..6f10e651fe25 100644 --- a/dpnp/backend/extensions/ufunc/elementwise_functions/erf_funcs.cpp +++ b/dpnp/backend/extensions/ufunc/elementwise_functions/erf_funcs.cpp @@ -184,8 +184,7 @@ using ew_cmn_ns::unary_strided_impl_fn_ptr_t; }; template