remove and deprecate remaining references to program

Author: ndgrigorian

CHANGELOG.md

@@ -22,6 +22,7 @@ and this project adheres to [Semantic Versioning](https://semver.org/spec/v2.0.0
 
 ### Deprecated
 * Deprecated `dpctl.program` submodule in favor of `dpctl.compiler`, which provides a better description of the purpose of the submodule in exposing DPC++ compilation-related functionality [gh-2317](https://github.com/IntelPython/dpctl/pull/2317)
+* Deprecated `DPCTL_ENABLE_L0_PROGRAM_CREATION` CMake option in favor of `DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION` [gh-2317](https://github.com/IntelPython/dpctl/pull/2317)
 
 ### Maintenance
 * Updated pybind11 version used by `dpctl` and examples [gh-2357](https://github.com/IntelPython/dpctl/pull/2357)

docs/doc_sources/contributor_guides/building.rst

@@ -250,7 +250,7 @@ library.
         -DCMAKE_CXX_COMPILER=icpx                               \
         -DCMAKE_INSTALL_PREFIX=${INSTALL_PREFIX}                \
         -DCMAKE_PREFIX_PATH=${INSTALL_PREFIX}                   \
-        -DDPCTL_ENABLE_L0_PROGRAM_CREATION=ON                   \
+        -DDPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION=ON             \
         -DDPCTL_BUILD_CAPI_TESTS=ON                             \
         -DDPCTL_GENERATE_COVERAGE=ON                            \
         ..

dpctl/apis/include/dpctl4pybind11.hpp

@@ -88,7 +88,7 @@ class dpctl_capi
                               DPCTLSyclQueueRef,
                               PyObject *);
 
-    // program
+    // compiler
     DPCTLSyclKernelRef (*SyclKernel_GetKernelRef_)(PySyclKernelObject *);
     PySyclKernelObject *(*SyclKernel_Make_)(DPCTLSyclKernelRef, const char *);
 

dpctl/tests/test_sycl_program.py dpctl/tests/test_sycl_compiler.pydpctl/tests/test_sycl_program.py renamed to dpctl/tests/test_sycl_compiler.py

@@ -32,14 +32,14 @@ def get_spirv_abspath(fn):
     return spirv_file
 
 
-def _check_cpython_api_SyclKernelBundle_GetKernelBundleRef(sycl_prog):
+def _check_cpython_api_SyclKernelBundle_GetKernelBundleRef(kb):
     """Checks Cython-generated C-API function
-    `SyclKernelBundle_GetKernelBundleRef` defined in _program.pyx"""
+    `SyclKernelBundle_GetKernelBundleRef` defined in _compiler.pyx"""
     import ctypes
     import sys
 
-    assert type(sycl_prog) is dpc.SyclKernelBundle
-    mod = sys.modules[sycl_prog.__class__.__module__]
+    assert type(kb) is dpc.SyclKernelBundle
+    mod = sys.modules[kb.__class__.__module__]
     # get capsule storing SyclKernelBundle_GetKernelBundleRef function ptr
     kb_ref_fn_cap = mod.__pyx_capi__["SyclKernelBundle_GetKernelBundleRef"]
     # construct Python callable to invoke "SyclKernelBundle_GetKernelBundleRef"
@@ -54,41 +54,41 @@ def _check_cpython_api_SyclKernelBundle_GetKernelBundleRef(sycl_prog):
     callable_maker = ctypes.PYFUNCTYPE(ctypes.c_void_p, ctypes.py_object)
     get_kernel_bundle_ref_fn = callable_maker(kb_ref_fn_ptr)
 
-    r2 = sycl_prog.addressof_ref()
-    r1 = get_kernel_bundle_ref_fn(sycl_prog)
+    r2 = kb.addressof_ref()
+    r1 = get_kernel_bundle_ref_fn(kb)
     assert r1 == r2
 
 
-def _check_cpython_api_SyclKernelBundle_Make(sycl_prog):
+def _check_cpython_api_SyclKernelBundle_Make(kb):
     """Checks Cython-generated C-API function
-    `SyclKernelBundle_Make` defined in _program.pyx"""
+    `SyclKernelBundle_Make` defined in _compiler.pyx"""
     import ctypes
     import sys
 
-    assert type(sycl_prog) is dpc.SyclKernelBundle
-    mod = sys.modules[sycl_prog.__class__.__module__]
+    assert type(kb) is dpc.SyclKernelBundle
+    mod = sys.modules[kb.__class__.__module__]
     # get capsule storing SyclKernelBundle_Make function ptr
-    make_prog_fn_cap = mod.__pyx_capi__["SyclKernelBundle_Make"]
+    make_kb_fn_cap = mod.__pyx_capi__["SyclKernelBundle_Make"]
     # construct Python callable to invoke "SyclKernelBundle_Make"
     cap_ptr_fn = ctypes.pythonapi.PyCapsule_GetPointer
     cap_ptr_fn.restype = ctypes.c_void_p
     cap_ptr_fn.argtypes = [ctypes.py_object, ctypes.c_char_p]
-    make_prog_fn_ptr = cap_ptr_fn(
-        make_prog_fn_cap,
+    make_kb_fn_ptr = cap_ptr_fn(
+        make_kb_fn_cap,
         b"struct PySyclKernelBundleObject *(DPCTLSyclKernelBundleRef)",
     )
     # PYFUNCTYPE(result_type, *arg_types)
     callable_maker = ctypes.PYFUNCTYPE(ctypes.py_object, ctypes.c_void_p)
-    make_prog_fn = callable_maker(make_prog_fn_ptr)
+    make_kb_fn = callable_maker(make_kb_fn_ptr)
 
-    p2 = make_prog_fn(sycl_prog.addressof_ref())
-    assert p2.has_sycl_kernel("add")
-    assert p2.has_sycl_kernel("axpy")
+    kb2 = make_kb_fn(kb.addressof_ref())
+    assert kb2.has_sycl_kernel("add")
+    assert kb2.has_sycl_kernel("axpy")
 
 
 def _check_cpython_api_SyclKernel_GetKernelRef(krn):
     """Checks Cython-generated C-API function
-    `SyclKernel_GetKernelRef` defined in _program.pyx"""
+    `SyclKernel_GetKernelRef` defined in _compiler.pyx"""
     import ctypes
     import sys
 
@@ -114,7 +114,7 @@ def _check_cpython_api_SyclKernel_GetKernelRef(krn):
 
 def _check_cpython_api_SyclKernel_Make(krn):
     """Checks Cython-generated C-API function
-    `SyclKernel_Make` defined in _program.pyx"""
+    `SyclKernel_Make` defined in _compiler.pyx"""
     import ctypes
     import sys
 
@@ -149,7 +149,7 @@ def _check_cpython_api_SyclKernel_Make(krn):
     assert krn.work_group_size == k3.work_group_size
 
 
-def _check_multi_kernel_program(kb):
+def _check_multi_kernel_bundle(kb):
     assert type(kb) is dpc.SyclKernelBundle
 
     assert type(kb.addressof_ref()) is int
@@ -206,7 +206,7 @@ def test_create_kernel_bundle_from_source_ocl():
     except dpctl.SyclQueueCreationError:
         pytest.skip("No OpenCL queue is available")
     kb = dpc.create_kernel_bundle_from_source(q, oclSrc)
-    _check_multi_kernel_program(kb)
+    _check_multi_kernel_bundle(kb)
 
 
 def test_create_kernel_bundle_from_spirv_ocl():
@@ -218,7 +218,7 @@ def test_create_kernel_bundle_from_spirv_ocl():
     with open(spirv_file, "rb") as fin:
         spirv = fin.read()
     kb = dpc.create_kernel_bundle_from_spirv(q, spirv)
-    _check_multi_kernel_program(kb)
+    _check_multi_kernel_bundle(kb)
 
 
 def test_create_kernel_bundle_from_spirv_l0():
@@ -230,7 +230,7 @@ def test_create_kernel_bundle_from_spirv_l0():
     with open(spirv_file, "rb") as fin:
         spirv = fin.read()
     kb = dpc.create_kernel_bundle_from_spirv(q, spirv)
-    _check_multi_kernel_program(kb)
+    _check_multi_kernel_bundle(kb)
 
 
 @pytest.mark.xfail(
@@ -251,7 +251,7 @@ def test_create_kernel_bundle_from_source_l0():
         c[index] = a[index] + d*b[index];                                  \
     }"
     kb = dpc.create_kernel_bundle_from_source(q, oclSrc)
-    _check_multi_kernel_program(kb)
+    _check_multi_kernel_bundle(kb)
 
 
 def test_create_kernel_bundle_from_invalid_src_ocl():

examples/pybind11/use_dpctl_sycl_kernel/example.py

@@ -30,12 +30,12 @@
 with open("resource/double_it.spv", "br") as fh:
     il = fh.read()
 
-# Build the program for the selected device
-pr = dpc.create_kernel_bundle_from_spirv(q, il, "")
-assert pr.has_sycl_kernel("double_it")
+# Build the kernel bundle for the selected device
+kb = dpc.create_kernel_bundle_from_spirv(q, il, "")
+assert kb.has_sycl_kernel("double_it")
 
-# Retrieve the kernel from the problem
-krn = pr.get_sycl_kernel("double_it")
+# Retrieve the kernel from the kernel bundle
+krn = kb.get_sycl_kernel("double_it")
 assert krn.num_args == 2
 
 # Construct the argument, and allocate memory for the result

libsyclinterface/CMakeLists.txt

@@ -15,12 +15,26 @@ if(NOT DEFINED IntelSYCL_FOUND OR NOT IntelSYCL_FOUND)
    find_package(IntelSYCL REQUIRED)
 endif()
 
-# Option to turn on support for creating Level Zero interoperability programs
-# from a SPIR-V binary file.
-option(DPCTL_ENABLE_L0_PROGRAM_CREATION
-    "Enable Level Zero Program creation from SPIR-V"
+# Option to turn on support for creating Level Zero interoperability kernel
+# bundles from a SPIR-V binary file.
+option(DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
+    "Enable Level Zero kernel bundle creation from SPIR-V"
     ON
 )
+if(DEFINED DPCTL_ENABLE_L0_PROGRAM_CREATION)
+    message(DEPRECATION
+        "DPCTL_ENABLE_L0_PROGRAM_CREATION is deprecated and will be removed in "
+        "a future release. Use DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION instead."
+    )
+    set(DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
+        ${DPCTL_ENABLE_L0_PROGRAM_CREATION}
+        CACHE BOOL
+        "Enable Level Zero kernel bundle creation from SPIR-V"
+        FORCE
+    )
+    unset(DPCTL_ENABLE_L0_PROGRAM_CREATION)
+    unset(DPCTL_ENABLE_L0_PROGRAM_CREATION CACHE)
+endif()
 # Option to generate code coverage report using llvm-cov and lcov.
 option(DPCTL_GENERATE_COVERAGE
     "Build dpctl C API with coverage instrumentation"
@@ -59,8 +73,8 @@ else()
     find_package(IntelSyclCompiler REQUIRED)
 endif()
 
-if(DPCTL_ENABLE_L0_PROGRAM_CREATION)
-    set(DPCTL_ENABLE_L0_PROGRAM_CREATION 1)
+if(DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION)
+    set(DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION 1)
     if(DPCTL_LEVEL_ZERO_INCLUDE_DIR)
         set(LEVEL_ZERO_INCLUDE_DIR ${DPCTL_LEVEL_ZERO_INCLUDE_DIR})
     else()
@@ -371,7 +385,7 @@ if (build_so_version)
     )
 endif()
 
-if(DPCTL_ENABLE_L0_PROGRAM_CREATION)
+if(DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION)
     target_include_directories(DPCTLSyclInterface
         SYSTEM PRIVATE
         ${LEVEL_ZERO_INCLUDE_DIR}

libsyclinterface/dbg_build.sh

@@ -20,7 +20,7 @@ cmake                                                       \
     -DCMAKE_CXX_FLAGS=-fsycl                                \
     -DCMAKE_INSTALL_PREFIX="${INSTALL_PREFIX}"              \
     -DCMAKE_PREFIX_PATH="${INSTALL_PREFIX}"                 \
-    -DDPCTL_ENABLE_L0_PROGRAM_CREATION=ON                   \
+    -DDPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION=ON             \
     -DDPCTL_BUILD_CAPI_TESTS=ON                             \
     -DDPCTL_GENERATE_COVERAGE=OFF                           \
     ..

libsyclinterface/dbg_build_custom.sh

@@ -17,7 +17,7 @@ cmake                                                       \
     -DCMAKE_PREFIX_PATH="${INSTALL_PREFIX}"                 \
     -DDPCTL_CUSTOM_DPCPP_INSTALL_DIR="${DPCPP_HOME}"        \
     -DCMAKE_LINKER:PATH="${DPCPP_HOME}/bin/lld"             \
-    -DDPCTL_ENABLE_L0_PROGRAM_CREATION="${USE_LO_HEADERS}"  \
+    -DDPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION="${USE_LO_HEADERS}" \
     -DDPCTL_BUILD_CAPI_TESTS=ON                             \
     -DDPCTL_GENERATE_COVERAGE=ON                            \
     ..

libsyclinterface/include/syclinterface/Config/dpctl_config.h.in

@@ -25,8 +25,9 @@
 
 #pragma once
 
-/* Defined when dpctl was built with level zero program creation enabled. */
-#cmakedefine DPCTL_ENABLE_L0_PROGRAM_CREATION 1
+/* Defined when dpctl was built with Level Zero kernel bundle creation
+ * enabled. */
+#cmakedefine DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION 1
 
 #define __SYCL_COMPILER_VERSION_REQUIRED 20221201L
 

libsyclinterface/source/dpctl_sycl_kernel_bundle_interface.cpp

@@ -38,7 +38,7 @@
 #include <sycl/sycl.hpp> /* Sycl headers       */
 #include <utility>
 
-#ifdef DPCTL_ENABLE_L0_PROGRAM_CREATION
+#ifdef DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
 // Note: include ze_api.h before level_zero.hpp. Make sure clang-format does
 // not reorder the includes.
 // clang-format off
@@ -63,7 +63,7 @@ static const int clLibLoadFlags = RTLD_NOLOAD | RTLD_NOW | RTLD_LOCAL;
 static const char *clLoaderName = "OpenCL.dll";
 static const int clLibLoadFlags = 0;
 #else
-#error "OpenCL program compilation is unavailable for this platform"
+#error "OpenCL kernel bundle compilation is unavailable for this platform"
 #endif
 
 #define CodeStringSuffix(code)                                                 \
@@ -408,7 +408,7 @@ _GetKernel_ocl_impl(const kernel_bundle<bundle_state::executable> &kb,
     }
 }
 
-#ifdef DPCTL_ENABLE_L0_PROGRAM_CREATION
+#ifdef DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
 
 #ifdef __linux__
 static const char *zeLoaderName = DPCTL_LIBZE_LOADER_FILENAME;
@@ -417,7 +417,7 @@ static const int zeLibLoadFlags = RTLD_NOLOAD | RTLD_NOW | RTLD_LOCAL;
 static const char *zeLoaderName = "ze_loader.dll";
 static const int zeLibLoadFlags = 0;
 #else
-#error "Level Zero program compilation is unavailable for this platform"
+#error "Level Zero kernel bundle compilation is unavailable for this platform"
 #endif
 
 static constexpr sycl::backend ze_be = sycl::backend::ext_oneapi_level_zero;
@@ -676,7 +676,7 @@ bool _HasKernel_ze_impl(const kernel_bundle<bundle_state::executable> &kb,
     return false;
 }
 
-#endif /* #ifdef DPCTL_ENABLE_L0_PROGRAM_CREATION */
+#endif /* #ifdef DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION */
 
 } /* end of anonymous namespace */
 
@@ -691,19 +691,19 @@ DPCTLKernelBundle_CreateFromSpirv(__dpctl_keep const DPCTLSyclContextRef CtxRef,
 {
     DPCTLSyclKernelBundleRef KBRef = nullptr;
     if (!CtxRef) {
-        error_handler("Cannot create program from SPIR-V as the supplied SYCL "
-                      "context is NULL.",
+        error_handler("Cannot create kernel bundle from SPIR-V as the supplied "
+                      "SYCL context is NULL.",
                       __FILE__, __func__, __LINE__);
         return KBRef;
     }
     if (!DevRef) {
-        error_handler("Cannot create program from SPIR-V as the supplied SYCL "
-                      "device is NULL.",
+        error_handler("Cannot create kernel bundle from SPIR-V as the supplied "
+                      "SYCL device is NULL.",
                       __FILE__, __func__, __LINE__);
         return KBRef;
     }
     if ((!IL) || (length == 0)) {
-        error_handler("Cannot create program from null SPIR-V buffer.",
+        error_handler("Cannot create kernel bundle from null SPIR-V buffer.",
                       __FILE__, __func__, __LINE__);
         return KBRef;
     }
@@ -720,7 +720,7 @@ DPCTLKernelBundle_CreateFromSpirv(__dpctl_keep const DPCTLSyclContextRef CtxRef,
                 SpecConsts);
             break;
         case backend::ext_oneapi_level_zero:
-#ifdef DPCTL_ENABLE_L0_PROGRAM_CREATION
+#ifdef DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
             KBRef = _CreateKernelBundleWithIL_ze_impl(
                 *SyclCtx, *SyclDev, IL, length, CompileOpts, NumSpecConsts,
                 SpecConsts);
@@ -807,7 +807,7 @@ DPCTLKernelBundle_GetKernel(__dpctl_keep DPCTLSyclKernelBundleRef KBRef,
     case sycl::backend::opencl:
         return _GetKernel_ocl_impl(*SyclKB, KernelName);
     case sycl::backend::ext_oneapi_level_zero:
-#ifdef DPCTL_ENABLE_L0_PROGRAM_CREATION
+#ifdef DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
         return _GetKernel_ze_impl(*SyclKB, KernelName);
 #endif
     default:
@@ -837,7 +837,7 @@ bool DPCTLKernelBundle_HasKernel(__dpctl_keep DPCTLSyclKernelBundleRef KBRef,
     case sycl::backend::opencl:
         return _HasKernel_ocl_impl(*SyclKB, KernelName);
     case sycl::backend::ext_oneapi_level_zero:
-#ifdef DPCTL_ENABLE_L0_PROGRAM_CREATION
+#ifdef DPCTL_ENABLE_L0_KERNEL_BUNDLE_CREATION
         return _HasKernel_ze_impl(*SyclKB, KernelName);
 #endif
     default: