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..78c9f24bf1 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 diff --git a/test/conformance/program/program_adapter_native_cpu.match b/test/conformance/program/program_adapter_native_cpu.match index fa17ed17f7..f02fc11cee 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,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_ 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..bafea3960a 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)); @@ -30,6 +42,79 @@ TEST_P(urProgramSetSpecializationConstantsTest, Success) { Launch1DRange(1); ValidateBuffer(buffer, sizeof(spec_value), spec_value); } +/* +TEST_P(urProgramSetSpecializationConstantsTest, UseDefaultValue) { + 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, @@ -46,3 +131,4 @@ TEST_P(urProgramSetSpecializationConstantsTest, InvalidSizeCount) { ASSERT_EQ_RESULT(UR_RESULT_ERROR_INVALID_SIZE, urProgramSetSpecializationConstants(program, 0, &info)); } +*/