From 989538b8ac3ebdc0decdbd8c2d4916c67f1ab32c Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 15:52:04 +0200
Subject: [PATCH 01/12] Sort NaNs to the end for descending order

`dpnp.sort`/`dpnp.argsort` (and their `dpnp.ndarray`/`dpnp.tensor`
counterparts) placed `NaN` values first when sorting in descending
order, while NumPy 2.5 keeps `NaN` at the end for both ascending and
descending order.

Fix the merge-sort comparators so that only the comparison between
non-NaN values is reversed for descending order, and make the radix-sort
float casts map `NaN` to the maximum key so it lands in the last bucket
regardless of direction.
---
 CHANGELOG.md                                  |  1 +
 .../include/kernels/sorting/radix_sort.hpp    | 22 ++++++++----
 .../include/utils/rich_comparisons.hpp        | 35 +++++++++++++++++--
 dpnp/tests/tensor/test_usm_ndarray_sorting.py |  3 +-
 4 files changed, 50 insertions(+), 11 deletions(-)

diff --git a/CHANGELOG.md b/CHANGELOG.md
index 0f47c903d54d..c30e2af86895 100644
--- a/CHANGELOG.md
+++ b/CHANGELOG.md
@@ -99,6 +99,7 @@ This release is compatible with NumPy 2.5.
 * Fixed `dpnp.cumsum`, `dpnp.cumprod`, and their `nan`/`cumulative_*` variants (including `dpnp.tensor.cumulative_sum`/`cumulative_prod`) silently returning incorrect results when accumulating along an axis of an array with more than one row [#3063](https://github.com/IntelPython/dpnp/pull/3063)
 * Fixed `dpnp.einsum` returning a result whose memory layout differs from NumPy for the default `order="K"`, and ignoring `out` and `order` for a contraction over a size-0 dimension [#3058](https://github.com/IntelPython/dpnp/pull/3058)
 * Fixed operations on a boolean array whose bytes are not `0x00`/`0x01` [#3055](https://github.com/IntelPython/dpnp/pull/3055)
+* Fixed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts placing `NaN` values first instead of last when sorting in descending order
 
 ### Security
 
diff --git a/dpnp/tensor/libtensor/include/kernels/sorting/radix_sort.hpp b/dpnp/tensor/libtensor/include/kernels/sorting/radix_sort.hpp
index 163f2ae64dcc..d1a09a9f139b 100644
--- a/dpnp/tensor/libtensor/include/kernels/sorting/radix_sort.hpp
+++ b/dpnp/tensor/libtensor/include/kernels/sorting/radix_sort.hpp
@@ -170,9 +170,11 @@ std::uint16_t order_preserving_cast(sycl::half val)
 {
     using UIntT = std::uint16_t;
 
-    const UIntT uint_val = sycl::bit_cast<UIntT>(
-        (sycl::isnan(val)) ? std::numeric_limits<sycl::half>::quiet_NaN()
-                           : val);
+    // NaNs sort to the end for both orders
+    if (sycl::isnan(val))
+        return std::numeric_limits<UIntT>::max();
+
+    const UIntT uint_val = sycl::bit_cast<UIntT>(val);
     UIntT mask;
 
     // test the sign bit of the original value
@@ -203,8 +205,11 @@ std::uint32_t order_preserving_cast(FloatT val)
 {
     using UIntT = std::uint32_t;
 
-    UIntT uint_val = sycl::bit_cast<UIntT>(
-        (sycl::isnan(val)) ? std::numeric_limits<FloatT>::quiet_NaN() : val);
+    // NaNs sort to the end for both orders
+    if (sycl::isnan(val))
+        return std::numeric_limits<UIntT>::max();
+
+    const UIntT uint_val = sycl::bit_cast<UIntT>(val);
 
     UIntT mask;
 
@@ -231,8 +236,11 @@ std::uint64_t order_preserving_cast(FloatT val)
 {
     using UIntT = std::uint64_t;
 
-    UIntT uint_val = sycl::bit_cast<UIntT>(
-        (sycl::isnan(val)) ? std::numeric_limits<FloatT>::quiet_NaN() : val);
+    // NaNs sort to the end for both orders
+    if (sycl::isnan(val))
+        return std::numeric_limits<UIntT>::max();
+
+    const UIntT uint_val = sycl::bit_cast<UIntT>(val);
     UIntT mask;
 
     // test the sign bit of the original value
diff --git a/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp b/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp
index 3544fbddc008..93029b9d4107 100644
--- a/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp
+++ b/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp
@@ -60,9 +60,10 @@ struct ExtendedRealFPLess
 template <typename fpT>
 struct ExtendedRealFPGreater
 {
+    /* [R, nan] — NaNs sort to the end, as in ascending order */
     bool operator()(const fpT v1, const fpT v2) const
     {
-        return (!std::isnan(v2) && (std::isnan(v1) || (v2 < v1)));
+        return (!std::isnan(v1) && (std::isnan(v2) || (v2 < v1)));
     }
 };
 
@@ -106,10 +107,38 @@ struct ExtendedComplexFPLess
 template <typename cT>
 struct ExtendedComplexFPGreater
 {
+    /* [(R, R), (R, nan), (nan, R), (nan, nan)] — NaN-containing values keep
+       the same trailing groups as ascending order; only the finite-component
+       comparison within a group is reversed */
     bool operator()(const cT &v1, const cT &v2) const
     {
-        auto less_ = ExtendedComplexFPLess<cT>{};
-        return less_(v2, v1);
+        using realT = typename cT::value_type;
+
+        const realT real1 = std::real(v1);
+        const realT real2 = std::real(v2);
+
+        const bool r1_nan = std::isnan(real1);
+        const bool r2_nan = std::isnan(real2);
+
+        const realT imag1 = std::imag(v1);
+        const realT imag2 = std::imag(v2);
+
+        const bool i1_nan = std::isnan(imag1);
+        const bool i2_nan = std::isnan(imag2);
+
+        const int idx1 = ((r1_nan) ? 2 : 0) + ((i1_nan) ? 1 : 0);
+        const int idx2 = ((r2_nan) ? 2 : 0) + ((i2_nan) ? 1 : 0);
+
+        const bool res =
+            !(r1_nan && i1_nan) &&
+            ((idx1 < idx2) ||
+             ((idx1 == idx2) &&
+              ((r1_nan && !i1_nan && (imag2 < imag1)) ||
+               (!r1_nan && i1_nan && (real2 < real1)) ||
+               (!r1_nan && !i1_nan &&
+                ((real2 < real1) || (!(real1 < real2) && (imag2 < imag1)))))));
+
+        return res;
     }
 };
 
diff --git a/dpnp/tests/tensor/test_usm_ndarray_sorting.py b/dpnp/tests/tensor/test_usm_ndarray_sorting.py
index af96811bf2f9..9917f93bdca1 100644
--- a/dpnp/tests/tensor/test_usm_ndarray_sorting.py
+++ b/dpnp/tests/tensor/test_usm_ndarray_sorting.py
@@ -309,8 +309,9 @@ def test_sort_real_fp_nan(dtype, kind):
 
     s = dpt.sort(x, descending=True, kind=kind)
 
+    # NaNs sort to the end for descending order too matching NumPy
     expected = dpt.asarray(
-        [dpt.nan, dpt.nan, 0.2, 0.1, -0.0, 0.0, -0.1, -0.3], dtype=dtype
+        [0.2, 0.1, -0.0, 0.0, -0.1, -0.3, dpt.nan, dpt.nan], dtype=dtype
     )
 
     assert dpt.allclose(s, expected, equal_nan=True)

From 6872854dd5a84188ac73613048c1f453dceb5bf2 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 15:54:49 +0200
Subject: [PATCH 02/12] Port CuPy descending sort/argsort tests

Bring in the descending-order sort/argsort test coverage from cupy#10088,
adapted to dpnp's signature (`order` remains a positional parameter, so the
keyword-only checks pass an explicit `order=None`).
---
 .../cupy/sorting_tests/test_sort.py           | 161 ++++++++++++++++--
 1 file changed, 145 insertions(+), 16 deletions(-)

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 196f64fffa74..382393de7216 100644
--- a/dpnp/tests/third_party/cupy/sorting_tests/test_sort.py
+++ b/dpnp/tests/third_party/cupy/sorting_tests/test_sort.py
@@ -23,6 +23,14 @@ def get_array_module(*args):
 
 class TestSort(unittest.TestCase):
 
+    def _sort(self, xp, a, use_method, axis=-1, descending=None):
+        kwargs = {} if descending is None else {"descending": descending}
+        if use_method:
+            a.sort(axis=axis, **kwargs)
+            return a
+        else:
+            return xp.sort(a, axis=axis, **kwargs)
+
     # Test ranks
 
     def test_sort_zero_dim(self):
@@ -59,14 +67,13 @@ def test_external_sort_two_or_more_dim(self, xp):
     @testing.numpy_cupy_array_equal()
     def test_sort_dtype(self, xp, dtype):
         a = testing.shaped_random((10,), xp, dtype)
-        a.sort()
-        return a
+        return self._sort(xp, a, use_method=True)
 
     @testing.for_all_dtypes()
     @testing.numpy_cupy_array_equal()
     def test_external_sort_dtype(self, xp, dtype):
         a = testing.shaped_random((10,), xp, dtype)
-        return xp.sort(a)
+        return self._sort(xp, a, use_method=False)
 
     # Test contiguous arrays
 
@@ -103,8 +110,7 @@ def test_sort_axis1(self, xp):
     @testing.numpy_cupy_array_equal()
     def test_sort_axis2(self, xp):
         a = testing.shaped_random((2, 3, 4), xp)
-        a.sort(axis=1)
-        return a
+        return self._sort(xp, a, use_method=True, axis=1)
 
     @testing.numpy_cupy_array_equal()
     def test_sort_axis3(self, xp):
@@ -115,7 +121,7 @@ def test_sort_axis3(self, xp):
     @testing.numpy_cupy_array_equal()
     def test_external_sort_axis(self, xp):
         a = testing.shaped_random((2, 3, 3), xp)
-        return xp.sort(a, axis=0)
+        return self._sort(xp, a, use_method=False, axis=0)
 
     @testing.numpy_cupy_array_equal()
     def test_sort_negative_axis(self, xp):
@@ -184,8 +190,7 @@ def test_external_sort_invalid_negative_axis2(self):
     def test_nan1(self, xp, dtype):
         a = testing.shaped_random((10,), xp, dtype)
         a[2] = a[6] = xp.nan
-        out = xp.sort(a)
-        return out
+        return self._sort(xp, a, use_method=False)
 
     @testing.for_dtypes("efdFD")
     @testing.numpy_cupy_array_equal()
@@ -211,6 +216,68 @@ def test_nan4(self, xp, dtype):
         out = xp.sort(a, axis=2)
         return out
 
+    # Test descending order
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_all_dtypes()
+    @testing.numpy_cupy_array_equal()
+    def test_sort_descending_dtype(self, xp, dtype):
+        a = testing.shaped_random((10,), xp, dtype)
+        return self._sort(xp, a, use_method=True, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_all_dtypes()
+    @testing.numpy_cupy_array_equal()
+    def test_external_sort_descending_dtype(self, xp, dtype):
+        a = testing.shaped_random((10,), xp, dtype)
+        return self._sort(xp, a, use_method=False, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_all_dtypes()
+    @testing.numpy_cupy_array_equal()
+    def test_sort_descending_axis(self, xp, dtype):
+        # Enough rows, and long enough rows, that a segmented sort emitting
+        # them in the wrong order cannot coincidentally match -- including
+        # for bool, where each sorted row collapses to a count of `True`s.
+        a = testing.shaped_random((4, 5, 6), xp, dtype)
+        return self._sort(xp, a, use_method=True, axis=1, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_all_dtypes()
+    @testing.numpy_cupy_array_equal()
+    def test_external_sort_descending_axis(self, xp, dtype):
+        a = testing.shaped_random((4, 5, 6), xp, dtype)
+        return self._sort(xp, a, use_method=False, axis=1, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.numpy_cupy_array_equal()
+    def test_sort_descending_false_matches_default(self, xp):
+        a = testing.shaped_random((10,), xp)
+        return self._sort(xp, a, use_method=False, descending=False)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_dtypes("efdFD")
+    @testing.numpy_cupy_array_equal()
+    def test_sort_descending_nan(self, xp, dtype):
+        a = testing.shaped_random((10,), xp, dtype)
+        a[2] = a[6] = xp.nan
+        return self._sort(xp, a, use_method=False, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_dtypes("efdFD")
+    @testing.numpy_cupy_array_equal()
+    def test_sort_descending_nan_axis(self, xp, dtype):
+        a = testing.shaped_random((4, 5, 6), xp, dtype)
+        a[0, 2, 1] = a[1, 0, 3] = a[3, 4, 5] = xp.nan
+        return self._sort(xp, a, use_method=False, axis=1, descending=True)
+
+    def test_sort_descending_keyword_only(self):
+        a = cupy.arange(3)
+        with pytest.raises(TypeError):
+            cupy.sort(a, -1, None, None, True)
+        with pytest.raises(TypeError):
+            a.sort(-1, None, None, True)
+
     # Large case
 
     @testing.slow
@@ -308,14 +375,23 @@ def test_F_order(self, xp):
 )
 class TestArgsort(unittest.TestCase):
 
-    def argsort(self, a, axis=-1):
-        if self.external:
+    def argsort(self, a, axis=-1, descending=None):
+        xp = cupy.get_array_module(a)
+        if descending is None:
             # Need to explicitly specify kind="stable"
             # numpy uses "quicksort" as default
-            xp = cupy.get_array_module(a)
-            return xp.argsort(a, axis=axis, kind="stable")
+            kwargs = {"kind": "stable"}
+        else:
+            # numpy rejects `kind` combined with `descending`; `stable=True`
+            # is its replacement for forcing determinism. cupy's argsort is
+            # always stable regardless, and doesn't accept `stable`.
+            kwargs = {"descending": descending}
+            if xp is numpy:
+                kwargs["stable"] = True
+        if self.external:
+            return xp.argsort(a, axis=axis, **kwargs)
         else:
-            return a.argsort(axis=axis, kind="stable")
+            return a.argsort(axis=axis, **kwargs)
 
     # Test base cases
 
@@ -425,6 +501,59 @@ def test_nan2(self, xp, dtype):
         a[0, 2, 1] = a[1, 1, 3] = xp.nan
         return self.argsort(a)
 
+    # Test descending order
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_all_dtypes()
+    @testing.numpy_cupy_array_equal()
+    def test_argsort_descending_dtype(self, xp, dtype):
+        a = testing.shaped_random((10,), xp, dtype)
+        return self.argsort(a, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_all_dtypes()
+    @testing.numpy_cupy_array_equal()
+    def test_argsort_descending_axis(self, xp, dtype):
+        a = testing.shaped_random((4, 5, 6), xp, dtype)
+        return self.argsort(a, axis=0, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.numpy_cupy_array_equal()
+    def test_argsort_descending_false_matches_default(self, xp):
+        a = testing.shaped_random((10,), xp)
+        return self.argsort(a, descending=False)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.numpy_cupy_array_equal()
+    def test_argsort_descending_stable(self, xp):
+        # Repeated values must keep their original relative (ascending
+        # index) order, not just be the reverse of the ascending argsort.
+        a = xp.array([3, 1, 3, 1, 2, 3, 1, 2])
+        return self.argsort(a, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_dtypes("efdFD")
+    @testing.numpy_cupy_array_equal()
+    def test_argsort_descending_nan(self, xp, dtype):
+        a = testing.shaped_random((10,), xp, dtype)
+        a[2] = a[6] = xp.nan
+        return self.argsort(a, descending=True)
+
+    @testing.with_requires("numpy>=2.5")
+    @testing.for_dtypes("efdFD")
+    @testing.numpy_cupy_array_equal()
+    def test_argsort_descending_nan_axis(self, xp, dtype):
+        a = testing.shaped_random((4, 5, 6), xp, dtype)
+        a[0, 2, 1] = a[1, 0, 3] = a[3, 4, 5] = xp.nan
+        return self.argsort(a, axis=1, descending=True)
+
+    def test_argsort_descending_keyword_only(self):
+        a = cupy.arange(3)
+        with pytest.raises(TypeError):
+            cupy.argsort(a, -1, None, None, True)
+        with pytest.raises(TypeError):
+            a.argsort(-1, None, None, True)
+
 
 class TestSort_complex(unittest.TestCase):
 
@@ -515,9 +644,9 @@ def test_lexsort_workspace_oom(self):
             )
         )
 
-    # thread_unsafe marker requires pytest-run-parallel, not used by dpnp
     # @pytest.mark.thread_unsafe(
-    #     reason="contextlib.redirect_stderr replaces sys.stderr globally")
+    #     reason="contextlib.redirect_stderr replaces sys.stderr globally"
+    # )
     def test_no_stderr_noise_on_workspace_oom(self):
         # The thrust allocator's `noexcept`-driven stderr trace was
         # confusing to users (cupy/cupy#9894).  After the fix, OOM produces a
@@ -697,7 +826,7 @@ def test_partition_invalid_negative_axis2(self):
         }
     )
 )
-@pytest.mark.skip("not supported yet")
+@pytest.mark.skip("argpartition isn't supported yet")
 class TestArgpartition(unittest.TestCase):
 
     def argpartition(self, a, kth, axis=-1):

From 5210a9d5be89a3d5c3452eac6d5eff8f86d25934 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 17:30:02 +0200
Subject: [PATCH 03/12] Extend descending sort/argsort tests with NaN and
 stability coverage

Backport NumPy 2.5 descending-order test coverage (numpy gh-31345,
gh-31476, gh-31557) into dpnp's own sort test suite, adapted to the
dtypes dpnp supports:

* NaNs sort to the end for both ascending and descending float sorts.
* NaN-containing complex values sort to the end in the same groups for
  both orders, with finite values kept in lexicographic order.
* a stable (arg)sort keeps the original relative order of equal
  elements in both directions.

Expected results are delegated to `numpy.sort`/`numpy.argsort` with
`stable=True, descending=...`, so the tests require NumPy >= 2.5.
---
 dpnp/tests/test_sort.py | 77 +++++++++++++++++++++++++++++------------
 1 file changed, 54 insertions(+), 23 deletions(-)

diff --git a/dpnp/tests/test_sort.py b/dpnp/tests/test_sort.py
index 27a2afe79b6a..cda4ec81aa58 100644
--- a/dpnp/tests/test_sort.py
+++ b/dpnp/tests/test_sort.py
@@ -59,28 +59,26 @@ def test_kind(self, kind):
         expected = numpy.argsort(a, kind="stable")
         assert_array_equal(result, expected)
 
+    @testing.with_requires("numpy>=2.5")
     @pytest.mark.parametrize("descending", [False, True])
-    def test_descending(self, descending):
-        a = numpy.repeat(numpy.arange(10), 10)
+    @pytest.mark.parametrize(
+        "dtype", get_integer_dtypes(all_int_types=True) + [dpnp.bool]
+    )
+    def test_descending_duplicates(self, dtype, descending):
+        # a stable argsort keeps the original relative order of equal
+        # elements in both ascending and descending order
+        if dtype == dpnp.bool:
+            values = [False, True]
+        else:
+            info = numpy.iinfo(dtype)
+            values = [info.min, 1, info.max]
+        a = numpy.array(values * 2, dtype=dtype)
         ia = dpnp.array(a)
 
         result = dpnp.argsort(ia, descending=descending)
-        if not descending:
-            expected = numpy.argsort(a, kind="stable")
-        else:
-            expected = numpy.flip(numpy.argsort(numpy.flip(a), kind="stable"))
-            expected = (a.shape[0] - 1) - expected
-        assert_array_equal(result, expected)
-
-        # test ndarray method
-        result = ia.argsort(descending=descending)
-        if not descending:
-            expected = a.argsort(kind="stable")
-        else:
-            a = numpy.flip(a)
-            expected = numpy.flip(a.argsort(kind="stable"))
-            expected = (a.shape[0] - 1) - expected
+        expected = numpy.argsort(a, stable=True, descending=descending)
         assert_array_equal(result, expected)
+        assert_array_equal(dpnp.sort(ia, descending=descending), a[expected])
 
     # `stable` keyword is supported in numpy 2.0 and above
     @testing.with_requires("numpy>=2.0")
@@ -545,24 +543,57 @@ def test_kind(self, kind):
         expected = numpy.sort(a, kind="stable")
         assert_array_equal(result, expected)
 
+    @testing.with_requires("numpy>=2.5")
     @pytest.mark.parametrize("descending", [False, True])
     def test_descending(self, descending):
         a = numpy.repeat(numpy.arange(10), 10)
         ia = dpnp.array(a)
 
         result = dpnp.sort(ia, descending=descending)
-        expected = numpy.sort(a, kind="stable")
-        if descending:
-            expected = numpy.flip(expected)
+        expected = numpy.sort(a, stable=True, descending=descending)
         assert_array_equal(result, expected)
 
         # test ndarray method
         ia.sort(descending=descending)
-        a.sort(kind="stable")
-        if descending:
-            a = numpy.flip(a)
+        a.sort(stable=True, descending=descending)
         assert_array_equal(ia, a)
 
+    @testing.with_requires("numpy>=2.5")
+    @pytest.mark.parametrize("descending", [False, True])
+    @pytest.mark.parametrize("dtype", get_float_dtypes(no_float16=False))
+    def test_descending_nan(self, dtype, descending):
+        # NaNs are sorted to the end for both ascending and descending order
+        a = numpy.linspace(-50, 50, 101).astype(dtype)
+        a[::10] = numpy.nan
+        ia = dpnp.array(a)
+
+        result = dpnp.sort(ia, descending=descending)
+        expected = numpy.sort(a, stable=True, descending=descending)
+        assert_array_equal(result, expected)
+
+    @testing.with_requires("numpy>=2.5")
+    @pytest.mark.parametrize("descending", [False, True])
+    @pytest.mark.parametrize("dtype", get_complex_dtypes())
+    def test_descending_complex_nan(self, dtype, descending):
+        # NaN-containing complex values sort to the end in groups
+        # (no nan) -> (imag nan) -> (real nan) -> (all nan) for both orders;
+        # finite values keep lexicographic order (real part more significant)
+        arange = numpy.tile(numpy.arange(25), 4)
+        no_nans = arange + 1j * arange
+        im_nans = arange + complex(0, numpy.nan)
+        re_nans = complex(numpy.nan, 0) + 1j * arange
+        all_nans = numpy.full(100, complex(numpy.nan, numpy.nan))
+        a = numpy.concatenate((no_nans, im_nans, re_nans, all_nans))
+        a = a.astype(dtype)
+
+        rng = numpy.random.default_rng(0)
+        rng.shuffle(a)
+        ia = dpnp.array(a)
+
+        result = dpnp.sort(ia, descending=descending)
+        expected = numpy.sort(a, stable=True, descending=descending)
+        assert_array_equal(result, expected)
+
     # `stable` keyword is supported in numpy 2.0 and above
     @testing.with_requires("numpy>=2.0")
     @pytest.mark.parametrize("stable", [None, False, True])

From 8383540bc312205498be81dee78a37dea5ad250a Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 17:32:54 +0200
Subject: [PATCH 04/12] Add PR reference to the changelog entry

---
 CHANGELOG.md | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/CHANGELOG.md b/CHANGELOG.md
index c30e2af86895..30fc1cb2420a 100644
--- a/CHANGELOG.md
+++ b/CHANGELOG.md
@@ -99,7 +99,7 @@ This release is compatible with NumPy 2.5.
 * Fixed `dpnp.cumsum`, `dpnp.cumprod`, and their `nan`/`cumulative_*` variants (including `dpnp.tensor.cumulative_sum`/`cumulative_prod`) silently returning incorrect results when accumulating along an axis of an array with more than one row [#3063](https://github.com/IntelPython/dpnp/pull/3063)
 * Fixed `dpnp.einsum` returning a result whose memory layout differs from NumPy for the default `order="K"`, and ignoring `out` and `order` for a contraction over a size-0 dimension [#3058](https://github.com/IntelPython/dpnp/pull/3058)
 * Fixed operations on a boolean array whose bytes are not `0x00`/`0x01` [#3055](https://github.com/IntelPython/dpnp/pull/3055)
-* Fixed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts placing `NaN` values first instead of last when sorting in descending order
+* Fixed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts placing `NaN` values first instead of last when sorting in descending order [#3066](https://github.com/IntelPython/dpnp/pull/3066)
 
 ### Security
 

From 7e8cadf0c3f482d8a6e64d364aca8ca9b2a167a5 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 17:38:03 +0200
Subject: [PATCH 05/12] Document NaN ordering for descending sort/argsort

Note in the `dpnp` and `dpnp.tensor` sort/argsort docstrings that NaN
values (and complex values with a NaN component) are ordered to the end
regardless of the `descending` flag.
---
 dpnp/dpnp_iface_sorting.py | 6 ++++--
 dpnp/tensor/_sorting.py    | 8 ++++++--
 2 files changed, 10 insertions(+), 4 deletions(-)

diff --git a/dpnp/dpnp_iface_sorting.py b/dpnp/dpnp_iface_sorting.py
index 5feffffbc432..7aaf2852ec57 100644
--- a/dpnp/dpnp_iface_sorting.py
+++ b/dpnp/dpnp_iface_sorting.py
@@ -119,7 +119,8 @@ def argsort(
     descending : bool, optional
         Sort order. If ``True``, the array must be sorted in descending order
         (by value). If ``False``, the array must be sorted in ascending order
-        (by value).
+        (by value). NaN values (and complex values with a NaN component) are
+        ordered to the end regardless of `descending`.
 
         Default: ``False``.
     stable : {None, bool}, optional
@@ -348,7 +349,8 @@ def sort(a, axis=-1, kind=None, order=None, *, descending=False, stable=None):
     descending : bool, optional
         Sort order. If ``True``, the array must be sorted in descending order
         (by value). If ``False``, the array must be sorted in ascending order
-        (by value).
+        (by value). NaN values (and complex values with a NaN component) are
+        ordered to the end regardless of `descending`.
 
         Default: ``False``.
     stable : {None, bool}, optional
diff --git a/dpnp/tensor/_sorting.py b/dpnp/tensor/_sorting.py
index c912b4f77cdf..6d9476c5e802 100644
--- a/dpnp/tensor/_sorting.py
+++ b/dpnp/tensor/_sorting.py
@@ -73,7 +73,9 @@ def sort(x, /, *, axis=-1, descending=False, stable=True, kind=None):
         descending (Optional[bool]):
             sort order. If `True`, the array must be sorted in descending
             order (by value). If `False`, the array must be sorted in
-            ascending order (by value). Default: `False`.
+            ascending order (by value). NaN values (and complex values with
+            a NaN component) are ordered to the end regardless of
+            `descending`. Default: `False`.
         stable (Optional[bool]):
             sort stability. If `True`, the returned array must maintain the
             relative order of `x` values which compare as equal. If `False`,
@@ -185,7 +187,9 @@ def argsort(x, axis=-1, descending=False, stable=True, kind=None):
         descending (Optional[bool]):
             sort order. If `True`, the array must be sorted in descending
             order (by value). If `False`, the array must be sorted in
-            ascending order (by value). Default: `False`.
+            ascending order (by value). NaN values (and complex values with
+            a NaN component) are ordered to the end regardless of
+            `descending`. Default: `False`.
         stable (Optional[bool]):
             sort stability. If `True`, the returned array must maintain the
             relative order of `x` values which compare as equal. If `False`,

From a4079d41f6a4b20e00693b91311cca46d0c200e0 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 17:57:51 +0200
Subject: [PATCH 06/12] Cover top_k NaN ordering under the descending sort fix

The descending sort fix routes float and complex `top_k(mode="largest")`
through the same comparator, so NaN values (and complex values with a NaN
component) are now treated as the smallest and no longer surface ahead of
finite values. Add a `dpnp.tensor.top_k` NaN test, document the behavior
in the `top_k` docstring, and add a changelog note.
---
 CHANGELOG.md                                |  3 +-
 dpnp/tensor/_sorting.py                     |  4 +++
 dpnp/tests/tensor/test_usm_ndarray_top_k.py | 33 +++++++++++++++++++++
 3 files changed, 39 insertions(+), 1 deletion(-)

diff --git a/CHANGELOG.md b/CHANGELOG.md
index 30fc1cb2420a..c0a18d079b33 100644
--- a/CHANGELOG.md
+++ b/CHANGELOG.md
@@ -48,6 +48,8 @@ This release is compatible with NumPy 2.5.
 * `dpnp` uses pybind11 3.1.0 [#3015](https://github.com/IntelPython/dpnp/pull/3015)
 * Reworked the ASV benchmarks and added end-to-end workload benchmarks derived from dpBench [#2996](https://github.com/IntelPython/dpnp/pull/2996)
 * Reduced allocations in `dpnp.linalg.norm` by reusing the reduction result as the `sqrt` output buffer in the 2-norm and Frobenius-norm branches [#3062](https://github.com/IntelPython/dpnp/pull/3062)
+* Changed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts placing `NaN` values first instead of last when sorting in descending order [#3066](https://github.com/IntelPython/dpnp/pull/3066)
+* `dpnp.tensor.top_k` now treats `NaN` values (and complex values with a `NaN` component) as the smallest, consistent with `dpnp.tensor.sort`, so `mode="largest"` no longer returns them ahead of finite values [#3066](https://github.com/IntelPython/dpnp/pull/3066)
 
 ### Deprecated
 
@@ -99,7 +101,6 @@ This release is compatible with NumPy 2.5.
 * Fixed `dpnp.cumsum`, `dpnp.cumprod`, and their `nan`/`cumulative_*` variants (including `dpnp.tensor.cumulative_sum`/`cumulative_prod`) silently returning incorrect results when accumulating along an axis of an array with more than one row [#3063](https://github.com/IntelPython/dpnp/pull/3063)
 * Fixed `dpnp.einsum` returning a result whose memory layout differs from NumPy for the default `order="K"`, and ignoring `out` and `order` for a contraction over a size-0 dimension [#3058](https://github.com/IntelPython/dpnp/pull/3058)
 * Fixed operations on a boolean array whose bytes are not `0x00`/`0x01` [#3055](https://github.com/IntelPython/dpnp/pull/3055)
-* Fixed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts placing `NaN` values first instead of last when sorting in descending order [#3066](https://github.com/IntelPython/dpnp/pull/3066)
 
 ### Security
 
diff --git a/dpnp/tensor/_sorting.py b/dpnp/tensor/_sorting.py
index 6d9476c5e802..a78eb8585f91 100644
--- a/dpnp/tensor/_sorting.py
+++ b/dpnp/tensor/_sorting.py
@@ -319,6 +319,10 @@ def top_k(x, k, /, *, axis=None, mode="largest"):
             - `"largest"`: return the `k` largest elements.
             - `"smallest"`: return the `k` smallest elements.
 
+            NaN values (and complex values with a NaN component) are treated
+            as the smallest, so `"largest"` does not return them ahead of
+            finite values.
+
             Default: `"largest"`.
 
     Returns:
diff --git a/dpnp/tests/tensor/test_usm_ndarray_top_k.py b/dpnp/tests/tensor/test_usm_ndarray_top_k.py
index 1c04c1fff57a..4a7be3149aaf 100644
--- a/dpnp/tests/tensor/test_usm_ndarray_top_k.py
+++ b/dpnp/tests/tensor/test_usm_ndarray_top_k.py
@@ -179,6 +179,39 @@ def test_top_k_1d_smallest(dtype, n):
     assert dpt.all(s.indices == expected_inds), (s.indices, expected_inds)
 
 
+@pytest.mark.parametrize("dtype", ["f2", "f4", "f8", "c8", "c16"])
+@pytest.mark.parametrize("mode", ["largest", "smallest"])
+def test_top_k_nan(dtype, mode):
+    # NaNs (and complex values with a NaN component) are ordered to the end
+    # for both modes, so top_k excludes them until k reaches the NaN region
+    q = get_queue_or_skip()
+    skip_if_dtype_not_supported(dtype, q)
+
+    is_complex = dtype in ("c8", "c16")
+    nan = complex(dpt.nan, dpt.nan) if is_complex else dpt.nan
+
+    def has_nan(a):
+        if is_complex:
+            return dpt.any(dpt.isnan(dpt.real(a)) | dpt.isnan(dpt.imag(a)))
+        return dpt.any(dpt.isnan(a))
+
+    # 5 distinct finite values followed by 2 NaNs, then rolled to interleave
+    x = dpt.roll(dpt.asarray([3, 1, 5, 2, 4, nan, nan], dtype=dtype), 3)
+
+    # k within the finite region: NaNs are excluded from the result
+    r = dpt.top_k(x, 3, mode=mode)
+    assert not has_nan(r.values)
+    assert dpt.all(r.values == x[r.indices])
+    extreme = [5, 4, 3] if mode == "largest" else [1, 2, 3]
+    expected = dpt.asarray(extreme, dtype=dtype)
+    assert dpt.all(dpt.sort(r.values) == dpt.sort(expected))
+
+    # k reaching into the NaN region: the 2 NaNs are ordered last
+    r = dpt.top_k(x, 7, mode=mode)
+    assert has_nan(r.values[-2:])
+    assert not has_nan(r.values[:5])
+
+
 @pytest.mark.parametrize(
     "dtype",
     [

From ce730165956d182f2a73fbe6c71c68b61a9f6199 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Thu, 10 Sep 2026 18:06:49 +0200
Subject: [PATCH 07/12] Cover descending complex NaN ordering in dpnp.tensor
 sort test

Extend `test_sort_complex_fp_nan` to also sort with `descending=True` and
compare against NumPy, so the dpnp.tensor layer exercises the complex
NaN-at-end ordering for both directions. Guarded on numpy>=2.5.
---
 dpnp/tests/tensor/test_usm_ndarray_sorting.py | 16 ++++++++++++++++
 1 file changed, 16 insertions(+)

diff --git a/dpnp/tests/tensor/test_usm_ndarray_sorting.py b/dpnp/tests/tensor/test_usm_ndarray_sorting.py
index 9917f93bdca1..7311dea78358 100644
--- a/dpnp/tests/tensor/test_usm_ndarray_sorting.py
+++ b/dpnp/tests/tensor/test_usm_ndarray_sorting.py
@@ -33,6 +33,7 @@
 from numpy.testing import assert_array_equal
 
 import dpnp.tensor as dpt
+from dpnp.tests.helper import numpy_version
 
 from .helper import (
     get_queue_or_skip,
@@ -357,6 +358,21 @@ def test_sort_complex_fp_nan(dtype):
             r1.view(np.int64), r2.view(np.int64)
         ), f"Failed for {i} and {j}"
 
+    # complex values with a NaN component sort to the end for descending
+    # order too, matching NumPy (`descending` requires numpy>=2.5)
+    if numpy_version() >= "2.5.0":
+        s = dpt.sort(inp, descending=True)
+        expected = np.sort(dpt.asnumpy(inp), descending=True)
+        assert np.allclose(dpt.asnumpy(s), expected, equal_nan=True)
+
+        m1 = dpt.asnumpy(dpt.sort(sub_arrs, axis=1, descending=True))
+        m2 = np.sort(dpt.asnumpy(sub_arrs), axis=1, descending=True)
+        for k in range(len(pairs)):
+            i, j = pairs[k]
+            assert np.array_equal(
+                m1[k].view(np.int64), m2[k].view(np.int64)
+            ), f"Failed for {i} and {j}"
+
 
 def test_radix_sort_size_1_axis():
     get_queue_or_skip()

From 68af50f4621e852b86fabd997357ed913fcd490d Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Fri, 11 Sep 2026 14:46:03 +0200
Subject: [PATCH 08/12] Clarify changelog wording for descending NaN ordering

The entry now states the new behavior (NaN values placed last instead of
first) rather than reading as if they are placed first.
---
 CHANGELOG.md | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/CHANGELOG.md b/CHANGELOG.md
index c0a18d079b33..bf5d16cd1979 100644
--- a/CHANGELOG.md
+++ b/CHANGELOG.md
@@ -48,7 +48,7 @@ This release is compatible with NumPy 2.5.
 * `dpnp` uses pybind11 3.1.0 [#3015](https://github.com/IntelPython/dpnp/pull/3015)
 * Reworked the ASV benchmarks and added end-to-end workload benchmarks derived from dpBench [#2996](https://github.com/IntelPython/dpnp/pull/2996)
 * Reduced allocations in `dpnp.linalg.norm` by reusing the reduction result as the `sqrt` output buffer in the 2-norm and Frobenius-norm branches [#3062](https://github.com/IntelPython/dpnp/pull/3062)
-* Changed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts placing `NaN` values first instead of last when sorting in descending order [#3066](https://github.com/IntelPython/dpnp/pull/3066)
+* Changed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts to place `NaN` values last instead of first when sorting in descending order [#3066](https://github.com/IntelPython/dpnp/pull/3066)
 * `dpnp.tensor.top_k` now treats `NaN` values (and complex values with a `NaN` component) as the smallest, consistent with `dpnp.tensor.sort`, so `mode="largest"` no longer returns them ahead of finite values [#3066](https://github.com/IntelPython/dpnp/pull/3066)
 
 ### Deprecated

From e385256ca456f9313d4a4e67e8b877af33fb09dd Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Fri, 11 Sep 2026 14:52:08 +0200
Subject: [PATCH 09/12] Correct top_k NaN wording in changelog and docstring

NaN values are ordered to the end for both modes, so describing them as
"the smallest" was inaccurate for mode="smallest". Scope the changelog
entry to mode="largest" (the only behavior that changed) and state in the
top_k docstring that NaNs are ordered last for both modes.
---
 CHANGELOG.md            | 2 +-
 dpnp/tensor/_sorting.py | 6 +++---
 2 files changed, 4 insertions(+), 4 deletions(-)

diff --git a/CHANGELOG.md b/CHANGELOG.md
index bf5d16cd1979..579da1496bee 100644
--- a/CHANGELOG.md
+++ b/CHANGELOG.md
@@ -49,7 +49,7 @@ This release is compatible with NumPy 2.5.
 * Reworked the ASV benchmarks and added end-to-end workload benchmarks derived from dpBench [#2996](https://github.com/IntelPython/dpnp/pull/2996)
 * Reduced allocations in `dpnp.linalg.norm` by reusing the reduction result as the `sqrt` output buffer in the 2-norm and Frobenius-norm branches [#3062](https://github.com/IntelPython/dpnp/pull/3062)
 * Changed `dpnp.sort`, `dpnp.argsort`, and their `dpnp.ndarray`/`dpnp.tensor` counterparts to place `NaN` values last instead of first when sorting in descending order [#3066](https://github.com/IntelPython/dpnp/pull/3066)
-* `dpnp.tensor.top_k` now treats `NaN` values (and complex values with a `NaN` component) as the smallest, consistent with `dpnp.tensor.sort`, so `mode="largest"` no longer returns them ahead of finite values [#3066](https://github.com/IntelPython/dpnp/pull/3066)
+* Changed `dpnp.tensor.top_k` with `mode="largest"` to no longer return `NaN` values (or complex values with a `NaN` component) ahead of finite values, matching the `NaN`-last order of `dpnp.tensor.sort` [#3066](https://github.com/IntelPython/dpnp/pull/3066)
 
 ### Deprecated
 
diff --git a/dpnp/tensor/_sorting.py b/dpnp/tensor/_sorting.py
index a78eb8585f91..e2e6654c5c1f 100644
--- a/dpnp/tensor/_sorting.py
+++ b/dpnp/tensor/_sorting.py
@@ -319,9 +319,9 @@ def top_k(x, k, /, *, axis=None, mode="largest"):
             - `"largest"`: return the `k` largest elements.
             - `"smallest"`: return the `k` smallest elements.
 
-            NaN values (and complex values with a NaN component) are treated
-            as the smallest, so `"largest"` does not return them ahead of
-            finite values.
+            NaN values (and complex values with a NaN component) are ordered
+            last for both modes, so they are not returned ahead of finite
+            values.
 
             Default: `"largest"`.
 

From 808365e7f040126dfd74787a444125779292a563 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Fri, 11 Sep 2026 14:55:24 +0200
Subject: [PATCH 10/12] Document NaN ordering on dpnp.ndarray sort/argsort
 methods

Add the same NaN-ordering note to the `dpnp.ndarray.sort` and
`dpnp.ndarray.argsort` docstrings that the module-level `dpnp.sort` and
`dpnp.argsort` already carry.
---
 dpnp/dpnp_array.py | 6 ++++--
 1 file changed, 4 insertions(+), 2 deletions(-)

diff --git a/dpnp/dpnp_array.py b/dpnp/dpnp_array.py
index b225fb2c7329..b346ef064f93 100644
--- a/dpnp/dpnp_array.py
+++ b/dpnp/dpnp_array.py
@@ -898,7 +898,8 @@ def argsort(
         descending : bool, optional
             Sort order. If ``True``, the array must be sorted in descending
             order (by value). If ``False``, the array must be sorted in
-            ascending order (by value).
+            ascending order (by value). NaN values (and complex values with a
+            NaN component) are ordered to the end regardless of `descending`.
 
             Default: ``False``.
         stable : {None, bool}, optional
@@ -1967,7 +1968,8 @@ def sort(
         descending : bool, optional
             Sort order. If ``True``, the array must be sorted in descending
             order (by value). If ``False``, the array must be sorted in
-            ascending order (by value).
+            ascending order (by value). NaN values (and complex values with a
+            NaN component) are ordered to the end regardless of `descending`.
 
             Default: ``False``.
         stable : {None, bool}, optional

From 490c858ea225bd946556a2409323f81878e972b0 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Fri, 11 Sep 2026 15:00:40 +0200
Subject: [PATCH 11/12] Exercise all sort kinds in descending NaN float test

Parametrize `test_descending_nan` over `kind` so the descending NaN-at-end
behavior is validated for the radix-sort path as well as merge sort.
---
 dpnp/tests/test_sort.py | 5 +++--
 1 file changed, 3 insertions(+), 2 deletions(-)

diff --git a/dpnp/tests/test_sort.py b/dpnp/tests/test_sort.py
index cda4ec81aa58..3834d488b8a7 100644
--- a/dpnp/tests/test_sort.py
+++ b/dpnp/tests/test_sort.py
@@ -559,15 +559,16 @@ def test_descending(self, descending):
         assert_array_equal(ia, a)
 
     @testing.with_requires("numpy>=2.5")
+    @pytest.mark.parametrize("kind", [None, "stable", "mergesort", "radixsort"])
     @pytest.mark.parametrize("descending", [False, True])
     @pytest.mark.parametrize("dtype", get_float_dtypes(no_float16=False))
-    def test_descending_nan(self, dtype, descending):
+    def test_descending_nan(self, dtype, descending, kind):
         # NaNs are sorted to the end for both ascending and descending order
         a = numpy.linspace(-50, 50, 101).astype(dtype)
         a[::10] = numpy.nan
         ia = dpnp.array(a)
 
-        result = dpnp.sort(ia, descending=descending)
+        result = dpnp.sort(ia, descending=descending, kind=kind)
         expected = numpy.sort(a, stable=True, descending=descending)
         assert_array_equal(result, expected)
 

From f709514e101d98bf2275cf7f0d55ededb1e75249 Mon Sep 17 00:00:00 2001
From: Anton Volkov <antonwolfy@gmail.com>
Date: Fri, 11 Sep 2026 15:45:59 +0200
Subject: [PATCH 12/12] Simplify descending complex comparator via operand
 negation

Implement `ExtendedComplexFPGreater` as `ExtendedComplexFPLess{}(-v1, -v2)`
instead of duplicating the comparison logic with operands swapped. Negation
preserves NaN-ness, so the NaN-based grouping (and its ordering to the end)
is unchanged while the finite-component comparison is reversed.
---
 .../include/utils/rich_comparisons.hpp        | 33 ++-----------------
 1 file changed, 3 insertions(+), 30 deletions(-)

diff --git a/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp b/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp
index 93029b9d4107..0886de7451c4 100644
--- a/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp
+++ b/dpnp/tensor/libtensor/include/utils/rich_comparisons.hpp
@@ -107,38 +107,11 @@ struct ExtendedComplexFPLess
 template <typename cT>
 struct ExtendedComplexFPGreater
 {
-    /* [(R, R), (R, nan), (nan, R), (nan, nan)] — NaN-containing values keep
-       the same trailing groups as ascending order; only the finite-component
-       comparison within a group is reversed */
+    /* Negating both operands reverses the finite comparison but preserves
+       NaN-ness, so NaN groups stay ordered to the end. */
     bool operator()(const cT &v1, const cT &v2) const
     {
-        using realT = typename cT::value_type;
-
-        const realT real1 = std::real(v1);
-        const realT real2 = std::real(v2);
-
-        const bool r1_nan = std::isnan(real1);
-        const bool r2_nan = std::isnan(real2);
-
-        const realT imag1 = std::imag(v1);
-        const realT imag2 = std::imag(v2);
-
-        const bool i1_nan = std::isnan(imag1);
-        const bool i2_nan = std::isnan(imag2);
-
-        const int idx1 = ((r1_nan) ? 2 : 0) + ((i1_nan) ? 1 : 0);
-        const int idx2 = ((r2_nan) ? 2 : 0) + ((i2_nan) ? 1 : 0);
-
-        const bool res =
-            !(r1_nan && i1_nan) &&
-            ((idx1 < idx2) ||
-             ((idx1 == idx2) &&
-              ((r1_nan && !i1_nan && (imag2 < imag1)) ||
-               (!r1_nan && i1_nan && (real2 < real1)) ||
-               (!r1_nan && !i1_nan &&
-                ((real2 < real1) || (!(real1 < real2) && (imag2 < imag1)))))));
-
-        return res;
+        return ExtendedComplexFPLess<cT>{}(-v1, -v2);
     }
 };