[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.

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.

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_DEFAULT_DIR="${UR_CONFORMANCE_DEVICE_BINARIES_DIR}")
+        target_include_directories(${target}
+            PRIVATE ${UR_CONFORMANCE_DEVICE_BINARIES_DIR})
+        add_dependencies(${target} generate_device_binaries)
+    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()

test/adapters/level_zero/CMakeLists.txt

@@ -0,0 +1,20 @@
+# 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
+
+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 kernel_names_header)

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);
+}

test/conformance/device_code/CMakeLists.txt

@@ -93,6 +93,7 @@ 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)

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;

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;
+}

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_

test/conformance/kernel/urKernelGetInfo.cpp

@@ -15,6 +15,14 @@ UUR_TEST_SUITE_P(
                       UR_KERNEL_INFO_NUM_REGS),
     uur::deviceTestWithParamPrinter<ur_kernel_info_t>);
 
+struct urKernelGetInfoSingleTest : uur::urBaseKernelTest {
+    void SetUp() override {
+        UUR_RETURN_ON_FATAL_FAILURE(urBaseKernelTest::SetUp());
+        urBaseKernelTest::Build();
+    }
+};
+UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urKernelGetInfoSingleTest);
+
 TEST_P(urKernelGetInfoTest, Success) {
     auto property_name = getParam();
     size_t property_size = 0;
@@ -66,3 +74,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_EQ(kernel_name, std::string{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);
+}

test/conformance/program/program_adapter_cuda.match

@@ -1,5 +1,7 @@
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullHandleContext/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramCreateWithNativeHandleTest.InvalidNullPointerProgram/NVIDIA_CUDA_BACKEND___{{.*}}_
+urProgramCreateWithBinaryTest.BuildInvalidProgramBinary/NVIDIA_CUDA_BACKEND___{{.*}}_
+urProgramCreateWithILTest.BuildInvalidProgram/NVIDIA_CUDA_BACKEND___{{.*}}_
 {{OPT}}urProgramGetBuildInfoTest.Success/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_BINARY_TYPE
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleProgram/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_STATUS
 {{OPT}}urProgramGetBuildInfoTest.InvalidNullHandleProgram/NVIDIA_CUDA_BACKEND___{{.*}}___UR_PROGRAM_BUILD_INFO_OPTIONS

test/conformance/program/program_adapter_native_cpu.match

@@ -12,12 +12,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 +39,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 +115,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 +131,3 @@
 {{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.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_

test/conformance/program/program_adapter_opencl.match

@@ -1,3 +1,4 @@
+urProgramCreateWithILTest.BuildInvalidProgram/Intel_R__OpenCL___{{.*}}_
 urProgramGetFunctionPointerTest.InvalidFunctionName/Intel_R__OpenCL___{{.*}}_
 urProgramGetInfoTest.Success/Intel_R__OpenCL___{{.*}}___UR_PROGRAM_INFO_SOURCE
 urProgramGetInfoTest.Success/Intel_R__OpenCL___{{.*}}___UR_PROGRAM_INFO_BINARIES

test/conformance/program/urProgramCreateWithBinary.cpp

@@ -94,3 +94,11 @@ TEST_P(urProgramCreateWithBinaryTest, InvalidSizePropertyCount) {
                                                binary.data(), &properties,
                                                &binary_program));
 }
+
+TEST_P(urProgramCreateWithBinaryTest, BuildInvalidProgramBinary) {
+    ur_program_handle_t program = nullptr;
+    uint8_t binary[] = {0, 1, 2, 3, 4};
+    ASSERT_EQ_RESULT(UR_RESULT_ERROR_INVALID_BINARY,
+                     urProgramCreateWithBinary(context, device, 5, binary,
+                                               nullptr, &program));
+}

test/conformance/program/urProgramCreateWithIL.cpp

@@ -73,3 +73,11 @@ TEST_P(urProgramCreateWithILTest, InvalidNullPointerProgram) {
                                            il_binary->size(), nullptr,
                                            nullptr));
 }
+
+TEST_P(urProgramCreateWithILTest, BuildInvalidProgram) {
+    ur_program_handle_t program = nullptr;
+    char binary[] = {0, 1, 2, 3, 4};
+    ASSERT_EQ_RESULT(
+        UR_RESULT_ERROR_INVALID_BINARY,
+        urProgramCreateWithIL(context, &binary, 5, nullptr, &program));
+}

test/conformance/program/urProgramGetBuildInfo.cpp

