From 2b1e40df91c141c819f7061056d240045ef47b6e Mon Sep 17 00:00:00 2001 From: Ross Brunton Date: Wed, 13 Mar 2024 17:26:10 +0000 Subject: [PATCH] [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. --- test/adapters/CMakeLists.txt | 13 +++ test/adapters/level_zero/CMakeLists.txt | 20 ++++ test/adapters/level_zero/urProgramLink.cpp | 31 ++++++ test/conformance/device_code/CMakeLists.txt | 1 + .../conformance/device_code/spec_constant.cpp | 2 +- .../device_code/spec_constant_multiple.cpp | 43 +++++++++ .../kernel/kernel_adapter_native_cpu.match | 2 + test/conformance/kernel/urKernelGetInfo.cpp | 28 ++++++ .../program/program_adapter_cuda.match | 6 ++ .../program/program_adapter_hip.match | 3 + .../program/program_adapter_native_cpu.match | 9 ++ .../program/program_adapter_opencl.match | 1 + .../program/urProgramCreateWithBinary.cpp | 8 ++ .../program/urProgramCreateWithIL.cpp | 8 ++ .../program/urProgramGetBuildInfo.cpp | 23 +++++ test/conformance/program/urProgramGetInfo.cpp | 41 ++++++++ .../urProgramSetSpecializationConstants.cpp | 94 +++++++++++++++++++ 17 files changed, 332 insertions(+), 1 deletion(-) create mode 100644 test/adapters/level_zero/CMakeLists.txt create mode 100644 test/adapters/level_zero/urProgramLink.cpp create mode 100644 test/conformance/device_code/spec_constant_multiple.cpp diff --git a/test/adapters/CMakeLists.txt b/test/adapters/CMakeLists.txt index 768a70d879..adb4f60c01 100644 --- 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_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() diff --git a/test/adapters/level_zero/CMakeLists.txt b/test/adapters/level_zero/CMakeLists.txt new file mode 100644 index 0000000000..6e2837ae40 --- /dev/null +++ b/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_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) diff --git a/test/adapters/level_zero/urProgramLink.cpp b/test/adapters/level_zero/urProgramLink.cpp new file mode 100644 index 0000000000..77ce3b8f49 --- /dev/null +++ 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 + +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 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 index 450733d5ed..1788b6e147 100644 --- a/test/conformance/device_code/CMakeLists.txt +++ b/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) diff --git a/test/conformance/device_code/spec_constant.cpp b/test/conformance/device_code/spec_constant.cpp index fb5c5a13a4..a5908eca10 100644 --- 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 spec_const; +constexpr specialization_id 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 new file mode 100644 index 0000000000..9a8379cdc7 --- /dev/null +++ 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 +#include + +using namespace sycl; + +constexpr specialization_id a_spec; +constexpr specialization_id b_spec; +constexpr specialization_id c_spec; + +int main() { + queue myQueue; + uint64_t out_val = 0; + buffer out(&out_val, sycl::range<1>{1}); + + myQueue.submit([&](handler &cgh) { + accessor out_acc{out, cgh, write_only}; + + cgh.set_specialization_constant(0); + cgh.set_specialization_constant(0); + cgh.set_specialization_constant(false); + + cgh.parallel_for( + out.get_range(), [=](item<1> item_id, kernel_handler h) { + uint32_t a = h.get_specialization_constant(); + uint64_t b = h.get_specialization_constant(); + bool c = h.get_specialization_constant(); + 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 index 4d9e3d3536..54baf7630e 100644 --- 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 index add18cd27d..81fb28a77a 100644 --- a/test/conformance/kernel/urKernelGetInfo.cpp +++ b/test/conformance/kernel/urKernelGetInfo.cpp @@ -15,6 +15,14 @@ UUR_TEST_SUITE_P( UR_KERNEL_INFO_NUM_REGS), uur::deviceTestWithParamPrinter); +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 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); +} diff --git a/test/conformance/program/program_adapter_cuda.match b/test/conformance/program/program_adapter_cuda.match index 77f14ed0bb..d353284dc2 100644 --- a/test/conformance/program/program_adapter_cuda.match +++ b/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 @@ -9,6 +11,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 +25,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 index 9aa6e56ef6..2329935d87 100644 --- a/test/conformance/program/program_adapter_hip.match +++ b/test/conformance/program/program_adapter_hip.match @@ -1,3 +1,4 @@ +urProgramCreateWithBinaryTest.BuildInvalidProgramBinary/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 @@ -20,6 +22,7 @@ {{OPT}}urProgramGetInfoTest.InvalidNullHandleProgram/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_INFO_BINARIES {{OPT}}urProgramGetInfoTest.InvalidNullHandleProgram/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_INFO_NUM_KERNELS {{OPT}}urProgramGetInfoTest.InvalidNullHandleProgram/AMD_HIP_BACKEND___{{.*}}___UR_PROGRAM_INFO_KERNEL_NAMES +urProgramGetInfoSingleTest.NumDevicesMatchesDeviceArray/AMD_HIP_BACKEND___{{.*}}_ {{OPT}}urProgramLinkTest.Success/AMD_HIP_BACKEND___{{.*}}_ {{OPT}}urProgramSetSpecializationConstantsTest.Success/AMD_HIP_BACKEND___{{.*}}_ {{OPT}}{{Segmentation fault|Aborted}} diff --git a/test/conformance/program/program_adapter_native_cpu.match b/test/conformance/program/program_adapter_native_cpu.match index fa17ed17f7..09a414d04d 100644 --- a/test/conformance/program/program_adapter_native_cpu.match +++ b/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,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_ diff --git a/test/conformance/program/program_adapter_opencl.match b/test/conformance/program/program_adapter_opencl.match index 0d429016ee..d35966d283 100644 --- a/test/conformance/program/program_adapter_opencl.match +++ b/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 diff --git a/test/conformance/program/urProgramCreateWithBinary.cpp b/test/conformance/program/urProgramCreateWithBinary.cpp index 3fb6e3a268..4a20651b52 100644 --- a/test/conformance/program/urProgramCreateWithBinary.cpp +++ b/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)); +} diff --git a/test/conformance/program/urProgramCreateWithIL.cpp b/test/conformance/program/urProgramCreateWithIL.cpp index 00f41da6ef..2d8d978ce6 100644 --- a/test/conformance/program/urProgramCreateWithIL.cpp +++ b/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)); +} diff --git a/test/conformance/program/urProgramGetBuildInfo.cpp b/test/conformance/program/urProgramGetBuildInfo.cpp index ddc0ff998c..ac56f42ea1 100644 --- a/test/conformance/program/urProgramGetBuildInfo.cpp +++ b/test/conformance/program/urProgramGetBuildInfo.cpp @@ -21,6 +21,15 @@ UUR_TEST_SUITE_P(urProgramGetBuildInfoTest, UR_PROGRAM_BUILD_INFO_BINARY_TYPE), uur::deviceTestWithParamPrinter); +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 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'); +} diff --git a/test/conformance/program/urProgramGetInfo.cpp b/test/conformance/program/urProgramGetInfo.cpp index 80d00072e7..6505a376f4 100644 --- a/test/conformance/program/urProgramGetInfo.cpp +++ b/test/conformance/program/urProgramGetInfo.cpp @@ -23,6 +23,15 @@ UUR_TEST_SUITE_P( UR_PROGRAM_INFO_KERNEL_NAMES), uur::deviceTestWithParamPrinter); +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); +} diff --git a/test/conformance/program/urProgramSetSpecializationConstants.cpp b/test/conformance/program/urProgramSetSpecializationConstants.cpp index 4149711889..74c602cbe3 100644 --- a/test/conformance/program/urProgramSetSpecializationConstants.cpp +++ b/test/conformance/program/urProgramSetSpecializationConstants.cpp @@ -12,11 +12,23 @@ struct urProgramSetSpecializationConstantsTest : uur::urKernelExecutionTest { } uint32_t spec_value = 42; + uint32_t default_spec_value = 1000; // Must match the one in the SYCL source ur_specialization_constant_info_t info = {0, sizeof(spec_value), &spec_value}; }; UUR_INSTANTIATE_KERNEL_TEST_SUITE_P(urProgramSetSpecializationConstantsTest); +struct urProgramSetMultipleSpecializationConstantsTest + : uur::urKernelExecutionTest { + // The types of spec constants in this program are {uint32_t, uint64_t, bool} + void SetUp() override { + program_name = "spec_constant_multiple"; + UUR_RETURN_ON_FATAL_FAILURE(urProgramTest::SetUp()); + } +}; +UUR_INSTANTIATE_KERNEL_TEST_SUITE_P( + urProgramSetMultipleSpecializationConstantsTest); + TEST_P(urProgramSetSpecializationConstantsTest, Success) { ASSERT_SUCCESS(urProgramSetSpecializationConstants(program, 1, &info)); ASSERT_SUCCESS(urProgramBuild(context, program, nullptr)); @@ -31,6 +43,88 @@ TEST_P(urProgramSetSpecializationConstantsTest, Success) { ValidateBuffer(buffer, sizeof(spec_value), spec_value); } +TEST_P(urProgramSetSpecializationConstantsTest, UseDefaultValue) { + ur_platform_backend_t backend; + ASSERT_SUCCESS(urPlatformGetInfo(platform, UR_PLATFORM_INFO_BACKEND, + sizeof(ur_platform_backend_t), &backend, + nullptr)); + if (backend == UR_PLATFORM_BACKEND_CUDA) { + GTEST_FAIL() + << "This test is known to cause crashes on Nvidia; not running."; + } + + ASSERT_SUCCESS(urProgramBuild(context, program, nullptr)); + auto entry_points = + uur::KernelsEnvironment::instance->GetEntryPointNames(program_name); + kernel_name = entry_points[0]; + ASSERT_SUCCESS(urKernelCreate(program, kernel_name.data(), &kernel)); + + ur_mem_handle_t buffer; + AddBuffer1DArg(sizeof(spec_value), &buffer); + Launch1DRange(1); + ValidateBuffer(buffer, sizeof(spec_value), default_spec_value); +} + +TEST_P(urProgramSetMultipleSpecializationConstantsTest, MultipleCalls) { + uint32_t a = 100; + uint64_t b = 200; + bool c = false; + uint64_t output = 0; + + ur_specialization_constant_info_t info_a = {0, sizeof(uint32_t), &a}; + ASSERT_SUCCESS(urProgramSetSpecializationConstants(program, 1, &info_a)); + + ur_specialization_constant_info_t info_c = {2, sizeof(bool), &c}; + ASSERT_SUCCESS(urProgramSetSpecializationConstants(program, 1, &info_c)); + + ur_specialization_constant_info_t info_b = {1, sizeof(uint64_t), &b}; + ASSERT_SUCCESS(urProgramSetSpecializationConstants(program, 1, &info_b)); + + ASSERT_SUCCESS(urProgramBuild(context, program, nullptr)); + auto entry_points = + uur::KernelsEnvironment::instance->GetEntryPointNames(program_name); + kernel_name = entry_points[0]; + ASSERT_SUCCESS(urKernelCreate(program, kernel_name.data(), &kernel)); + + ur_mem_handle_t buffer; + AddBuffer1DArg(sizeof(uint64_t), &buffer); + Launch1DRange(1); + + ASSERT_SUCCESS(urEnqueueMemBufferRead(queue, buffer, true, 0, + sizeof(uint64_t), &output, 0, nullptr, + nullptr)); + ASSERT_EQ(output, 300); +} + +TEST_P(urProgramSetMultipleSpecializationConstantsTest, SingleCall) { + uint32_t a = 200; + uint64_t b = 300; + bool c = true; + uint64_t output = 0; + + ur_specialization_constant_info_t info[3] = { + {0, sizeof(uint32_t), &a}, + {2, sizeof(bool), &c}, + {1, sizeof(uint64_t), &b}, + }; + ASSERT_SUCCESS(urProgramSetSpecializationConstants(program, 3, &info[0])); + + ASSERT_SUCCESS(urProgramBuild(context, program, nullptr)); + auto entry_points = + uur::KernelsEnvironment::instance->GetEntryPointNames(program_name); + kernel_name = entry_points[0]; + ASSERT_SUCCESS(urKernelCreate(program, kernel_name.data(), &kernel)); + + ur_mem_handle_t buffer; + AddBuffer1DArg(sizeof(uint64_t), &buffer); + Launch1DRange(1); + + ASSERT_SUCCESS(urEnqueueMemBufferRead(queue, buffer, true, 0, + sizeof(uint64_t), &output, 0, nullptr, + nullptr)); + ASSERT_EQ(output, 100); +} + TEST_P(urProgramSetSpecializationConstantsTest, InvalidNullHandleProgram) { ASSERT_EQ_RESULT(UR_RESULT_ERROR_INVALID_NULL_HANDLE, urProgramSetSpecializationConstants(nullptr, 1, &info));