Updates to kernel tests.

Author: Diptorup Deb

numba_dpex/tests/kernel_tests/test_func_qualname_disambiguation.py

@@ -2,13 +2,11 @@
 #
 # SPDX-License-Identifier: Apache-2.0
 
-import dpctl
 import dpctl.tensor as dpt
+import dpnp
 import numpy as np
-import pytest
 
 import numba_dpex as ndpx
-from numba_dpex.tests._helper import filter_strings
 
 
 def make_write_values_kernel(n_rows):
@@ -80,8 +78,7 @@ def write_values_inner(array_in, row_idx):
     return write_values_inner
 
 
-@pytest.mark.parametrize("offload_device", filter_strings)
-def test_qualname_basic(offload_device):
+def test_qualname_basic():
     """A basic test function to test
     qualified name disambiguation.
     """
@@ -92,32 +89,11 @@ def test_qualname_basic(offload_device):
         else:
             ans[i, 0] = 1
 
-    a = np.zeros((10, 10), dtype=dpt.int64)
-
-    device = dpctl.SyclDevice(offload_device)
-    queue = dpctl.SyclQueue(device)
-
-    da = dpt.usm_ndarray(
-        a.shape,
-        dtype=a.dtype,
-        buffer="device",
-        buffer_ctor_kwargs={"queue": queue},
-    )
-    da.usm_data.copy_from_host(a.reshape((-1)).view("|u1"))
+    a = dpnp.zeros((10, 10), dtype=dpt.int64)
 
     kernel = make_write_values_kernel(10)
-    kernel(da)
-
-    result = np.zeros_like(a)
-    da.usm_data.copy_to_host(result.reshape((-1)).view("|u1"))
-
-    print(ans)
-    print(result)
-
-    assert np.array_equal(result, ans)
+    kernel(a)
 
+    result = dpnp.zeros_like(a)
 
-if __name__ == "__main__":
-    test_qualname_basic("level_zero:gpu:0")
-    test_qualname_basic("opencl:gpu:0")
-    test_qualname_basic("opencl:cpu:0")
+    assert np.array_equal(dpnp.asnumpy(result), ans)

numba_dpex/tests/kernel_tests/test_func_specialization.py

@@ -2,7 +2,7 @@
 #
 # SPDX-License-Identifier: Apache-2.0
 
-import dpctl.tensor as dpt
+import dpnp
 import numpy as np
 import pytest
 
@@ -32,12 +32,12 @@ def kernel_function(a, b):
 
     k = dpex.kernel(kernel_function)
 
-    a = dpt.ones(N)
-    b = dpt.ones(N)
+    a = dpnp.ones(N)
+    b = dpnp.ones(N)
 
-    k[N](a, b)
+    k[dpex.Range(N)](a, b)
 
-    assert np.array_equal(dpt.asnumpy(b), dpt.asnumpy(a) + 1)
+    assert np.array_equal(dpnp.asnumpy(b), dpnp.asnumpy(a) + 1)
 
 
 def test_single_signature():
@@ -53,19 +53,19 @@ def kernel_function(a, b):
     k = dpex.kernel(kernel_function)
 
     # Test with int32, should work
-    a = dpt.ones(N, dtype=dpt.int32)
-    b = dpt.ones(N, dtype=dpt.int32)
+    a = dpnp.ones(N, dtype=dpnp.int32)
+    b = dpnp.ones(N, dtype=dpnp.int32)
 
-    k[N](a, b)
+    k[dpex.Range(N)](a, b)
 
-    assert np.array_equal(dpt.asnumpy(b), dpt.asnumpy(a) + 1)
+    assert np.array_equal(dpnp.asnumpy(b), dpnp.asnumpy(a) + 1)
 
     # Test with int64, should fail
-    a = dpt.ones(N, dtype=dpt.int64)
-    b = dpt.ones(N, dtype=dpt.int64)
+    a = dpnp.ones(N, dtype=dpnp.int64)
+    b = dpnp.ones(N, dtype=dpnp.int64)
 
     with pytest.raises(Exception) as e:
-        k[N](a, b)
+        k[dpex.Range(N)](a, b)
 
     assert " >>> <unknown function>(int64)" in e.value.args[0]
 
@@ -83,26 +83,26 @@ def kernel_function(a, b):
     k = dpex.kernel(kernel_function)
 
     # Test with int32, should work
-    a = dpt.ones(N, dtype=dpt.int32)
-    b = dpt.ones(N, dtype=dpt.int32)
+    a = dpnp.ones(N, dtype=dpnp.int32)
+    b = dpnp.ones(N, dtype=dpnp.int32)
 
-    k[N](a, b)
+    k[dpex.Range(N)](a, b)
 
