[Testing] Improve "program" testing to better match the DPC++ e2e tests

This adds a number of conformance tests to test things that are required for the "program" DPC++ e2e tests. Note that these tests don't all pass. Adapter testing infrastructure for level zero has been added, and a single test has been added to check their specific handling of linker errors (they write it to the build log of the output program). Other than that, test additions are as follows: * A few specialization constant tests to test usage of a kernel with multiple specialization constants defined. * Testing of the default specialization values. * KernelGetInfo outputs the correct kernel name and context pointer. * Compiling an (valid or invalid) program produces a valid (though unspecified) build log. * The "num devices" program info is sensible.
oneapi-src · Apr 29, 2024 · 1b80a22 · 1b80a22
1 parent a6cdcf4
commit 1b80a22
Show file tree

Hide file tree

Showing 25 changed files with 416 additions and 2 deletions.
diff --git a/include/ur_api.h b/include/ur_api.h
@@ -4076,6 +4076,8 @@ typedef struct ur_program_properties_t {
 ///
 /// @details
 ///     - The application may call this function from simultaneous threads.
+///     - The adapter may (but is not required to) perform validation of the
+///       provided module during this call.
 ///
 /// @remarks
 ///   _Analogues_
@@ -4118,6 +4120,8 @@ urProgramCreateWithIL(
 ///       ::UR_PROGRAM_BINARY_TYPE_LIBRARY for `hDevice`.
 ///     - The device specified by `hDevice` must be device associated with
 ///       context.
+///     - The adapter may (but is not required to) perform validation of the
+///       provided module during this call.
 ///
 /// @remarks
 ///   _Analogues_

diff --git a/scripts/core/program.yml b/scripts/core/program.yml
@@ -89,6 +89,7 @@ analogue:
     - "**clCreateProgramWithIL**"
 details:
     - "The application may call this function from simultaneous threads."
+    - "The adapter may (but is not required to) perform validation of the provided module during this call."
 params:
     - type: $x_context_handle_t
       name: hContext
@@ -129,6 +130,7 @@ details:
     - "The application may call this function from simultaneous threads."
     - "Following a successful call to this entry point, `phProgram` will contain a binary of type $X_PROGRAM_BINARY_TYPE_COMPILED_OBJECT or $X_PROGRAM_BINARY_TYPE_LIBRARY for `hDevice`."
     - "The device specified by `hDevice` must be device associated with context."
+    - "The adapter may (but is not required to) perform validation of the provided module during this call."
 params:
     - type: $x_context_handle_t
       name: hContext

diff --git a/source/loader/ur_libapi.cpp b/source/loader/ur_libapi.cpp
@@ -2890,6 +2890,8 @@ ur_result_t UR_APICALL urPhysicalMemRelease(
 ///
 /// @details
 ///     - The application may call this function from simultaneous threads.
+///     - The adapter may (but is not required to) perform validation of the
+///       provided module during this call.
 ///
 /// @remarks
 ///   _Analogues_
@@ -2942,6 +2944,8 @@ ur_result_t UR_APICALL urProgramCreateWithIL(
 ///       ::UR_PROGRAM_BINARY_TYPE_LIBRARY for `hDevice`.
 ///     - The device specified by `hDevice` must be device associated with
 ///       context.
+///     - The adapter may (but is not required to) perform validation of the
+///       provided module during this call.
 ///
 /// @remarks
 ///   _Analogues_

diff --git a/source/ur_api.cpp b/source/ur_api.cpp
@@ -2459,6 +2459,8 @@ ur_result_t UR_APICALL urPhysicalMemRelease(
 ///
 /// @details
 ///     - The application may call this function from simultaneous threads.
+///     - The adapter may (but is not required to) perform validation of the
+///       provided module during this call.
 ///
 /// @remarks
 ///   _Analogues_
@@ -2505,6 +2507,8 @@ ur_result_t UR_APICALL urProgramCreateWithIL(
 ///       ::UR_PROGRAM_BINARY_TYPE_LIBRARY for `hDevice`.
 ///     - The device specified by `hDevice` must be device associated with
 ///       context.
+///     - The adapter may (but is not required to) perform validation of the
+///       provided module during this call.
 ///
 /// @remarks
 ///   _Analogues_

diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt
@@ -17,9 +17,12 @@ set(gtest_force_shared_crt ON CACHE BOOL "" FORCE)
 FetchContent_MakeAvailable(googletest)
 enable_testing()
 
+# Conformance defines the generate_device_binaries target which should be
+# imported first
+add_subdirectory(conformance)
+
 add_subdirectory(loader)
 add_subdirectory(adapters)
-add_subdirectory(conformance)
 add_subdirectory(usm)
 add_subdirectory(layers)
 add_subdirectory(unit)

diff --git a/test/adapters/CMakeLists.txt b/test/adapters/CMakeLists.txt
@@ -24,6 +24,15 @@ function(add_adapter_test name)
 
     target_compile_definitions(${target} PRIVATE
         ${args_FIXTURE}_ENVIRONMENT)
+
+    if(${args_FIXTURE} STREQUAL "KERNELS")
+        target_compile_definitions(${target} PRIVATE KERNELS_ENVIRONMENT PRIVATE
+            KERNELS_DEFAULT_DIR="${UR_CONFORMANCE_DEVICE_BINARIES_DIR}")
+        target_include_directories(${target}
+            PRIVATE ${UR_CONFORMANCE_DEVICE_BINARIES_DIR})
+        add_dependencies(${target} generate_device_binaries kernel_names_header)
+    endif()
+
     target_link_libraries(${target} PRIVATE
         ${PROJECT_NAME}::loader
         ${PROJECT_NAME}::headers
@@ -46,3 +55,7 @@ endif()
 if(UR_BUILD_ADAPTER_HIP OR UR_BUILD_ADAPTER_ALL)
     add_subdirectory(hip)
 endif()
+
+if(UR_BUILD_ADAPTER_L0 OR UR_BUILD_ADAPTER_ALL)
+    add_subdirectory(level_zero)
+endif()
diff --git a/test/adapters/level_zero/CMakeLists.txt b/test/adapters/level_zero/CMakeLists.txt
@@ -0,0 +1,27 @@
+# Copyright (C) 2024 Intel Corporation
+# Part of the Unified-Runtime Project, under the Apache License v2.0 with LLVM Exceptions.
+# See LICENSE.TXT
+# SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+if(NOT UR_DPCXX)
+    # Tests that require kernels can't be used if we aren't generating
+    # device binaries
+    message(WARNING
+        "UR_DPCXX is not defined, skipping adapter tests for level_zero")
+else()
+    add_adapter_test(level_zero
+        FIXTURE KERNELS
+        SOURCES
+            urProgramLink.cpp
+        ENVIRONMENT
+            "UR_ADAPTERS_FORCE_LOAD=\"$<TARGET_FILE:ur_adapter_level_zero>\""
+    )
+
+    target_include_directories(test-adapter-level_zero PRIVATE
+        ${PROJECT_SOURCE_DIR}/source
+        ${PROJECT_SOURCE_DIR}/source/adapters/level_zero
+    )
+
+    add_dependencies(test-adapter-level_zero
+        generate_device_binaries kernel_names_header)
+endif()
diff --git a/test/adapters/level_zero/urProgramLink.cpp b/test/adapters/level_zero/urProgramLink.cpp
@@ -0,0 +1,31 @@
+// Copyright (C) 2024 Intel Corporation
+// Part of the Unified-Runtime Project, under the Apache License v2.0 with LLVM Exceptions.
+// See LICENSE.TXT
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+#include <uur/fixtures.h>
+
+using urLevelZeroProgramLinkTest = uur::urProgramTest;
+UUR_INSTANTIATE_DEVICE_TEST_SUITE_P(urLevelZeroProgramLinkTest);
+
+TEST_P(urLevelZeroProgramLinkTest, InvalidLinkOptionsPrintedInLog) {
+    ur_program_handle_t linked_program = nullptr;
+    ASSERT_SUCCESS(urProgramCompile(context, program, "-foo"));
+    ASSERT_EQ_RESULT(
+        UR_RESULT_ERROR_PROGRAM_LINK_FAILURE,
+        urProgramLink(context, 1, &program, "-foo", &linked_program));
+
+    size_t logSize;
+    std::vector<char> log;
+
+    ASSERT_SUCCESS(urProgramGetBuildInfo(linked_program, device,
+                                         UR_PROGRAM_BUILD_INFO_LOG, 0, nullptr,
+                                         &logSize));
+    log.resize(logSize);
+    log[logSize - 1] = 'x';
+    ASSERT_SUCCESS(urProgramGetBuildInfo(linked_program, device,
+                                         UR_PROGRAM_BUILD_INFO_LOG, logSize,
+                                         log.data(), nullptr));
+    ASSERT_EQ(log[logSize - 1], '\0');
+    ASSERT_NE(std::string{log.data()}.find("-foo"), std::string::npos);
+}
diff --git a/test/conformance/device_code/CMakeLists.txt b/test/conformance/device_code/CMakeLists.txt
@@ -67,6 +67,15 @@ macro(add_device_binary SOURCE_FILE)
         if(${TRIPLE} MATCHES "amd" AND ${KERNEL_NAME} MATCHES "image_copy")
             continue()
         endif()
+
+        # This seems to fail to build the SYCL binary due to the invalid asm
+        if(${TRIPLE} MATCHES "cuda" AND ${KERNEL_NAME} MATCHES "build_failure")
+            continue()
+        endif()
+        if(${TRIPLE} MATCHES "amd" AND ${KERNEL_NAME} MATCHES "build_failure")
+            continue()
+        endif()
+
         add_custom_command(OUTPUT ${EXE_PATH}
             COMMAND ${UR_DPCXX} -fsycl -fsycl-targets=${TRIPLE} -fsycl-device-code-split=off 
             ${AMD_TARGET_BACKEND} ${AMD_OFFLOAD_ARCH} ${AMD_NOGPULIB}
@@ -93,10 +102,12 @@ add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/foo.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/image_copy.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/mean.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/spec_constant.cpp)
+add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/spec_constant_multiple.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/usm_ll.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/saxpy.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/saxpy_usm.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/indexers_usm.cpp)
+add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/build_failure.cpp)
 
 set(KERNEL_HEADER ${UR_CONFORMANCE_DEVICE_BINARIES_DIR}/kernel_entry_points.h)
 add_custom_command(OUTPUT ${KERNEL_HEADER}

diff --git a/test/conformance/device_code/build_failure.cpp b/test/conformance/device_code/build_failure.cpp
@@ -0,0 +1,21 @@
+// Copyright (C) 2023 Intel Corporation
+// Part of the Unified-Runtime Project, under the Apache License v2.0 with LLVM Exceptions.
+// See LICENSE.TXT
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+#include <sycl/sycl.hpp>
+
+int main() {
+    sycl::queue deviceQueue;
+
+    auto Kernel = []() {
+#ifdef __SYCL_DEVICE_ONLY__
+        asm volatile("undefined\n");
+#endif // __SYCL_DEVICE_ONLY__
+    };
+
+    deviceQueue.submit(
+        [&](sycl::handler &cgh) { cgh.single_task<class Foo>(Kernel); });
+
+    return 0;
+}
diff --git a/test/conformance/device_code/spec_constant.cpp b/test/conformance/device_code/spec_constant.cpp
@@ -8,7 +8,7 @@
 
 using namespace sycl;
 
-constexpr specialization_id<uint32_t> spec_const;
+constexpr specialization_id<uint32_t> spec_const{1000};
 
 int main() {
     queue myQueue;

diff --git a/test/conformance/device_code/spec_constant_multiple.cpp b/test/conformance/device_code/spec_constant_multiple.cpp
@@ -0,0 +1,43 @@
+// Copyright (C) 2024 Intel Corporation
+// Part of the Unified-Runtime Project, under the Apache License v2.0 with LLVM Exceptions.
+// See LICENSE.TXT
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+#include <cstdint>
+#include <sycl/sycl.hpp>
+
+using namespace sycl;
+
+constexpr specialization_id<uint32_t> a_spec;
+constexpr specialization_id<uint64_t> b_spec;
+constexpr specialization_id<bool> c_spec;
+
+int main() {
+    queue myQueue;
+    uint64_t out_val = 0;
+    buffer<uint64_t, 1> out(&out_val, sycl::range<1>{1});
+
+    myQueue.submit([&](handler &cgh) {
+        accessor out_acc{out, cgh, write_only};
+
+        cgh.set_specialization_constant<a_spec>(0);
+        cgh.set_specialization_constant<b_spec>(0);
+        cgh.set_specialization_constant<c_spec>(false);
+
+        cgh.parallel_for<class SpecConstant>(
+            out.get_range(), [=](item<1> item_id, kernel_handler h) {
+                uint32_t a = h.get_specialization_constant<a_spec>();
+                uint64_t b = h.get_specialization_constant<b_spec>();
+                bool c = h.get_specialization_constant<c_spec>();
+                if (c) {
+                    out_acc[0] = b - a;
+                } else {
+                    out_acc[0] = b + a;
+                }
+            });
+    });
+
+    myQueue.wait();
+
+    return 0;
+}
diff --git a/test/conformance/kernel/kernel_adapter_native_cpu.match b/test/conformance/kernel/kernel_adapter_native_cpu.match
@@ -80,6 +80,8 @@ urKernelGetInfoTest.InvalidNullPointerPropSizeRet/SYCL_NATIVE_CPU___SYCL_Native_
 urKernelGetInfoTest.InvalidNullPointerPropSizeRet/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_KERNEL_INFO_PROGRAM
 urKernelGetInfoTest.InvalidNullPointerPropSizeRet/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_KERNEL_INFO_ATTRIBUTES
 urKernelGetInfoTest.InvalidNullPointerPropSizeRet/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_KERNEL_INFO_NUM_REGS
+urKernelGetInfoSingleTest.KernelNameCorrect/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+urKernelGetInfoSingleTest.KernelContextCorrect/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 urKernelGetNativeHandleTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 urKernelGetNativeHandleTest.InvalidNullHandleKernel/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 urKernelGetNativeHandleTest.InvalidNullPointerNativeKernel/SYCL_NATIVE_CPU___SYCL_Native_CPU_

diff --git a/test/conformance/kernel/urKernelGetInfo.cpp b/test/conformance/kernel/urKernelGetInfo.cpp
@@ -15,6 +15,9 @@ UUR_TEST_SUITE_P(
                       UR_KERNEL_INFO_NUM_REGS),
     uur::deviceTestWithParamPrinter<ur_kernel_info_t>);
 
+using urKernelGetInfoSingleTest = uur::urKernelExecutionTest;
+UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urKernelGetInfoSingleTest);
+
 TEST_P(urKernelGetInfoTest, Success) {
     auto property_name = getParam();
     size_t property_size = 0;
@@ -88,3 +91,23 @@ TEST_P(urKernelGetInfoTest, InvalidNullPointerPropSizeRet) {
         urKernelGetInfo(kernel, UR_KERNEL_INFO_NUM_ARGS, 0, nullptr, nullptr),
         UR_RESULT_ERROR_INVALID_NULL_POINTER);
 }
+
+TEST_P(urKernelGetInfoSingleTest, KernelNameCorrect) {
+    size_t name_size = 0;
+    std::vector<char> name_data;
+    ASSERT_SUCCESS(urKernelGetInfo(kernel, UR_KERNEL_INFO_FUNCTION_NAME, 0,
+                                   nullptr, &name_size));
+    name_data.resize(name_size);
+    ASSERT_SUCCESS(urKernelGetInfo(kernel, UR_KERNEL_INFO_FUNCTION_NAME,
+                                   name_size, name_data.data(), nullptr));
+    ASSERT_EQ(name_data[name_size - 1], '\0');
+    ASSERT_STREQ(kernel_name.c_str(), name_data.data());
+}
+
+TEST_P(urKernelGetInfoSingleTest, KernelContextCorrect) {
+    ur_context_handle_t info_context;
+    ASSERT_SUCCESS(urKernelGetInfo(kernel, UR_KERNEL_INFO_CONTEXT,
+                                   sizeof(ur_context_handle_t), &info_context,
+                                   nullptr));
+    ASSERT_EQ(context, info_context);
+}
diff --git a/test/conformance/program/program_adapter_cuda.match b/test/conformance/program/program_adapter_cuda.match
@@ -1,3 +1,4 @@
+urProgramBuildTest.BuildFailure/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullHandleContext/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullPointerProgram/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramGetBuildInfoTest.Success/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_BINARY_TYPE
@@ -9,6 +10,7 @@
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleDevice/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_OPTIONS
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleDevice/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_LOG
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleDevice/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_BINARY_TYPE
+{{OPT}}urProgramGetBuildInfoSingleTest.LogIsNullTerminated/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramGetInfoTest.Success/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_INFO_NUM_KERNELS
 {{OPT}}urProgramGetInfoTest.Success/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_INFO_KERNEL_NAMES
 {{OPT}}urProgramGetInfoTest.InvalidNullHandleProgram/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_INFO_REFERENCE_COUNT
@@ -22,3 +24,6 @@
 {{OPT}}urProgramGetInfoTest.InvalidNullHandleProgram/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_INFO_KERNEL_NAMES
 {{OPT}}urProgramLinkTest.Success/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramSetSpecializationConstantsTest.Success/NVIDIA_CUDA_BACKEND___{{.*}}_
+{{OPT}}urProgramSetSpecializationConstantsTest.UseDefaultValue/NVIDIA_CUDA_BACKEND___{{.*}}_
+urProgramSetMultipleSpecializationConstantsTest.MultipleCalls/NVIDIA_CUDA_BACKEND___{{.*}}_
+urProgramSetMultipleSpecializationConstantsTest.SingleCall/NVIDIA_CUDA_BACKEND___{{.*}}_
diff --git a/test/conformance/program/program_adapter_hip.match b/test/conformance/program/program_adapter_hip.match
@@ -1,3 +1,4 @@
+urProgramBuildTest.BuildFailure/AMD_HIP_BACKEND___{{.*}}_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullHandleContext/AMD_HIP_BACKEND___{{.*}}_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullPointerProgram/AMD_HIP_BACKEND___{{.*}}_
 {{OPT}}urProgramGetBuildInfoTest.Success/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_BINARY_TYPE
@@ -9,6 +10,7 @@
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleDevice/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_OPTIONS
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleDevice/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_LOG
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleDevice/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_BINARY_TYPE
+{{OPT}}urProgramGetBuildInfoSingleTest.LogIsNullTerminated/AMD_HIP_BACKEND___{{.*}}_
 {{OPT}}urProgramGetInfoTest.Success/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_INFO_NUM_KERNELS
 {{OPT}}urProgramGetInfoTest.Success/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_INFO_KERNEL_NAMES
 {{OPT}}urProgramGetInfoTest.InvalidNullHandleProgram/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_INFO_REFERENCE_COUNT

diff --git a/test/conformance/program/program_adapter_native_cpu.match b/test/conformance/program/program_adapter_native_cpu.match
@@ -2,6 +2,7 @@
 {{OPT}}urProgramBuildTest.SuccessWithOptions/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramBuildTest.InvalidNullHandleContext/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramBuildTest.InvalidNullHandleProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramBuildTest.BuildFailure/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCompileTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCompileTest.InvalidNullHandleContext/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCompileTest.InvalidNullHandleProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
@@ -12,12 +13,14 @@
 {{OPT}}urProgramCreateWithBinaryTest.InvalidNullPointerProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithBinaryTest.InvalidNullPointerMetadata/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithBinaryTest.InvalidSizePropertyCount/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramCreateWithBinaryTest.BuildInvalidProgramBinary/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithILTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithILTest.SuccessWithProperties/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithILTest.InvalidNullHandle/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithILTest.InvalidNullPointerSource/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithILTest.InvalidSizeLength/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithILTest.InvalidNullPointerProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramCreateWithILTest.BuildInvalidProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithNativeHandleTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullHandleContext/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullPointerProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
@@ -37,6 +40,7 @@
 {{OPT}}urProgramGetBuildInfoTest.InvalidEnumeration/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_PROGRAM_BUILD_INFO_OPTIONS
 {{OPT}}urProgramGetBuildInfoTest.InvalidEnumeration/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_PROGRAM_BUILD_INFO_LOG
 {{OPT}}urProgramGetBuildInfoTest.InvalidEnumeration/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_PROGRAM_BUILD_INFO_BINARY_TYPE
+{{OPT}}urProgramGetBuildInfoSingleTest.LogIsNullTerminated/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramGetFunctionPointerTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramGetFunctionPointerTest.InvalidFunctionName/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramGetFunctionPointerTest.InvalidNullHandleDevice/SYCL_NATIVE_CPU___SYCL_Native_CPU_
@@ -112,6 +116,9 @@
 {{OPT}}urProgramGetInfoTest.InvalidNullPointerPropValueRet/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_PROGRAM_INFO_BINARIES
 {{OPT}}urProgramGetInfoTest.InvalidNullPointerPropValueRet/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_PROGRAM_INFO_NUM_KERNELS
 {{OPT}}urProgramGetInfoTest.InvalidNullPointerPropValueRet/SYCL_NATIVE_CPU___SYCL_Native_CPU___UR_PROGRAM_INFO_KERNEL_NAMES
+{{OPT}}urProgramGetInfoSingleTest.NumDevicesIsNonzero/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramGetInfoSingleTest.NumDevicesMatchesDeviceArray/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramGetInfoSingleTest.NumDevicesMatchesContextNumDevices/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramGetNativeHandleTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramGetNativeHandleTest.InvalidNullHandleProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramGetNativeHandleTest.InvalidNullPointerNativeProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
@@ -125,6 +132,9 @@
 {{OPT}}urProgramRetainTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramRetainTest.InvalidNullHandleProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramSetSpecializationConstantsTest.Success/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramSetSpecializationConstantsTest.UseDefaultValue/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramSetSpecializationConstantsTest.InvalidNullHandleProgram/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramSetSpecializationConstantsTest.InvalidNullPointerSpecConstants/SYCL_NATIVE_CPU___SYCL_Native_CPU_
 {{OPT}}urProgramSetSpecializationConstantsTest.InvalidSizeCount/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramSetMultipleSpecializationConstantsTest.MultipleCalls/SYCL_NATIVE_CPU___SYCL_Native_CPU_
+{{OPT}}urProgramSetMultipleSpecializationConstantsTest.SingleCall/SYCL_NATIVE_CPU___SYCL_Native_CPU_