Dppy vs numba debug test

numba_dppy/examples/_helper.py

@@ -13,16 +13,204 @@
 # See the License for the specific language governing permissions and
 # limitations under the License.
 
+import contextlib
+import unittest
+import os
+import pathlib
+import re
+
 import dpctl
+import numba_dppy
+
+from subprocess import Popen, PIPE
+from numba.tests.support import captured_stdout
+
+from numba_dppy import config
 
 
-def has_gpu(backend="opencl"):
+def has_gpu_queues(backend="opencl"):
+    """
+    Checks if dpctl is able to select a GPU device that defaults to
+    an OpenCL GPU.
+    """
     return bool(dpctl.get_num_devices(backend=backend, device_type="gpu"))
 
 
-def has_cpu(backend="opencl"):
+def has_cpu_queues(backend="opencl"):
+    """
+    Checks if dpctl is able to select a CPU device that defaults to
+    an OpenCL CPU.
+    """
     return bool(dpctl.get_num_devices(backend=backend, device_type="cpu"))
 
 
 def has_sycl_platforms():
-    return bool(len(dpctl.get_platforms()))
+    """
+    Checks if dpctl is able to identify a non-host SYCL platform.
+    """
+    platforms = dpctl.get_platforms()
+    for p in platforms:
+        if p.backend is not dpctl.backend_type.host:
+            return True
+    return False
+
+
+def is_gen12(device_type):
+    with dpctl.device_context(device_type):
+        q = dpctl.get_current_queue()
+        device = q.get_sycl_device()
+        name = device.name
+        if "Gen12" in name:
+            return True
+
+        return False
+
+
+def platform_not_supported(device_type):
+    import platform
+
+    platform = platform.system()
+    device = device_type.split(":")[0]
+
+    if device == "level_zero" and platform == "Windows":
+        return True
+
+    return False
+
+
+def skip_test(device_type):
+    skip = False
+    try:
+        with dpctl.device_context(device_type):
+            pass
+    except Exception:
+        skip = True
+
+    if not skip:
+        if platform_not_supported(device_type):
+            skip = True
+
+    return skip
+
+
+@contextlib.contextmanager
+def override_config(name, value, config=config):
+    """
+    Extends `numba/tests/support.py:override_config()` with argument `config`
+    which is `numba_dppy.config` by default.
+    """
+    old_value = getattr(config, name)
+    setattr(config, name, value)
+    try:
+        yield
+    finally:
+        setattr(config, name, old_value)
+
+
+def _id(obj):
+    return obj
+
+
+def expectedFailureIf(condition):
+    """
+    Expected failure for a test if the condition is true.
+    """
+    if condition:
+        return unittest.expectedFailure
+    return _id
+
+
+def ensure_dpnp():
+    try:
+        from numba_dppy.dpnp_glue import dpnp_fptr_interface as dpnp_glue
+
+        return True
+    except:
+        return False
+
+
+@contextlib.contextmanager
+def dpnp_debug():
+    import numba_dppy.dpnp_glue as dpnp_lowering
+
+    old, dpnp_lowering.DEBUG = dpnp_lowering.DEBUG, 1
+    yield
+    dpnp_lowering.DEBUG = old
+
+
+@contextlib.contextmanager
+def assert_dpnp_implementaion():
+    from numba.tests.support import captured_stdout
+
+    with captured_stdout() as stdout, dpnp_debug():
+        yield
+
+    assert "dpnp implementation" in stdout.getvalue(), "dpnp implementation is not used"
+
+
+@contextlib.contextmanager
+def assert_auto_offloading(parfor_offloaded=1, parfor_offloaded_failure=0):
+    """
+    If ``parfor_offloaded`` is not provided this context_manager
+    will check for 1 occurrance of success message. Developers
+    can always specify how many parfor offload success message
+    is expected.
+    If ``parfor_offloaded_failure`` is not provided the default
+    behavior is to expect 0 failure message, in other words, we
+    expect all parfors present in the code to be successfully
+    offloaded to GPU.
+    """
+    old_debug = config.DEBUG
+    config.DEBUG = 1
+
+    with captured_stdout() as stdout:
+        yield
+
+    config.DEBUG = old_debug
+
+    got_parfor_offloaded = stdout.getvalue().count("Parfor offloaded to")
+    assert parfor_offloaded == got_parfor_offloaded, (
+            "Expected %d parfor(s) to be auto offloaded, instead got %d parfor(s) auto offloaded"
+            % (parfor_offloaded, got_parfor_offloaded)
+    )
+
+    got_parfor_offloaded_failure = stdout.getvalue().count(
+        "Failed to offload parfor to"
+    )
+    assert parfor_offloaded_failure == got_parfor_offloaded_failure, (
+            "Expected %d parfor(s) to be not auto offloaded, instead got %d parfor(s) not auto offloaded"
+            % (parfor_offloaded_failure, got_parfor_offloaded_failure)
+    )
+
+
+def make_check(text, val_to_search):
+    m = re.search(val_to_search, text, re.I)
+    got = m is not None
+    return got
+
+
+@contextlib.contextmanager
+def run_debug_command(command_name):
+    process_env = os.environ.copy()
+    process_env["NUMBA_OPT"] = "0"
+
+    run_dir = pathlib.Path(numba_dppy.__file__).parents[0] / 'examples/debug/'
+    command_path = run_dir / "commands" / command_name
+
+    process = Popen(
+        [
+            "gdb-oneapi",
+            "-q",
+            "-command",
+            command_path,
+            "python",
+        ],
+        stdout=PIPE,
+        stderr=PIPE,
+        env=process_env,
+        cwd=run_dir,
+    )
+    (output, err) = process.communicate()
+    process.wait()
+
+    yield str(output)