-    assert np.array_equal(dpt.asnumpy(b), dpt.asnumpy(a) + 1)
+    assert np.array_equal(dpnp.asnumpy(b), dpnp.asnumpy(a) + 1)
 
     # Test with float32, should work
-    a = dpt.ones(N, dtype=dpt.float32)
-    b = dpt.ones(N, dtype=dpt.float32)
+    a = dpnp.ones(N, dtype=dpnp.float32)
+    b = dpnp.ones(N, dtype=dpnp.float32)
 
-    k[N](a, b)
+    k[dpex.Range(N)](a, b)
 
-    assert np.array_equal(dpt.asnumpy(b), dpt.asnumpy(a) + 1)
+    assert np.array_equal(dpnp.asnumpy(b), dpnp.asnumpy(a) + 1)
 
     # Test with int64, should fail
-    a = dpt.ones(N, dtype=dpt.int64)
-    b = dpt.ones(N, dtype=dpt.int64)
+    a = dpnp.ones(N, dtype=dpnp.int64)
+    b = dpnp.ones(N, dtype=dpnp.int64)
 
     with pytest.raises(Exception) as e:
-        k[N](a, b)
+        k[dpex.Range(N)](a, b)
 
     assert " >>> <unknown function>(int64)" in e.value.args[0]

numba_dpex/tests/kernel_tests/test_kernel_has_return_value_error.py

@@ -2,8 +2,7 @@
 #
 # SPDX-License-Identifier: Apache-2.0
 
-import dpctl.tensor as dpt
-import numpy as np
+import dpnp
 import pytest
 
 import numba_dpex as dpex
@@ -28,8 +27,8 @@ def sig(request):
 
 
 def test_return(sig):
-    a = dpt.arange(1024, dtype=dpt.int32, device="0")
+    a = dpnp.arange(1024, dtype=dpnp.int32)
 
     with pytest.raises(dpex.core.exceptions.KernelHasReturnValueError):
-        kernel = dpex.kernel(sig)(f)
-        kernel[a.size, dpex.DEFAULT_LOCAL_SIZE](a)
+        kernel_fn = dpex.kernel(sig)(f)
+        kernel_fn[dpex.Range(a.size)](a)

numba_dpex/tests/kernel_tests/test_math_functions.py

@@ -4,12 +4,11 @@
 
 import math
 
-import dpctl
-import numpy as np
+import dpnp
+import numpy
 import pytest
 
 import numba_dpex as dpex
-from numba_dpex.tests._helper import filter_strings
 
 list_of_unary_ops = ["fabs", "exp", "log", "sqrt", "sin", "cos", "tan"]
 
@@ -20,35 +19,35 @@ def unary_op(request):
 
 
 list_of_dtypes = [
-    np.float32,
-    np.float64,
+    dpnp.float32,
+    dpnp.float64,
 ]
 
 
 @pytest.fixture(params=list_of_dtypes)
 def input_arrays(request):
     # The size of input and out arrays to be used
     N = 2048
-    a = np.array(np.random.random(N), request.param)
-    b = np.array(np.random.random(N), request.param)
+    a = dpnp.arange(N, dtype=request.param)
+    b = dpnp.arange(N, dtype=request.param)
     return a, b
 
 
-@pytest.mark.parametrize("filter_str", filter_strings)
-def test_binary_ops(filter_str, unary_op, input_arrays):
-    a, actual = input_arrays
+def test_binary_ops(unary_op, input_arrays):
+    a, b = input_arrays
     uop = getattr(math, unary_op)
-    np_uop = getattr(np, unary_op)
+    dpnp_uop = getattr(dpnp, unary_op)
 
     @dpex.kernel
     def f(a, b):
         i = dpex.get_global_id(0)
         b[i] = uop(a[i])
 
-    device = dpctl.SyclDevice(filter_str)
-    with dpctl.device_context(device):
-        f[a.size, dpex.DEFAULT_LOCAL_SIZE](a, actual)
+    f[dpex.Range(a.size)](a, b)
 
-    expected = np_uop(a)
+    expected = dpnp_uop(a)
 
-    np.testing.assert_allclose(actual, expected, rtol=1e-5, atol=0)
+    np_expected = dpnp.asnumpy(expected)
+    np_actual = dpnp.asnumpy(b)
+
+    assert numpy.allclose(np_expected, np_actual)

numba_dpex/tests/kernel_tests/test_scalar_arg_types.py

@@ -2,8 +2,6 @@
 #
 # SPDX-License-Identifier: Apache-2.0
 
-import dpctl
-import dpctl.tensor as dpt
 import dpnp
 import numpy
 import pytest