@@ -21,6 +21,15 @@ UUR_TEST_SUITE_P(urProgramGetBuildInfoTest,
                                    UR_PROGRAM_BUILD_INFO_BINARY_TYPE),
                  uur::deviceTestWithParamPrinter<ur_program_build_info_t>);
 
+struct urProgramGetBuildInfoSingleTest : uur::urProgramTest {
+    void SetUp() override {
+        UUR_RETURN_ON_FATAL_FAILURE(urProgramTest::SetUp());
+        // Some queries need the program to be built.
+        ASSERT_SUCCESS(urProgramBuild(this->context, program, nullptr));
+    }
+};
+UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urProgramGetBuildInfoSingleTest);
+
 TEST_P(urProgramGetBuildInfoTest, Success) {
     auto property_name = getParam();
     size_t property_size = 0;
@@ -60,3 +69,17 @@ TEST_P(urProgramGetBuildInfoTest, InvalidEnumeration) {
                                            UR_PROGRAM_BUILD_INFO_FORCE_UINT32,
                                            0, nullptr, &propSizeOut));
 }
+
+TEST_P(urProgramGetBuildInfoSingleTest, LogIsNullTerminated) {
+    size_t logSize;
+    std::vector<char> log;
+
+    ASSERT_SUCCESS(urProgramGetBuildInfo(
+        program, device, UR_PROGRAM_BUILD_INFO_LOG, 0, nullptr, &logSize));
+    log.resize(logSize);
+    log[logSize - 1] = 'x';
+    ASSERT_SUCCESS(urProgramGetBuildInfo(program, device,
+                                         UR_PROGRAM_BUILD_INFO_LOG, logSize,
+                                         log.data(), nullptr));
+    ASSERT_EQ(log[logSize - 1], '\0');
+}

test/conformance/program/urProgramGetInfo.cpp

@@ -23,6 +23,15 @@ UUR_TEST_SUITE_P(
                       UR_PROGRAM_INFO_KERNEL_NAMES),
     uur::deviceTestWithParamPrinter<ur_program_info_t>);
 
+struct urProgramGetInfoSingleTest : uur::urProgramTest {
+    void SetUp() override {
+        UUR_RETURN_ON_FATAL_FAILURE(urProgramTest::SetUp());
+        // Some queries need the program to be built.
+        ASSERT_SUCCESS(urProgramBuild(this->context, program, nullptr));
+    }
+};
+UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urProgramGetInfoSingleTest);
+
 TEST_P(urProgramGetInfoTest, Success) {
     auto property_name = getParam();
     size_t property_size = 0;
@@ -74,3 +83,35 @@ TEST_P(urProgramGetInfoTest, InvalidNullPointerPropValueRet) {
                                       0, nullptr, nullptr),
                      UR_RESULT_ERROR_INVALID_NULL_POINTER);
 }
+
+TEST_P(urProgramGetInfoSingleTest, NumDevicesIsNonzero) {
+    uint32_t count;
+    ASSERT_SUCCESS(urProgramGetInfo(program, UR_PROGRAM_INFO_NUM_DEVICES,
+                                    sizeof(uint32_t), &count, nullptr));
+    ASSERT_GE(count, 1);
+}
+
+TEST_P(urProgramGetInfoSingleTest, NumDevicesMatchesDeviceArray) {
+    uint32_t count;
+    ASSERT_SUCCESS(urProgramGetInfo(program, UR_PROGRAM_INFO_NUM_DEVICES,
+                                    sizeof(uint32_t), &count, nullptr));
+
+    size_t info_devices_size;
+    ASSERT_SUCCESS(urProgramGetInfo(program, UR_PROGRAM_INFO_DEVICES, 0,
+                                    nullptr, &info_devices_size));
+    ASSERT_EQ(count, info_devices_size / sizeof(ur_device_handle_t *));
+}
+
+TEST_P(urProgramGetInfoSingleTest, NumDevicesMatchesContextNumDevices) {
+    uint32_t count;
+    ASSERT_SUCCESS(urProgramGetInfo(program, UR_PROGRAM_INFO_NUM_DEVICES,
+                                    sizeof(uint32_t), &count, nullptr));
+
+    // The device count either matches the number of devices in the context or
+    // is 1, depending on how it was built
+    uint32_t info_context_devices_count;
+    ASSERT_SUCCESS(urContextGetInfo(context, UR_CONTEXT_INFO_NUM_DEVICES,
+                                    sizeof(uint32_t),
+                                    &info_context_devices_count, nullptr));
+    ASSERT_TRUE(count == 1 || count == info_context_devices_count);
+}