numba_dppy/examples/debug/commands/dppy_numba_jit

@@ -0,0 +1,21 @@
+# Run: NUMBA_OPT=0 gdb-oneapi -q -command commands/dppy_numba_jit python
+set trace-commands on
+set pagination off
+set breakpoint pending on
+break dppy_numba_basic.py:24
+run dppy_numba_basic.py
+# Expected:
+# ...
+# Thread 1.1 "python" hit Breakpoint 1, __main__::func$242 () at dppy_numba.py:24
+# 24          param_c = param_a + 10  # Set breakpoint
+continue
+# Expected:
+# Thread 1.2 "python" hit Breakpoint 1, __main__::func$242 () at dppy_numba.py:24
+# 24          param_c = param_a + 10  # Set breakpoint
+continue
+# Expected:
+# ...
+# Done...
+continue
+echo Done\n
+quit

numba_dppy/examples/debug/commands/dppy_numba_kernel

@@ -0,0 +1,20 @@
+# Run: NUMBA_OPT=0 gdb-oneapi -q -command commands/dppy_numba_kernel python
+set trace-commands on
+set pagination off
+set breakpoint pending on
+break dppy_numba_basic.py:24
+run dppy_numba_basic.py --dppy
+# Expected:
+# ...
+# Thread 2.1 hit Breakpoint 1, with SIMD lanes [0-7], __main__::func () at dppy_numba.py:24
+# 24          param_c = param_a + 10  # Set breakpoint
+continue
+# Expected:
+# Thread 2.2 hit Breakpoint 1, with SIMD lanes [0-7], __main__::func () at dppy_numba.py:24
+# 24          param_c = param_a + 10  # Set breakpoint
+continue
+# Expected:
+# ...
+# Done...
+echo Done\n
+quit

numba_dppy/examples/debug/dppy_numba_basic.py

@@ -0,0 +1,74 @@
+# Copyright 2020, 2021 Intel Corporation
+#
+# Licensed under the Apache License, Version 2.0 (the "License");
+# you may not use this file except in compliance with the License.
+# You may obtain a copy of the License at
+#
+#      http://www.apache.org/licenses/LICENSE-2.0
+#
+# Unless required by applicable law or agreed to in writing, software
+# distributed under the License is distributed on an "AS IS" BASIS,
+# WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+# See the License for the specific language governing permissions and
+# limitations under the License.
+
+import dpctl
+import numpy as np
+import numba
+import argparse
+
+import numba_dppy as dppy
+
+
+def func(param_a, param_b):
+    param_c = param_a + 10  # Set breakpoint
+    param_d = param_b * 0.5
+    result = param_c + param_d
+    return result
+
+
+dppy_func = dppy.func(debug=True)(func)
+numba_func = numba.njit(debug=True)(func)
+
+
+@dppy.kernel(debug=True)
+def dppy_kernel(a_in_kernel, b_in_kernel, c_in_kernel):
+    i = dppy.get_global_id(0)
+    c_in_kernel[i] = dppy_func(a_in_kernel[i], b_in_kernel[i])
+
+
+@numba.njit(debug=True)
+def numba_func_driver(a, b, c):
+    for i in range(len(c)):
+        c[i] = numba_func(a[i], b[i])
+
+
+def main():
+    parser = argparse.ArgumentParser()
+    parser.add_argument('--dppy',
+                        required=False,
+                        action='store_true',
+                        help="Start the dppy version of functions",
+    )
+
+    args = parser.parse_args()
+
+    global_size = 10
+    N = global_size
+
+    a = np.arange(N, dtype=np.float32)
+    b = np.arange(N, dtype=np.float32)
+    c = np.empty_like(a)
+
+    if args.dppy:
+        device = dpctl.select_default_device()
+        with dppy.offload_to_sycl_device(device):
+            dppy_kernel[global_size, dppy.DEFAULT_LOCAL_SIZE](a, b, c)
+    else:
+        numba_func_driver(a, b, c)
+
+    print("Done...")
+
+
+if __name__ == "__main__":
+    main()

numba_dppy/tests/_helper.py

@@ -15,8 +15,13 @@
 
 import contextlib
 import unittest
+import os
+import re
 
 import dpctl
+import numba_dppy
+
+from subprocess import Popen, PIPE
 from numba.tests.support import captured_stdout
 
 from numba_dppy import config
@@ -164,14 +169,45 @@ def assert_auto_offloading(parfor_offloaded=1, parfor_offloaded_failure=0):
 
     got_parfor_offloaded = stdout.getvalue().count("Parfor offloaded to")
     assert parfor_offloaded == got_parfor_offloaded, (
-        "Expected %d parfor(s) to be auto offloaded, instead got %d parfor(s) auto offloaded"
-        % (parfor_offloaded, got_parfor_offloaded)
+            "Expected %d parfor(s) to be auto offloaded, instead got %d parfor(s) auto offloaded"
+            % (parfor_offloaded, got_parfor_offloaded)
     )
 
     got_parfor_offloaded_failure = stdout.getvalue().count(
         "Failed to offload parfor to"
     )
     assert parfor_offloaded_failure == got_parfor_offloaded_failure, (
-        "Expected %d parfor(s) to be not auto offloaded, instead got %d parfor(s) not auto offloaded"
-        % (parfor_offloaded_failure, got_parfor_offloaded_failure)
+            "Expected %d parfor(s) to be not auto offloaded, instead got %d parfor(s) not auto offloaded"
+            % (parfor_offloaded_failure, got_parfor_offloaded_failure)
+    )
+
+
+def make_check(text, val_to_search):
+    m = re.search(val_to_search, text, re.I)
+    got = m is not None
+    return got
+
+
+@contextlib.contextmanager
+def run_debug_command(command_name):
+    process_env = os.environ.copy()
+    process_env["NUMBA_OPT"] = "0"
+
+    command_path = os.path.dirname(os.path.abspath(numba_dppy.__file__)) + "/examples/debug/commands/" + command_name
+
+    process = Popen(
+        [
+            "gdb-oneapi",
+            "-q",
+            "-command",
+            command_path,
+            "python",
+        ],
+        stdout=PIPE,
+        stderr=PIPE,
+        env=process_env,
     )
+    (output, err) = process.communicate()
+    process.wait()
+
+    yield str(output)