From d38b8b4ae83824adb8dd22ffbf84ce119cc5f425 Mon Sep 17 00:00:00 2001 From: Ahmed <36049290+AhmedAmraniAkdi@users.noreply.github.com> Date: Tue, 18 Mar 2025 15:55:27 +0000 Subject: [PATCH 01/11] Change relaxed exp exp2 embedded error to 4 + floor(fabs(2x)) (#2326) Reflects the changes to the specification: https://github.com/KhronosGroup/OpenCL-Docs/pull/1318 Relaxed embedded exp and exp2 ulps will be 4 + floor(fabs(2 * x)). Reciprocal and divide are unchanged because the code already handles the embedded profile case, see unary_float.c and binary_operator_float.c. --------- Co-authored-by: Sven van Haastregt --- test_conformance/math_brute_force/function_list.cpp | 5 ++--- test_conformance/math_brute_force/unary_float.cpp | 8 +------- 2 files changed, 3 insertions(+), 10 deletions(-) diff --git a/test_conformance/math_brute_force/function_list.cpp b/test_conformance/math_brute_force/function_list.cpp index 74f29930..90731ea0 100644 --- a/test_conformance/math_brute_force/function_list.cpp +++ b/test_conformance/math_brute_force/function_list.cpp @@ -291,13 +291,12 @@ const Func functionList[] = { ENTRY(erfc, 16.0f, 16.0f, 4.0f, 4.0f, FTZ_OFF, unaryF), ENTRY(erf, 16.0f, 16.0f, 4.0f, 4.0f, FTZ_OFF, unaryF), - // relaxed error is overwritten in unary.c as it is 3+floor(fabs(2*x)) + // floor(fabs(2*x)) is added to the relaxed error in unary.c ENTRY_EXT(exp, 3.0f, 4.0f, 2.0f, 3.0f, 3.0f, FTZ_OFF, unaryF, 4.0f), - // relaxed error is overwritten in unary.c as it is 3+floor(fabs(2*x)) + // floor(fabs(2*x)) is added to the relaxed error in unary.c ENTRY_EXT(exp2, 3.0f, 4.0f, 2.0f, 3.0f, 3.0f, FTZ_OFF, unaryF, 4.0f), - // relaxed error is overwritten in unary.c as it is 3+floor(fabs(2*x)) in derived mode; // in non-derived mode it uses the ulp error for half_exp10. ENTRY_EXT(exp10, 3.0f, 4.0f, 2.0f, 3.0f, 8192.0f, FTZ_OFF, unaryF, 8192.0f), diff --git a/test_conformance/math_brute_force/unary_float.cpp b/test_conformance/math_brute_force/unary_float.cpp index ae0a0646..ee8a61b8 100644 --- a/test_conformance/math_brute_force/unary_float.cpp +++ b/test_conformance/math_brute_force/unary_float.cpp @@ -300,13 +300,7 @@ cl_int Test(cl_uint job_id, cl_uint thread_id, void *data) if (strcmp(fname, "exp") == 0 || strcmp(fname, "exp2") == 0) { - // For full profile, ULP depends on input value. - // For embedded profile, ULP comes from functionList. - if (!gIsEmbedded) - { - ulps = 3.0f + floor(fabs(2 * s[j])); - } - + ulps += floor(fabs(2 * s[j])); fail = !(fabsf(err) <= ulps); } if (strcmp(fname, "tan") == 0) From b9301d1ae8a3ce79cd8137e17a88bade22604e38 Mon Sep 17 00:00:00 2001 From: Sven van Haastregt Date: Tue, 18 Mar 2025 17:02:31 +0100 Subject: [PATCH 02/11] basic: fix -Wformat-security warnings (#2284) Fix warnings such as: test_vloadstore.cpp:330:49: error: format string is not a string literal (potentially insecure) There were no security issues here as the format string arguments do not contain any conversion specifiers and are never written to. Make that latter fact explicit to avoid the warnings. Signed-off-by: Sven van Haastregt --- test_conformance/basic/test_vloadstore.cpp | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/test_conformance/basic/test_vloadstore.cpp b/test_conformance/basic/test_vloadstore.cpp index d34ecbf9..e076c081 100644 --- a/test_conformance/basic/test_vloadstore.cpp +++ b/test_conformance/basic/test_vloadstore.cpp @@ -46,10 +46,10 @@ char load_str[128] = { 0 }; extern cl_half_rounding_mode halfRoundingMode; // clang-format off -static const char *store_pattern= "results[ tid ] = tmp;\n"; -static const char *store_patternV3 = "results[3*tid] = tmp.s0; results[3*tid+1] = tmp.s1; results[3*tid+2] = tmp.s2;\n"; -static const char *load_pattern = "sSharedStorage[ i ] = src[ i ];\n"; -static const char *load_patternV3 = "sSharedStorage[3*i] = src[ 3*i]; sSharedStorage[3*i+1] = src[3*i+1]; sSharedStorage[3*i+2] = src[3*i+2];\n"; +static const char *const store_pattern= "results[ tid ] = tmp;\n"; +static const char *const store_patternV3 = "results[3*tid] = tmp.s0; results[3*tid+1] = tmp.s1; results[3*tid+2] = tmp.s2;\n"; +static const char *const load_pattern = "sSharedStorage[ i ] = src[ i ];\n"; +static const char *const load_patternV3 = "sSharedStorage[3*i] = src[ 3*i]; sSharedStorage[3*i+1] = src[3*i+1]; sSharedStorage[3*i+2] = src[3*i+2];\n"; static const char *kernel_pattern[] = { pragma_str, "#define STYPE %s\n" From 395648cfef20c8f4c653d714fc0b07cf847ee39e Mon Sep 17 00:00:00 2001 From: Starla Huang <82885378+starlahuang@users.noreply.github.com> Date: Tue, 18 Mar 2025 16:03:13 +0000 Subject: [PATCH 03/11] Use max work item size instead (#2294) The test is using CL_DEVICE_MAX_WORK_GROUP_SIZE as first dimension of local work size. But it can be bigger than the first dimension of CL_DEVICE_MAX_WORK_ITEM_SIZEs which results in failure. This patch corrects it to query and use the first dimension of CL_DEVICE_MAX_WORK_ITEM_SIZES instead. Signed-off-by: Xing Huang --- test_conformance/api/test_sub_group_dispatch.cpp | 7 +++++-- 1 file changed, 5 insertions(+), 2 deletions(-) diff --git a/test_conformance/api/test_sub_group_dispatch.cpp b/test_conformance/api/test_sub_group_dispatch.cpp index fb0401cf..70a78f4a 100644 --- a/test_conformance/api/test_sub_group_dispatch.cpp +++ b/test_conformance/api/test_sub_group_dispatch.cpp @@ -102,10 +102,13 @@ REGISTER_TEST_VERSION(sub_group_dispatch, Version(2, 1)) out = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, sizeof(size_t), NULL, &error); test_error(error, "clCreateBuffer failed"); - error = clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, - sizeof(size_t), &max_local, NULL); + size_t max_work_item_sizes[3] = {}; + error = clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES, + sizeof(max_work_item_sizes), &max_work_item_sizes, + nullptr); test_error(error, "clGetDeviceInfo failed"); + max_local = max_work_item_sizes[0]; error = clGetDeviceInfo(device, CL_DEVICE_PLATFORM, sizeof(platform), (void *)&platform, NULL); From d4c292ed4626bfe23bb3f05596eafce074701909 Mon Sep 17 00:00:00 2001 From: Ahmed Hesham <117350656+ahesham-arm@users.noreply.github.com> Date: Tue, 18 Mar 2025 16:04:14 +0000 Subject: [PATCH 04/11] Fix broken CI on MacOS (#2341) LunarG has changed the SDK file hosted for MacOS v1.3.275.0 from a .zip file to a .dmg file with the old hyperlink transparently redirecting to the new one. The script expects a .zip archive so it fails when it downloads a .dmg file. Bump the Vulkan SDK version to 1.4.309.0, which is the latest version and is provided as a .zip archive. Signed-off-by: Ahmed Hesham --- .github/workflows/presubmit.yml | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/.github/workflows/presubmit.yml b/.github/workflows/presubmit.yml index 665799c8..2c05c369 100644 --- a/.github/workflows/presubmit.yml +++ b/.github/workflows/presubmit.yml @@ -72,7 +72,7 @@ jobs: - name: Install Vulkan SDK uses: humbletim/install-vulkan-sdk@main with: - version: 1.3.275.0 + version: 1.4.309.0 cache: true - name: Install Android NDK if: ${{ matrix.arch == 'android-arm' || matrix.arch == 'android-aarch64' }} From 2131aff0e902c97eef796cc78677a1145bb9119b Mon Sep 17 00:00:00 2001 From: Ahmed Hesham <117350656+ahesham-arm@users.noreply.github.com> Date: Tue, 18 Mar 2025 17:10:21 +0000 Subject: [PATCH 05/11] Migrate compiler suite to the new test registration framework (#2319) Contributes to #2181. Signed-off-by: Ahmed Hesham --- test_conformance/compiler/main.cpp | 97 +--- test_conformance/compiler/procs.h | 272 ----------- test_conformance/compiler/testBase.h | 32 +- .../compiler/test_async_build.cpp | 7 +- .../compiler/test_build_helpers.cpp | 91 ++-- .../compiler/test_build_options.cpp | 48 +- test_conformance/compiler/test_compile.cpp | 405 +++++++--------- .../test_compiler_defines_for_extensions.cpp | 2 +- .../compiler/test_feature_macro.cpp | 115 ++--- .../compiler/test_image_macro.cpp | 12 +- .../compiler/test_opencl_c_versions.cpp | 3 +- .../compiler/test_pragma_unroll.cpp | 91 ++-- .../compiler/test_preprocessor.cpp | 451 ++++++++++-------- .../test_unload_platform_compiler.cpp | 40 +- 14 files changed, 664 insertions(+), 1002 deletions(-) delete mode 100644 test_conformance/compiler/procs.h diff --git a/test_conformance/compiler/main.cpp b/test_conformance/compiler/main.cpp index 8f9ffb15..20a62f26 100644 --- a/test_conformance/compiler/main.cpp +++ b/test_conformance/compiler/main.cpp @@ -13,103 +13,10 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "harness/compat.h" - #include "harness/testHarness.h" -#include "procs.h" -#include -#include - -#if !defined(_WIN32) -#include -#endif - -test_definition test_list[] = { - ADD_TEST(load_program_source), - ADD_TEST(load_multistring_source), - ADD_TEST(load_two_kernel_source), - ADD_TEST(load_null_terminated_source), - ADD_TEST(load_null_terminated_multi_line_source), - ADD_TEST(load_null_terminated_partial_multi_line_source), - ADD_TEST(load_discreet_length_source), - ADD_TEST(get_program_source), - ADD_TEST(get_program_build_info), - ADD_TEST(get_program_info), - ADD_TEST(get_program_info_kernel_names), - ADD_TEST(get_program_info_mult_devices), - - ADD_TEST(large_compile), - ADD_TEST(async_build), - - ADD_TEST(options_build_optimizations), - ADD_TEST(options_build_macro), - ADD_TEST(options_build_macro_existence), - ADD_TEST(options_include_directory), - ADD_TEST(options_denorm_cache), - - ADD_TEST(preprocessor_define_udef), - ADD_TEST(preprocessor_include), - ADD_TEST(preprocessor_line_error), - ADD_TEST(preprocessor_pragma), - - ADD_TEST(opencl_c_versions), - ADD_TEST(compiler_defines_for_extensions), - ADD_TEST(image_macro), - - ADD_TEST(simple_compile_only), - ADD_TEST(simple_static_compile_only), - ADD_TEST(simple_extern_compile_only), - ADD_TEST(simple_compile_with_callback), - ADD_TEST(simple_embedded_header_compile), - ADD_TEST(simple_link_only), - ADD_TEST(two_file_regular_variable_access), - ADD_TEST(two_file_regular_struct_access), - ADD_TEST(two_file_regular_function_access), - ADD_TEST(simple_link_with_callback), - ADD_TEST(simple_embedded_header_link), - ADD_TEST(execute_after_simple_compile_and_link), - ADD_TEST(execute_after_simple_compile_and_link_no_device_info), - ADD_TEST(execute_after_simple_compile_and_link_with_defines), - ADD_TEST(execute_after_simple_compile_and_link_with_callbacks), - ADD_TEST(execute_after_simple_library_with_link), - ADD_TEST(execute_after_two_file_link), - ADD_TEST(execute_after_embedded_header_link), - ADD_TEST(execute_after_included_header_link), - ADD_TEST(execute_after_serialize_reload_object), - ADD_TEST(execute_after_serialize_reload_library), - ADD_TEST(simple_library_only), - ADD_TEST(simple_library_with_callback), - ADD_TEST(simple_library_with_link), - ADD_TEST(two_file_link), - ADD_TEST(multi_file_libraries), - ADD_TEST(multiple_files), - ADD_TEST(multiple_libraries), - ADD_TEST(multiple_files_multiple_libraries), - ADD_TEST(multiple_embedded_headers), - - ADD_TEST(program_binary_type), - ADD_TEST(compile_and_link_status_options_log), - - ADD_TEST_VERSION(pragma_unroll, Version(2, 0)), - - ADD_TEST_VERSION(features_macro, Version(3, 0)), - ADD_TEST(features_macro_coupling), - - ADD_TEST(unload_valid), - // ADD_TEST(unload_invalid), // disabling temporarily, see GitHub #977 - ADD_TEST(unload_repeated), - ADD_TEST(unload_compile_unload_link), - ADD_TEST(unload_build_unload_create_kernel), - ADD_TEST(unload_link_different), - ADD_TEST(unload_build_threaded), - ADD_TEST(unload_build_info), - ADD_TEST(unload_program_binaries), - -}; - -const int test_num = ARRAY_SIZE(test_list); int main(int argc, const char *argv[]) { - return runTestHarness(argc, argv, test_num, test_list, false, 0); + return runTestHarness(argc, argv, test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), false, 0); } diff --git a/test_conformance/compiler/procs.h b/test_conformance/compiler/procs.h deleted file mode 100644 index 13420e9d..00000000 --- a/test_conformance/compiler/procs.h +++ /dev/null @@ -1,272 +0,0 @@ -// -// Copyright (c) 2017 The Khronos Group Inc. -// -// Licensed under the Apache License, Version 2.0 (the "License"); -// you may not use this file except in compliance with the License. -// You may obtain a copy of the License at -// -// http://www.apache.org/licenses/LICENSE-2.0 -// -// Unless required by applicable law or agreed to in writing, software -// distributed under the License is distributed on an "AS IS" BASIS, -// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -// See the License for the specific language governing permissions and -// limitations under the License. -// -#include "harness/conversions.h" -#include "harness/errorHelpers.h" -#include "harness/kernelHelpers.h" -#include "harness/mt19937.h" -#include "harness/typeWrappers.h" - -// This is a macro rather than a function to be able to use and act like the -// existing test_error macro. -// -// Not all compiler tests need to use this macro, only those that don't use the -// test harness compiler helpers. -#define check_compiler_available(DEVICE) \ - { \ - cl_bool compilerAvailable = CL_FALSE; \ - cl_int error = clGetDeviceInfo((DEVICE), CL_DEVICE_COMPILER_AVAILABLE, \ - sizeof(compilerAvailable), \ - &compilerAvailable, NULL); \ - test_error(error, "Unable to query CL_DEVICE_COMPILER_AVAILABLE"); \ - if (compilerAvailable == CL_FALSE) \ - { \ - log_info("Skipping test - no compiler is available.\n"); \ - return TEST_SKIPPED_ITSELF; \ - } \ - } - -extern int test_load_program_source(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_load_multistring_source(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_load_two_kernel_source(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_load_null_terminated_source(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_load_null_terminated_multi_line_source(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_load_null_terminated_partial_multi_line_source( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_load_discreet_length_source(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_get_program_source(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_get_program_build_info(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_get_program_info(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_get_program_info_kernel_names(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_get_program_info_mult_devices(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_large_compile(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_async_build(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - -extern int test_options_build_optimizations(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_options_build_macro(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_options_build_macro_existence(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_options_include_directory(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_options_denorm_cache(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - -extern int test_preprocessor_define_udef(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_preprocessor_include(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_preprocessor_line_error(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_preprocessor_pragma(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - -extern int test_opencl_c_versions(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_compiler_defines_for_extensions(cl_device_id device, - cl_context context, - cl_command_queue queue, - int n_elems); -extern int test_image_macro(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - -extern int test_simple_compile_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_simple_static_compile_only(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_simple_extern_compile_only(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_simple_compile_with_callback(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_simple_embedded_header_compile(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); - -extern int test_simple_link_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_two_file_regular_variable_access(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_two_file_regular_struct_access(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_two_file_regular_function_access(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); - -extern int test_simple_link_with_callback(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_simple_embedded_header_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); - -extern int test_execute_after_simple_compile_and_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_execute_after_simple_compile_and_link_no_device_info( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_execute_after_simple_compile_and_link_with_defines( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_execute_after_simple_compile_and_link_with_callbacks( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_execute_after_simple_library_with_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_execute_after_two_file_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_execute_after_embedded_header_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_execute_after_included_header_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_execute_after_serialize_reload_object(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_execute_after_serialize_reload_library(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); - -extern int test_simple_library_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_simple_library_with_callback(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_simple_library_with_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_two_file_link(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_multi_file_libraries(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_multiple_libraries(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_multiple_files(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_multiple_files_multiple_libraries(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_multiple_embedded_headers(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); - -extern int test_program_binary_type(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_compile_and_link_status_options_log(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); - -extern int test_pragma_unroll(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_features_macro(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_features_macro_coupling(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_unload_valid(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_unload_invalid(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_unload_repeated(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_unload_compile_unload_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_unload_build_unload_create_kernel(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_unload_link_different(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_unload_build_threaded(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_unload_build_info(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_unload_program_binaries(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); diff --git a/test_conformance/compiler/testBase.h b/test_conformance/compiler/testBase.h index 7edfb2da..70c42b5d 100644 --- a/test_conformance/compiler/testBase.h +++ b/test_conformance/compiler/testBase.h @@ -16,14 +16,34 @@ #ifndef _testBase_h #define _testBase_h -#include "harness/compat.h" - -#include -#include -#include #include +#include + +#include + +#include "harness/conversions.h" +#include "harness/testHarness.h" +#include "harness/typeWrappers.h" + +// This is a macro rather than a function to be able to use and act like the +// existing test_error macro. +// +// Not all compiler tests need to use this macro, only those that don't use the +// test harness compiler helpers. +#define check_compiler_available(DEVICE) \ + { \ + cl_bool compilerAvailable = CL_FALSE; \ + cl_int error = clGetDeviceInfo((DEVICE), CL_DEVICE_COMPILER_AVAILABLE, \ + sizeof(compilerAvailable), \ + &compilerAvailable, NULL); \ + test_error(error, "Unable to query CL_DEVICE_COMPILER_AVAILABLE"); \ + if (compilerAvailable == CL_FALSE) \ + { \ + log_info("Skipping test - no compiler is available.\n"); \ + return TEST_SKIPPED_ITSELF; \ + } \ + } -#include "procs.h" // scope guard helper to ensure proper releasing of sub devices struct SubDevicesScopeGuarded diff --git a/test_conformance/compiler/test_async_build.cpp b/test_conformance/compiler/test_async_build.cpp index d1533624..d6c4cf51 100644 --- a/test_conformance/compiler/test_async_build.cpp +++ b/test_conformance/compiler/test_async_build.cpp @@ -94,8 +94,7 @@ void CL_CALLBACK test_notify_build_complete(cl_program program, void *userData) } } -int test_async_build(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(async_build) { cl_int error; @@ -119,9 +118,9 @@ int test_async_build(cl_device_id deviceID, cl_context context, test_error(error, "Unable to create program from source"); // Start an asynchronous build, registering the completion callback - TestData testData = { deviceID, testDef.expectedStatus }; + TestData testData = { device, testDef.expectedStatus }; callbackResult = 0; - error = clBuildProgram(program, 1, &deviceID, NULL, + error = clBuildProgram(program, 1, &device, NULL, test_notify_build_complete, (void *)&testData); // Allow implementations to return synchronous build failures. // They still need to call the callback. diff --git a/test_conformance/compiler/test_build_helpers.cpp b/test_conformance/compiler/test_build_helpers.cpp index 05dac356..a4945350 100644 --- a/test_conformance/compiler/test_build_helpers.cpp +++ b/test_conformance/compiler/test_build_helpers.cpp @@ -88,7 +88,7 @@ __kernel void sample_test_C(__global float *src, __global int *dst) )"; -int test_load_program_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_program_source) { int error; clProgramWrapper program; @@ -132,7 +132,7 @@ int test_load_program_source(cl_device_id deviceID, cl_context context, cl_comma return 0; } -int test_load_multistring_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_multistring_source) { int error; clProgramWrapper program; @@ -159,7 +159,7 @@ int test_load_multistring_source(cl_device_id deviceID, cl_context context, cl_c } /* Try compiling */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build multi-line program source" ); /* Should probably check binary here to verify the same results... */ @@ -169,7 +169,7 @@ int test_load_multistring_source(cl_device_id deviceID, cl_context context, cl_c return 0; } -int test_load_two_kernel_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_two_kernel_source) { int error; cl_program program; @@ -195,7 +195,7 @@ int test_load_two_kernel_source(cl_device_id deviceID, cl_context context, cl_co } /* Try compiling */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build two-kernel program source" ); /* Should probably check binary here to verify the same results... */ @@ -207,7 +207,7 @@ int test_load_two_kernel_source(cl_device_id deviceID, cl_context context, cl_co return 0; } -int test_load_null_terminated_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_null_terminated_source) { int error; cl_program program; @@ -222,7 +222,7 @@ int test_load_null_terminated_source(cl_device_id deviceID, cl_context context, } /* Try compiling */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build null-terminated program source" ); /* Should probably check binary here to verify the same results... */ @@ -234,7 +234,7 @@ int test_load_null_terminated_source(cl_device_id deviceID, cl_context context, return 0; } -int test_load_null_terminated_multi_line_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_null_terminated_multi_line_source) { int error; cl_program program; @@ -252,7 +252,7 @@ int test_load_null_terminated_multi_line_source(cl_device_id deviceID, cl_contex } /* Try compiling */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build null-terminated program source" ); /* Should probably check binary here to verify the same results... */ @@ -264,7 +264,7 @@ int test_load_null_terminated_multi_line_source(cl_device_id deviceID, cl_contex return 0; } -int test_load_discreet_length_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_discreet_length_source) { int error; cl_program program; @@ -295,7 +295,7 @@ int test_load_discreet_length_source(cl_device_id deviceID, cl_context context, } /* Try compiling */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build null-terminated program source" ); /* Should probably check binary here to verify the same results... */ @@ -307,7 +307,7 @@ int test_load_discreet_length_source(cl_device_id deviceID, cl_context context, return 0; } -int test_load_null_terminated_partial_multi_line_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(load_null_terminated_partial_multi_line_source) { int error; cl_program program; @@ -337,7 +337,7 @@ int test_load_null_terminated_partial_multi_line_source(cl_device_id deviceID, c } /* Try compiling */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build null-terminated program source" ); /* Should probably check binary here to verify the same results... */ @@ -349,7 +349,7 @@ int test_load_null_terminated_partial_multi_line_source(cl_device_id deviceID, c return 0; } -int test_get_program_info(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(get_program_info) { int error; cl_program program; @@ -373,7 +373,7 @@ int test_get_program_info(cl_device_id deviceID, cl_context context, cl_command_ test_error( error, "Unable to get device of program" ); /* Object comparability test. */ - test_assert_error(device1 == deviceID, + test_assert_error(device1 == device, "Unexpected result returned by CL_PROGRAM_DEVICES query"); cl_uint devCount; @@ -439,9 +439,7 @@ int test_get_program_info(cl_device_id deviceID, cl_context context, cl_command_ return 0; } -int test_get_program_info_kernel_names(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(get_program_info_kernel_names) { int error = CL_SUCCESS; size_t total_kernels = 0; @@ -474,8 +472,7 @@ int test_get_program_info_kernel_names(cl_device_id deviceID, // returned. Query CL_PROGRAM_KERNEL_NAMES and check that the right // kernel names are returned. { - error = - clBuildProgram(program, 1, &deviceID, nullptr, nullptr, nullptr); + error = clBuildProgram(program, 1, &device, nullptr, nullptr, nullptr); test_error(error, "clBuildProgram failed"); error = clGetProgramInfo(program, CL_PROGRAM_NUM_KERNELS, @@ -517,7 +514,7 @@ int test_get_program_info_kernel_names(cl_device_id deviceID, // kernel names are returned. { const char *build_options = "-DUSE_SAMPLE_TEST_B"; - error = clBuildProgram(program, 1, &deviceID, build_options, nullptr, + error = clBuildProgram(program, 1, &device, build_options, nullptr, nullptr); test_error(error, "clBuildProgram failed"); @@ -553,14 +550,12 @@ int test_get_program_info_kernel_names(cl_device_id deviceID, return CL_SUCCESS; } -int test_get_program_info_mult_devices(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(get_program_info_mult_devices) { size_t size = 0; // query multi-device context and perform objects comparability test - cl_int err = clGetDeviceInfo(deviceID, CL_DEVICE_PARTITION_PROPERTIES, 0, + cl_int err = clGetDeviceInfo(device, CL_DEVICE_PARTITION_PROPERTIES, 0, nullptr, &size); test_error_fail(err, "clGetDeviceInfo failed"); @@ -572,7 +567,7 @@ int test_get_program_info_mult_devices(cl_device_id deviceID, std::vector supported_props( size / sizeof(cl_device_partition_property), 0); - err = clGetDeviceInfo(deviceID, CL_DEVICE_PARTITION_PROPERTIES, + err = clGetDeviceInfo(device, CL_DEVICE_PARTITION_PROPERTIES, supported_props.size() * sizeof(cl_device_partition_property), supported_props.data(), &size); @@ -585,7 +580,7 @@ int test_get_program_info_mult_devices(cl_device_id deviceID, } cl_uint maxComputeUnits = 0; - err = clGetDeviceInfo(deviceID, CL_DEVICE_MAX_COMPUTE_UNITS, + err = clGetDeviceInfo(device, CL_DEVICE_MAX_COMPUTE_UNITS, sizeof(maxComputeUnits), &maxComputeUnits, nullptr); test_error_ret(err, "Unable to get maximal number of compute units", TEST_FAIL); @@ -607,14 +602,14 @@ int test_get_program_info_mult_devices(cl_device_id deviceID, if (sup_prop == prop[0]) { // how many sub-devices can we create? - err = clCreateSubDevices(deviceID, prop.data(), 0, nullptr, + err = clCreateSubDevices(device, prop.data(), 0, nullptr, &num_devices); test_error_fail(err, "clCreateSubDevices failed"); if (num_devices < 2) continue; // get the list of subDevices scope_guard.reset(new SubDevicesScopeGuarded(num_devices)); - err = clCreateSubDevices(deviceID, prop.data(), num_devices, + err = clCreateSubDevices(device, prop.data(), num_devices, scope_guard->sub_devices.data(), &num_devices); test_error_fail(err, "clCreateSubDevices failed"); @@ -681,7 +676,7 @@ int test_get_program_info_mult_devices(cl_device_id deviceID, return TEST_PASS; } -int test_get_program_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(get_program_source) { cl_program program; int error; @@ -736,7 +731,7 @@ int test_get_program_source(cl_device_id deviceID, cl_context context, cl_comman return 0; } -int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(get_program_build_info) { cl_program program; int error; @@ -753,7 +748,8 @@ int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_co } /* Make sure getting the length works */ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, 0, NULL, &length ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_STATUS, 0, + NULL, &length); test_error( error, "Unable to get program build status length" ); if( length != sizeof( status ) ) { @@ -762,10 +758,11 @@ int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_co } /* Now actually build it and verify the status */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Unable to build program source" ); - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_STATUS, + sizeof(status), &status, NULL); test_error( error, "Unable to get program build status" ); if( status != CL_BUILD_SUCCESS ) { @@ -776,7 +773,8 @@ int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_co /***** Build log *****/ /* Try getting the length */ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &length ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 0, + NULL, &length); test_error( error, "Unable to get program build log length" ); log_info("Build log is %zu long.\n", length); @@ -784,7 +782,8 @@ int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_co buffer = (char*)malloc(length); /* Try normal source */ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, length, buffer, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, length, + buffer, NULL); test_error( error, "Unable to get program build log" ); if( buffer[length-1] != '\0' ) @@ -794,23 +793,27 @@ int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_co } /* Try both at once */ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, length, buffer, &newLength ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, length, + buffer, &newLength); test_error( error, "Unable to get program build log" ); free(buffer); /***** Build options *****/ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &length ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_OPTIONS, 0, + NULL, &length); test_error( error, "Unable to get program build options length" ); buffer = (char*)malloc(length); /* Try normal source */ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, length, buffer, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_OPTIONS, + length, buffer, NULL); test_error( error, "Unable to get program build options" ); /* Try both at once */ - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, length, buffer, &newLength ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_OPTIONS, + length, buffer, &newLength); test_error( error, "Unable to get program build options" ); free(buffer); @@ -826,19 +829,21 @@ int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_co return -1; } - error = clBuildProgram( program, 1, &deviceID, "-cl-opt-disable", NULL, NULL ); + error = clBuildProgram(program, 1, &device, "-cl-opt-disable", NULL, NULL); if( error != CL_SUCCESS ) { print_error( error, "Building with valid options failed!" ); return -1; } - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &length ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_OPTIONS, 0, + NULL, &length); test_error( error, "Unable to get program build options" ); buffer = (char*)malloc(length); - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, length, buffer, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_OPTIONS, + length, buffer, NULL); test_error( error, "Unable to get program build options" ); if( strcmp( (char *)buffer, "-cl-opt-disable" ) != 0 ) { diff --git a/test_conformance/compiler/test_build_options.cpp b/test_conformance/compiler/test_build_options.cpp index c25fd10f..daba1b0e 100644 --- a/test_conformance/compiler/test_build_options.cpp +++ b/test_conformance/compiler/test_build_options.cpp @@ -88,7 +88,7 @@ cl_int get_result_from_program( cl_context context, cl_command_queue queue, cl_p return CL_SUCCESS; } -int test_options_build_optimizations(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(options_build_optimizations) { int error; cl_build_status status; @@ -105,10 +105,12 @@ int test_options_build_optimizations(cl_device_id deviceID, cl_context context, /* Build with the macro defined */ log_info("Testing optimization option '%s'\n", optimization_options[i]); - error = clBuildProgram( program, 1, &deviceID, optimization_options[i], NULL, NULL ); + error = clBuildProgram(program, 1, &device, optimization_options[i], + NULL, NULL); test_error( error, "Test program did not properly build" ); - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_STATUS, + sizeof(status), &status, NULL); test_error( error, "Unable to get program build status" ); if( (int)status != CL_BUILD_SUCCESS ) @@ -121,7 +123,7 @@ int test_options_build_optimizations(cl_device_id deviceID, cl_context context, return 0; } -int test_options_build_macro(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(options_build_macro) { int error; clProgramWrapper program; @@ -136,10 +138,11 @@ int test_options_build_macro(cl_device_id deviceID, cl_context context, cl_comma } /* Build with the macro defined */ - error = clBuildProgram( program, 1, &deviceID, "-DTEST_MACRO=1 ", NULL, NULL ); + error = clBuildProgram(program, 1, &device, "-DTEST_MACRO=1 ", NULL, NULL); test_error( error, "Test program did not properly build" ); - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_STATUS, + sizeof(status), &status, NULL); test_error( error, "Unable to get program build status" ); if( (int)status != CL_BUILD_SUCCESS ) @@ -162,7 +165,7 @@ int test_options_build_macro(cl_device_id deviceID, cl_context context, cl_comma } // Rebuild with a different value for the define macro, to make sure caching behaves properly - error = clBuildProgram( program, 1, &deviceID, "-DTEST_MACRO=5 ", NULL, NULL ); + error = clBuildProgram(program, 1, &device, "-DTEST_MACRO=5 ", NULL, NULL); test_error( error, "Test program did not properly rebuild" ); error = get_result_from_program( context, queue, program, &secondResult ); @@ -180,7 +183,7 @@ int test_options_build_macro(cl_device_id deviceID, cl_context context, cl_comma return 0; } -int test_options_build_macro_existence(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(options_build_macro_existence) { int error; clProgramWrapper program; @@ -195,7 +198,7 @@ int test_options_build_macro_existence(cl_device_id deviceID, cl_context context } /* Build without the macro defined */ - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Test program did not properly build" ); // Go ahead and run the program to verify results @@ -211,7 +214,7 @@ int test_options_build_macro_existence(cl_device_id deviceID, cl_context context } // Now compile again with the macro defined and verify a change in results - error = clBuildProgram( program, 1, &deviceID, "-DTEST_MACRO", NULL, NULL ); + error = clBuildProgram(program, 1, &device, "-DTEST_MACRO", NULL, NULL); test_error( error, "Test program did not properly build" ); error = get_result_from_program( context, queue, program, &secondResult ); @@ -229,7 +232,7 @@ int test_options_build_macro_existence(cl_device_id deviceID, cl_context context return 0; } -int test_options_include_directory(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(options_include_directory) { int error; @@ -252,10 +255,12 @@ int test_options_include_directory(cl_device_id deviceID, cl_context context, cl include_dir = "-I " + path + sep + "includeTestDirectory"; // log_info("%s\n", include_dir); - error = clBuildProgram( program, 1, &deviceID, include_dir.c_str(), NULL, NULL ); + error = + clBuildProgram(program, 1, &device, include_dir.c_str(), NULL, NULL); test_error( error, "Test program did not properly build" ); - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_STATUS, + sizeof(status), &status, NULL); test_error( error, "Unable to get program build status" ); if( (int)status != CL_BUILD_SUCCESS ) @@ -278,7 +283,8 @@ int test_options_include_directory(cl_device_id deviceID, cl_context context, cl // Rebuild with a different include directory include_dir = "-I " + path + sep + "secondIncludeTestDirectory"; - error = clBuildProgram( program, 1, &deviceID, include_dir.c_str(), NULL, NULL ); + error = + clBuildProgram(program, 1, &device, include_dir.c_str(), NULL, NULL); test_error( error, "Test program did not properly rebuild" ); error = get_result_from_program( context, queue, program, &secondResult ); @@ -334,7 +340,7 @@ cl_int get_float_result_from_program( cl_context context, cl_command_queue queue return CL_SUCCESS; } -int test_options_denorm_cache(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(options_denorm_cache) { int error; @@ -344,7 +350,8 @@ int test_options_denorm_cache(cl_device_id deviceID, cl_context context, cl_comm // If denorms aren't even supported, testing this flag is pointless cl_device_fp_config floatCaps = 0; - error = clGetDeviceInfo( deviceID, CL_DEVICE_SINGLE_FP_CONFIG, sizeof(floatCaps), &floatCaps, NULL); + error = clGetDeviceInfo(device, CL_DEVICE_SINGLE_FP_CONFIG, + sizeof(floatCaps), &floatCaps, NULL); test_error( error, "Unable to get device FP config" ); if( ( floatCaps & CL_FP_DENORM ) == 0 ) { @@ -356,10 +363,12 @@ int test_options_denorm_cache(cl_device_id deviceID, cl_context context, cl_comm test_error( error, "Unable to create test program" ); // Build first WITH the denorm flush flag - error = clBuildProgram( program, 1, &deviceID, "-cl-denorms-are-zero", NULL, NULL ); + error = + clBuildProgram(program, 1, &device, "-cl-denorms-are-zero", NULL, NULL); test_error( error, "Test program did not properly build" ); - error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL ); + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_STATUS, + sizeof(status), &status, NULL); test_error( error, "Unable to get program build status" ); if( (int)status != CL_BUILD_SUCCESS ) @@ -382,7 +391,7 @@ int test_options_denorm_cache(cl_device_id deviceID, cl_context context, cl_comm // valid, there isn't anything we can to do validate results for now // Rebuild without flushing flag set - error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error( error, "Test program did not properly rebuild" ); error = get_float_result_from_program( context, queue, program, *input, *input, &secondResult ); @@ -406,4 +415,3 @@ int test_options_denorm_cache(cl_device_id deviceID, cl_context context, cl_comm return 0; } - diff --git a/test_conformance/compiler/test_compile.cpp b/test_conformance/compiler/test_compile.cpp index 0ca37b3f..907de9cb 100644 --- a/test_conformance/compiler/test_compile.cpp +++ b/test_conformance/compiler/test_compile.cpp @@ -155,8 +155,8 @@ const char *link_static_function_access = // use with compile_static_function "extern int foo(int, int);\n" "int access_foo() { int blah = foo(3, 4); return blah + 5; }\n"; -int test_large_single_compile(cl_context context, cl_device_id deviceID, - unsigned int numLines) +static int test_large_single_compile(cl_context context, cl_device_id deviceID, + unsigned int numLines) { int error; cl_program program; @@ -219,8 +219,7 @@ int test_large_single_compile(cl_context context, cl_device_id deviceID, return 0; } -int test_large_compile(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(large_compile) { unsigned int toTest[] = { 64, 128, 256, 512, 1024, 2048, 4096, 0 @@ -240,7 +239,7 @@ int test_large_compile(cl_device_id deviceID, cl_context context, gettimeofday(&time1, NULL); #endif - if (test_large_single_compile(context, deviceID, toTest[i]) != 0) + if (test_large_single_compile(context, device, toTest[i]) != 0) { log_error( "ERROR: long program test failed for %d lines! (in %s:%d)\n", @@ -270,8 +269,10 @@ static int verifyCopyBuffer(cl_context context, cl_command_queue queue, #define _strdup strdup #endif -int test_large_multi_file_library(cl_context context, cl_device_id deviceID, - cl_command_queue queue, unsigned int numLines) +static int test_large_multi_file_library(cl_context context, + cl_device_id deviceID, + cl_command_queue queue, + unsigned int numLines) { int error; cl_program program; @@ -419,8 +420,7 @@ int test_large_multi_file_library(cl_context context, cl_device_id deviceID, return 0; } -int test_multi_file_libraries(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(multi_file_libraries) { unsigned int toTest[] = { 2, 4, 8, 16, 32, 64, 128, 256, 0 @@ -440,7 +440,7 @@ int test_multi_file_libraries(cl_device_id deviceID, cl_context context, gettimeofday(&time1, NULL); #endif - if (test_large_multi_file_library(context, deviceID, queue, toTest[i]) + if (test_large_multi_file_library(context, device, queue, toTest[i]) != 0) { log_error("ERROR: multi-file library program test failed for %d " @@ -464,10 +464,10 @@ int test_multi_file_libraries(cl_device_id deviceID, cl_context context, return 0; } -int test_large_multiple_embedded_headers(cl_context context, - cl_device_id deviceID, - cl_command_queue queue, - unsigned int numLines) +static int test_large_multiple_embedded_headers(cl_context context, + cl_device_id deviceID, + cl_command_queue queue, + unsigned int numLines) { int error; cl_program program; @@ -641,8 +641,7 @@ int test_large_multiple_embedded_headers(cl_context context, return 0; } -int test_multiple_embedded_headers(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(multiple_embedded_headers) { unsigned int toTest[] = { 2, 4, 8, 16, 32, 64, 128, 256, 0 @@ -663,7 +662,7 @@ int test_multiple_embedded_headers(cl_device_id deviceID, cl_context context, gettimeofday(&time1, NULL); #endif - if (test_large_multiple_embedded_headers(context, deviceID, queue, + if (test_large_multiple_embedded_headers(context, device, queue, toTest[i]) != 0) { @@ -690,8 +689,10 @@ int test_multiple_embedded_headers(cl_device_id deviceID, cl_context context, double logbase(double a, double base) { return log(a) / log(base); } -int test_large_multiple_libraries(cl_context context, cl_device_id deviceID, - cl_command_queue queue, unsigned int numLines) +static int test_large_multiple_libraries(cl_context context, + cl_device_id deviceID, + cl_command_queue queue, + unsigned int numLines) { int error; cl_program *simple_kernels; @@ -846,8 +847,7 @@ int test_large_multiple_libraries(cl_context context, cl_device_id deviceID, return 0; } -int test_multiple_libraries(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(multiple_libraries) { unsigned int toTest[] = { 2, 8, 32, 128, 256, 0 @@ -867,7 +867,7 @@ int test_multiple_libraries(cl_device_id deviceID, cl_context context, gettimeofday(&time1, NULL); #endif - if (test_large_multiple_libraries(context, deviceID, queue, toTest[i]) + if (test_large_multiple_libraries(context, device, queue, toTest[i]) != 0) { log_error("ERROR: multiple library program test failed for %d " @@ -891,10 +891,10 @@ int test_multiple_libraries(cl_device_id deviceID, cl_context context, return 0; } -int test_large_multiple_files_multiple_libraries(cl_context context, - cl_device_id deviceID, - cl_command_queue queue, - unsigned int numLines) +static int test_large_multiple_files_multiple_libraries(cl_context context, + cl_device_id deviceID, + cl_command_queue queue, + unsigned int numLines) { int error; cl_program *simple_kernels; @@ -1060,10 +1060,7 @@ int test_large_multiple_files_multiple_libraries(cl_context context, return 0; } -int test_multiple_files_multiple_libraries(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(multiple_files_multiple_libraries) { unsigned int toTest[] = { 8, 32, 128, 256, 0 }; // 512, 2048, 8192, 32768, 0 }; @@ -1083,8 +1080,8 @@ int test_multiple_files_multiple_libraries(cl_device_id deviceID, gettimeofday(&time1, NULL); #endif - if (test_large_multiple_files_multiple_libraries(context, deviceID, - queue, toTest[i]) + if (test_large_multiple_files_multiple_libraries(context, device, queue, + toTest[i]) != 0) { log_error("ERROR: multiple files, multiple libraries program test " @@ -1108,8 +1105,9 @@ int test_multiple_files_multiple_libraries(cl_device_id deviceID, return 0; } -int test_large_multiple_files(cl_context context, cl_device_id deviceID, - cl_command_queue queue, unsigned int numLines) +static int test_large_multiple_files(cl_context context, cl_device_id deviceID, + cl_command_queue queue, + unsigned int numLines) { int error; const char **lines; @@ -1232,8 +1230,7 @@ int test_large_multiple_files(cl_context context, cl_device_id deviceID, return 0; } -int test_multiple_files(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(multiple_files) { unsigned int toTest[] = { 8, 32, 128, 256, 0 }; // 512, 2048, 8192, 32768, 0 }; @@ -1253,7 +1250,7 @@ int test_multiple_files(cl_device_id deviceID, cl_context context, gettimeofday(&time1, NULL); #endif - if (test_large_multiple_files(context, deviceID, queue, toTest[i]) != 0) + if (test_large_multiple_files(context, device, queue, toTest[i]) != 0) { log_error("ERROR: multiple files program test failed for %d lines! " "(in %s:%d)\n\n", @@ -1276,8 +1273,7 @@ int test_multiple_files(cl_device_id deviceID, cl_context context, return 0; } -int test_simple_compile_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_compile_only) { int error; cl_program program; @@ -1293,8 +1289,8 @@ int test_simple_compile_only(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); /* All done! */ @@ -1304,8 +1300,7 @@ int test_simple_compile_only(cl_device_id deviceID, cl_context context, return 0; } -int test_simple_static_compile_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_static_compile_only) { int error; cl_program program; @@ -1323,8 +1318,8 @@ int test_simple_static_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling a static variable...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple static variable program"); /* All done! */ @@ -1342,8 +1337,8 @@ int test_simple_static_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling a static struct...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple static variable program"); /* All done! */ @@ -1361,8 +1356,8 @@ int test_simple_static_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling a static function...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple static function program"); /* All done! */ @@ -1372,8 +1367,7 @@ int test_simple_static_compile_only(cl_device_id deviceID, cl_context context, return 0; } -int test_simple_extern_compile_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_extern_compile_only) { int error; cl_program program; @@ -1390,8 +1384,8 @@ int test_simple_extern_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling an extern kernel...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple extern kernel program"); /* All done! */ @@ -1409,8 +1403,8 @@ int test_simple_extern_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling an extern variable...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple extern variable program"); /* All done! */ @@ -1428,8 +1422,8 @@ int test_simple_extern_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling an extern struct...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple extern variable program"); /* All done! */ @@ -1447,8 +1441,8 @@ int test_simple_extern_compile_only(cl_device_id deviceID, cl_context context, } log_info("Compiling an extern function...\n"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple extern function program"); /* All done! */ @@ -1500,8 +1494,7 @@ static void CL_CALLBACK simple_compile_callback(cl_program program, "compile_program_completion_event!\n"); } -int test_simple_compile_with_callback(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_compile_with_callback) { int error; cl_program program; @@ -1525,7 +1518,7 @@ int test_simple_compile_with_callback(cl_device_id deviceID, cl_context context, once_upon_a_midnight_dreary, compile_program_completion_event }; - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, simple_compile_callback, (void *)&simple_compile_user_data); test_error(error, "Unable to compile a simple program with a callback"); @@ -1545,10 +1538,7 @@ int test_simple_compile_with_callback(cl_device_id deviceID, cl_context context, return 0; } -int test_simple_embedded_header_compile(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(simple_embedded_header_compile) { int error; cl_program program, header; @@ -1574,7 +1564,7 @@ int test_simple_embedded_header_compile(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 1, &header, + error = clCompileProgram(program, 1, &device, NULL, 1, &header, &simple_header_name, NULL, NULL); test_error(error, "Unable to compile a simple program with embedded header"); @@ -1589,8 +1579,7 @@ int test_simple_embedded_header_compile(cl_device_id deviceID, return 0; } -int test_simple_link_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_link_only) { int error; cl_program program; @@ -1606,12 +1595,12 @@ int test_simple_link_only(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_linked_program = clLinkProgram( - context, 1, &deviceID, NULL, 1, &program, NULL, NULL, &error); + context, 1, &device, NULL, 1, &program, NULL, NULL, &error); test_error(error, "Unable to link a simple program"); /* All done! */ @@ -1624,10 +1613,7 @@ int test_simple_link_only(cl_device_id deviceID, cl_context context, return 0; } -int test_two_file_regular_variable_access(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(two_file_regular_variable_access) { int error; cl_program program, second_program, my_newly_linked_program; @@ -1647,8 +1633,8 @@ int test_two_file_regular_variable_access(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program with regular function"); @@ -1662,14 +1648,14 @@ int test_two_file_regular_variable_access(cl_device_id deviceID, return -1; } - error = clCompileProgram(second_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(second_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error( error, "Unable to compile a program that tries to access a regular variable"); cl_program two_programs[2] = { program, second_program }; - my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, + my_newly_linked_program = clLinkProgram(context, 1, &device, NULL, 2, two_programs, NULL, NULL, &error); test_error(error, "clLinkProgram: Expected a different error code while linking a " @@ -1688,10 +1674,7 @@ int test_two_file_regular_variable_access(cl_device_id deviceID, return 0; } -int test_two_file_regular_struct_access(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(two_file_regular_struct_access) { int error; cl_program program, second_program, my_newly_linked_program; @@ -1711,8 +1694,8 @@ int test_two_file_regular_struct_access(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program with regular struct"); error = create_single_kernel_helper_create_program( @@ -1725,14 +1708,14 @@ int test_two_file_regular_struct_access(cl_device_id deviceID, return -1; } - error = clCompileProgram(second_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(second_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error( error, "Unable to compile a program that tries to access a regular struct"); cl_program two_programs[2] = { program, second_program }; - my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, + my_newly_linked_program = clLinkProgram(context, 1, &device, NULL, 2, two_programs, NULL, NULL, &error); test_error(error, "clLinkProgram: Expected a different error code while linking a " @@ -1752,10 +1735,7 @@ int test_two_file_regular_struct_access(cl_device_id deviceID, } -int test_two_file_regular_function_access(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(two_file_regular_function_access) { int error; cl_program program, second_program, my_newly_linked_program; @@ -1775,8 +1755,8 @@ int test_two_file_regular_function_access(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program with regular function"); @@ -1790,14 +1770,14 @@ int test_two_file_regular_function_access(cl_device_id deviceID, return -1; } - error = clCompileProgram(second_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(second_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error( error, "Unable to compile a program that tries to access a regular function"); cl_program two_programs[2] = { program, second_program }; - my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, + my_newly_linked_program = clLinkProgram(context, 1, &device, NULL, 2, two_programs, NULL, NULL, &error); test_error(error, "clLinkProgram: Expected a different error code while linking a " @@ -1816,8 +1796,7 @@ int test_two_file_regular_function_access(cl_device_id deviceID, return 0; } -int test_simple_embedded_header_link(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_embedded_header_link) { int error; cl_program program, header, simple_program; @@ -1843,7 +1822,7 @@ int test_simple_embedded_header_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 1, &header, + error = clCompileProgram(program, 1, &device, NULL, 1, &header, &simple_header_name, NULL, NULL); test_error(error, "Unable to compile a simple program with embedded header"); @@ -1858,13 +1837,13 @@ int test_simple_embedded_header_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(simple_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(simple_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program two_programs[2] = { program, simple_program }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error); + context, 1, &device, "", 2, two_programs, NULL, NULL, &error); test_error(error, "Unable to create an executable from two binaries, one compiled " "with embedded header"); @@ -1920,8 +1899,7 @@ static void CL_CALLBACK simple_link_callback(cl_program program, "link_program_completion_event event!\n"); } -int test_simple_link_with_callback(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_link_with_callback) { int error; cl_program program; @@ -1938,8 +1916,8 @@ int test_simple_link_with_callback(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); link_program_completion_event = clCreateUserEvent(context, &error); @@ -1949,7 +1927,7 @@ int test_simple_link_with_callback(cl_device_id deviceID, cl_context context, link_program_completion_event }; cl_program my_linked_library = clLinkProgram( - context, 1, &deviceID, NULL, 1, &program, simple_link_callback, + context, 1, &device, NULL, 1, &program, simple_link_callback, (void *)&simple_link_user_data, &error); test_error(error, "Unable to link a simple program"); @@ -2108,10 +2086,7 @@ static int verifyCopyBuffer(cl_context context, cl_command_queue queue, return result; } -int test_execute_after_simple_compile_and_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_simple_compile_and_link) { int error; cl_program program; @@ -2127,12 +2102,12 @@ int test_execute_after_simple_compile_and_link(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_linked_program = clLinkProgram( - context, 1, &deviceID, NULL, 1, &program, NULL, NULL, &error); + context, 1, &device, NULL, 1, &program, NULL, NULL, &error); test_error(error, "Unable to link a simple program"); cl_kernel kernel = @@ -2155,9 +2130,7 @@ int test_execute_after_simple_compile_and_link(cl_device_id deviceID, return 0; } -int test_execute_after_simple_compile_and_link_no_device_info( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_simple_compile_and_link_no_device_info) { int error; cl_program program; @@ -2201,9 +2174,7 @@ int test_execute_after_simple_compile_and_link_no_device_info( return 0; } -int test_execute_after_simple_compile_and_link_with_defines( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_simple_compile_and_link_with_defines) { int error; cl_program program; @@ -2221,12 +2192,12 @@ int test_execute_after_simple_compile_and_link_with_defines( return -1; } - error = clCompileProgram(program, 1, &deviceID, "-DFIRST=5 -DSECOND=37", 0, + error = clCompileProgram(program, 1, &device, "-DFIRST=5 -DSECOND=37", 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_linked_program = clLinkProgram( - context, 1, &deviceID, NULL, 1, &program, NULL, NULL, &error); + context, 1, &device, NULL, 1, &program, NULL, NULL, &error); test_error(error, "Unable to link a simple program"); cl_kernel kernel = @@ -2249,10 +2220,7 @@ int test_execute_after_simple_compile_and_link_with_defines( return 0; } -int test_execute_after_serialize_reload_object(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_serialize_reload_object) { int error; cl_program program; @@ -2271,8 +2239,8 @@ int test_execute_after_serialize_reload_object(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); // Get the size of the resulting binary (only one device) @@ -2320,13 +2288,12 @@ int test_execute_after_serialize_reload_object(cl_device_id deviceID, // use clCreateProgramWithBinary cl_program program_with_binary = clCreateProgramWithBinary( - context, 1, &deviceID, &binarySize, (const unsigned char **)buffers, + context, 1, &device, &binarySize, (const unsigned char **)buffers, loadErrors, &error); test_error(error, "Unable to create program with binary"); - cl_program my_newly_linked_program = - clLinkProgram(context, 1, &deviceID, NULL, 1, &program_with_binary, - NULL, NULL, &error); + cl_program my_newly_linked_program = clLinkProgram( + context, 1, &device, NULL, 1, &program_with_binary, NULL, NULL, &error); test_error(error, "Unable to link a simple program"); cl_kernel kernel = @@ -2354,10 +2321,7 @@ int test_execute_after_serialize_reload_object(cl_device_id deviceID, return 0; } -int test_execute_after_serialize_reload_library(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_serialize_reload_library) { int error; cl_program program, another_program; @@ -2377,13 +2341,13 @@ int test_execute_after_serialize_reload_library(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_minted_library = - clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, - NULL, NULL, &error); + clLinkProgram(context, 1, &device, "-create-library", 1, &program, NULL, + NULL, &error); test_error(error, "Unable to create a simple library"); @@ -2432,7 +2396,7 @@ int test_execute_after_serialize_reload_library(cl_device_id deviceID, // use clCreateProgramWithBinary cl_program library_with_binary = clCreateProgramWithBinary( - context, 1, &deviceID, &binarySize, (const unsigned char **)buffers, + context, 1, &device, &binarySize, (const unsigned char **)buffers, loadErrors, &error); test_error(error, "Unable to create program with binary"); @@ -2446,14 +2410,14 @@ int test_execute_after_serialize_reload_library(cl_device_id deviceID, return -1; } - error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(another_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program program_and_archive[2] = { another_program, library_with_binary }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error); + context, 1, &device, "", 2, program_and_archive, NULL, NULL, &error); test_error(error, "Unable to create an executable from a binary and a library"); @@ -2540,9 +2504,7 @@ static void CL_CALLBACK program_link_completion_callback(cl_program program, "link_program_completion_event event!\n"); } -int test_execute_after_simple_compile_and_link_with_callbacks( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_simple_compile_and_link_with_callbacks) { int error; cl_program program; @@ -2563,7 +2525,7 @@ int test_execute_after_simple_compile_and_link_with_callbacks( compile_program_completion_event = clCreateUserEvent(context, &error); test_error(error, "Unable to create a user event"); - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, program_compile_completion_callback, (void *)compile_program_completion_event); test_error(error, "Unable to compile a simple program"); @@ -2580,7 +2542,7 @@ int test_execute_after_simple_compile_and_link_with_callbacks( test_error(error, "Unable to create a user event"); cl_program my_newly_linked_program = - clLinkProgram(context, 1, &deviceID, NULL, 1, &program, + clLinkProgram(context, 1, &device, NULL, 1, &program, program_link_completion_callback, (void *)link_program_completion_event, &error); test_error(error, "Unable to link a simple program"); @@ -2613,8 +2575,7 @@ int test_execute_after_simple_compile_and_link_with_callbacks( return 0; } -int test_simple_library_only(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_library_only) { int error; cl_program program; @@ -2630,13 +2591,13 @@ int test_simple_library_only(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_minted_library = - clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, - NULL, NULL, &error); + clLinkProgram(context, 1, &device, "-create-library", 1, &program, NULL, + NULL, &error); test_error(error, "Unable to create a simple library"); /* All done! */ @@ -2649,8 +2610,7 @@ int test_simple_library_only(cl_device_id deviceID, cl_context context, return 0; } -int test_simple_library_with_callback(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_library_with_callback) { int error; cl_program program; @@ -2667,8 +2627,8 @@ int test_simple_library_with_callback(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); link_program_completion_event = clCreateUserEvent(context, &error); @@ -2678,7 +2638,7 @@ int test_simple_library_with_callback(cl_device_id deviceID, cl_context context, link_program_completion_event }; cl_program my_newly_minted_library = clLinkProgram( - context, 1, &deviceID, "-create-library", 1, &program, + context, 1, &device, "-create-library", 1, &program, simple_link_callback, (void *)&simple_link_user_data, &error); test_error(error, "Unable to create a simple library"); @@ -2700,8 +2660,7 @@ int test_simple_library_with_callback(cl_device_id deviceID, cl_context context, return 0; } -int test_simple_library_with_link(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(simple_library_with_link) { int error; cl_program program, another_program; @@ -2717,13 +2676,13 @@ int test_simple_library_with_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_minted_library = - clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, - NULL, NULL, &error); + clLinkProgram(context, 1, &device, "-create-library", 1, &program, NULL, + NULL, &error); test_error(error, "Unable to create a simple library"); error = create_single_kernel_helper_create_program( @@ -2736,14 +2695,14 @@ int test_simple_library_with_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(another_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program program_and_archive[2] = { another_program, my_newly_minted_library }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error); + context, 1, &device, "", 2, program_and_archive, NULL, NULL, &error); test_error(error, "Unable to create an executable from a binary and a library"); @@ -2763,10 +2722,7 @@ int test_simple_library_with_link(cl_device_id deviceID, cl_context context, return 0; } -int test_execute_after_simple_library_with_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_simple_library_with_link) { int error; cl_program program, another_program; @@ -2783,13 +2739,13 @@ int test_execute_after_simple_library_with_link(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program my_newly_minted_library = - clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, - NULL, NULL, &error); + clLinkProgram(context, 1, &device, "-create-library", 1, &program, NULL, + NULL, &error); test_error(error, "Unable to create a simple library"); error = create_single_kernel_helper_create_program( @@ -2802,14 +2758,14 @@ int test_execute_after_simple_library_with_link(cl_device_id deviceID, return -1; } - error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(another_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program program_and_archive[2] = { another_program, my_newly_minted_library }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error); + context, 1, &device, "", 2, program_and_archive, NULL, NULL, &error); test_error(error, "Unable to create an executable from a binary and a library"); @@ -2849,8 +2805,7 @@ int test_execute_after_simple_library_with_link(cl_device_id deviceID, return 0; } -int test_two_file_link(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(two_file_link) { int error; cl_program program, another_program; @@ -2866,8 +2821,8 @@ int test_two_file_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); @@ -2881,13 +2836,13 @@ int test_two_file_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(another_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program two_programs[2] = { program, another_program }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error); + context, 1, &device, "", 2, two_programs, NULL, NULL, &error); test_error(error, "Unable to create an executable from two binaries"); /* All done! */ @@ -2903,8 +2858,7 @@ int test_two_file_link(cl_device_id deviceID, cl_context context, return 0; } -int test_execute_after_two_file_link(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(execute_after_two_file_link) { int error; cl_program program, another_program; @@ -2921,8 +2875,8 @@ int test_execute_after_two_file_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); error = create_single_kernel_helper_create_program( @@ -2935,13 +2889,13 @@ int test_execute_after_two_file_link(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(another_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program two_programs[2] = { program, another_program }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error); + context, 1, &device, "", 2, two_programs, NULL, NULL, &error); test_error(error, "Unable to create an executable from two binaries"); cl_kernel kernel = @@ -2977,10 +2931,7 @@ int test_execute_after_two_file_link(cl_device_id deviceID, cl_context context, return 0; } -int test_execute_after_embedded_header_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_embedded_header_link) { int error; cl_program program, header, simple_program; @@ -3007,7 +2958,7 @@ int test_execute_after_embedded_header_link(cl_device_id deviceID, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 1, &header, + error = clCompileProgram(program, 1, &device, NULL, 1, &header, &simple_header_name, NULL, NULL); test_error(error, "Unable to compile a simple program with embedded header"); @@ -3022,13 +2973,13 @@ int test_execute_after_embedded_header_link(cl_device_id deviceID, return -1; } - error = clCompileProgram(simple_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(simple_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program two_programs[2] = { program, simple_program }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error); + context, 1, &device, "", 2, two_programs, NULL, NULL, &error); test_error(error, "Unable to create an executable from two binaries, one compiled " "with embedded header"); @@ -3078,10 +3029,7 @@ int test_execute_after_embedded_header_link(cl_device_id deviceID, #include #endif -int test_execute_after_included_header_link(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(execute_after_included_header_link) { int error; cl_program program, simple_program; @@ -3154,10 +3102,10 @@ int test_execute_after_included_header_link(cl_device_id deviceID, return -1; } #if (defined(__linux__) || defined(__APPLE__)) && (!defined(__ANDROID__)) - error = clCompileProgram(program, 1, &deviceID, "-I/tmp/foo/bar", 0, NULL, + error = clCompileProgram(program, 1, &device, "-I/tmp/foo/bar", 0, NULL, NULL, NULL, NULL); #else - error = clCompileProgram(program, 1, &deviceID, "-Ifoo/bar", 0, NULL, NULL, + error = clCompileProgram(program, 1, &device, "-Ifoo/bar", 0, NULL, NULL, NULL, NULL); #endif test_error(error, @@ -3206,13 +3154,13 @@ int test_execute_after_included_header_link(cl_device_id deviceID, return -1; } - error = clCompileProgram(simple_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(simple_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program two_programs[2] = { program, simple_program }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error); + context, 1, &device, "", 2, two_programs, NULL, NULL, &error); test_error(error, "Unable to create an executable from two binaries, one compiled " "with embedded header"); @@ -3250,8 +3198,7 @@ int test_execute_after_included_header_link(cl_device_id deviceID, return 0; } -int test_program_binary_type(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(program_binary_type) { int error; cl_program program, another_program, program_with_binary, @@ -3272,11 +3219,11 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, - NULL); + error = + clCompileProgram(program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); - error = clGetProgramBuildInfo(program, deviceID, CL_PROGRAM_BINARY_TYPE, + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); @@ -3335,12 +3282,12 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, // use clCreateProgramWithBinary program_with_binary = clCreateProgramWithBinary( - context, 1, &deviceID, &binarySize, (const unsigned char **)buffers, + context, 1, &device, &binarySize, (const unsigned char **)buffers, loadErrors, &error); test_error(error, "Unable to create program with binary"); error = clGetProgramBuildInfo( - program_with_binary, deviceID, CL_PROGRAM_BINARY_TYPE, + program_with_binary, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); if (program_type != CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT) @@ -3356,11 +3303,11 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, } cl_program my_newly_minted_library = - clLinkProgram(context, 1, &deviceID, "-create-library", 1, + clLinkProgram(context, 1, &device, "-create-library", 1, &program_with_binary, NULL, NULL, &error); test_error(error, "Unable to create a simple library"); error = clGetProgramBuildInfo( - my_newly_minted_library, deviceID, CL_PROGRAM_BINARY_TYPE, + my_newly_minted_library, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); if (program_type != CL_PROGRAM_BINARY_TYPE_LIBRARY) @@ -3417,11 +3364,11 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, // use clCreateProgramWithBinary cl_program library_with_binary = clCreateProgramWithBinary( - context, 1, &deviceID, &binarySize, (const unsigned char **)buffers, + context, 1, &device, &binarySize, (const unsigned char **)buffers, loadErrors, &error); test_error(error, "Unable to create program with binary"); error = clGetProgramBuildInfo( - library_with_binary, deviceID, CL_PROGRAM_BINARY_TYPE, + library_with_binary, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); if (program_type != CL_PROGRAM_BINARY_TYPE_LIBRARY) @@ -3444,19 +3391,19 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, return -1; } - error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, + error = clCompileProgram(another_program, 1, &device, NULL, 0, NULL, NULL, NULL, NULL); test_error(error, "Unable to compile a simple program"); cl_program program_and_archive[2] = { another_program, library_with_binary }; cl_program fully_linked_program = clLinkProgram( - context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error); + context, 1, &device, "", 2, program_and_archive, NULL, NULL, &error); test_error(error, "Unable to create an executable from a binary and a library"); error = clGetProgramBuildInfo( - fully_linked_program, deviceID, CL_PROGRAM_BINARY_TYPE, + fully_linked_program, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); if (program_type != CL_PROGRAM_BINARY_TYPE_EXECUTABLE) @@ -3515,12 +3462,12 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, // use clCreateProgramWithBinary fully_linked_program_with_binary = clCreateProgramWithBinary( - context, 1, &deviceID, &binarySize, (const unsigned char **)buffers, + context, 1, &device, &binarySize, (const unsigned char **)buffers, loadErrors, &error); test_error(error, "Unable to create program with binary"); error = clGetProgramBuildInfo( - fully_linked_program_with_binary, deviceID, CL_PROGRAM_BINARY_TYPE, + fully_linked_program_with_binary, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); if (program_type != CL_PROGRAM_BINARY_TYPE_EXECUTABLE) @@ -3535,7 +3482,7 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, free(binary); } - error = clBuildProgram(fully_linked_program_with_binary, 1, &deviceID, NULL, + error = clBuildProgram(fully_linked_program_with_binary, 1, &device, NULL, NULL, NULL); test_error(error, "Unable to build a simple program"); @@ -3575,9 +3522,9 @@ int test_program_binary_type(cl_device_id deviceID, cl_context context, return -1; } - error = clBuildProgram(program, 1, &deviceID, NULL, NULL, NULL); + error = clBuildProgram(program, 1, &device, NULL, NULL, NULL); test_error(error, "Unable to build a simple program"); - error = clGetProgramBuildInfo(program, deviceID, CL_PROGRAM_BINARY_TYPE, + error = clGetProgramBuildInfo(program, device, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL); test_error(error, "Unable to get program binary type"); @@ -3666,10 +3613,9 @@ void CL_CALLBACK test_notify_link_complete(cl_program program, void *userData) log_info("\n <-- program successfully linked\n"); } -int test_large_compile_and_link_status_options_log(cl_context context, - cl_device_id deviceID, - cl_command_queue queue, - unsigned int numLines) +static int test_large_compile_and_link_status_options_log( + cl_context context, cl_device_id deviceID, cl_command_queue queue, + unsigned int numLines) { int error; cl_program program; @@ -3987,10 +3933,7 @@ int test_large_compile_and_link_status_options_log(cl_context context, return 0; } -int test_compile_and_link_status_options_log(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(compile_and_link_status_options_log) { unsigned int toTest[] = { 256, 0 }; // 512, 1024, 8192, 16384, 32768, 0 }; unsigned int i; @@ -4009,7 +3952,7 @@ int test_compile_and_link_status_options_log(cl_device_id deviceID, gettimeofday(&time1, NULL); #endif - if (test_large_compile_and_link_status_options_log(context, deviceID, + if (test_large_compile_and_link_status_options_log(context, device, queue, toTest[i]) != 0) { diff --git a/test_conformance/compiler/test_compiler_defines_for_extensions.cpp b/test_conformance/compiler/test_compiler_defines_for_extensions.cpp index 0b5dd449..c5c62e17 100644 --- a/test_conformance/compiler/test_compiler_defines_for_extensions.cpp +++ b/test_conformance/compiler/test_compiler_defines_for_extensions.cpp @@ -131,7 +131,7 @@ bool string_has_prefix(const char *str, const char *prefix) return strncmp(str, prefix, strlen(prefix)) == 0; } -int test_compiler_defines_for_extensions(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(compiler_defines_for_extensions) { int error; diff --git a/test_conformance/compiler/test_feature_macro.cpp b/test_conformance/compiler/test_feature_macro.cpp index 2f973eec..93776a8b 100644 --- a/test_conformance/compiler/test_feature_macro.cpp +++ b/test_conformance/compiler/test_feature_macro.cpp @@ -205,10 +205,10 @@ int feature_macro_verify_results(std::string test_macro_name, return error; } -int test_feature_macro_atomic_order_acq_rel(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_atomic_order_acq_rel(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -233,10 +233,10 @@ int test_feature_macro_atomic_order_acq_rel(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_atomic_order_seq_cst(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_atomic_order_seq_cst(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -262,10 +262,10 @@ int test_feature_macro_atomic_order_seq_cst(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_atomic_scope_device(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_atomic_scope_device(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -289,10 +289,9 @@ int test_feature_macro_atomic_scope_device(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_atomic_scope_all_devices(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_atomic_scope_all_devices( + cl_device_id deviceID, cl_context context, std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -316,10 +315,10 @@ int test_feature_macro_atomic_scope_all_devices(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_3d_image_writes(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_3d_image_writes(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -343,9 +342,10 @@ int test_feature_macro_3d_image_writes(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_device_enqueue(cl_device_id deviceID, cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_device_enqueue(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -371,10 +371,10 @@ int test_feature_macro_device_enqueue(cl_device_id deviceID, cl_context context, compiler_status, supported); } -int test_feature_macro_generic_address_space(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_generic_address_space(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -398,8 +398,9 @@ int test_feature_macro_generic_address_space(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_pipes(cl_device_id deviceID, cl_context context, - std::string test_macro_name, cl_bool& supported) +static int test_feature_macro_pipes(cl_device_id deviceID, cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -423,7 +424,7 @@ int test_feature_macro_pipes(cl_device_id deviceID, cl_context context, compiler_status, supported); } -int test_feature_macro_program_scope_global_variables( +static int test_feature_macro_program_scope_global_variables( cl_device_id deviceID, cl_context context, std::string test_macro_name, cl_bool& supported) { @@ -449,10 +450,10 @@ int test_feature_macro_program_scope_global_variables( compiler_status, supported); } -int test_feature_macro_read_write_images(cl_device_id deviceID, - cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_read_write_images(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -476,9 +477,10 @@ int test_feature_macro_read_write_images(cl_device_id deviceID, compiler_status, supported); } -int test_feature_macro_subgroups(cl_device_id deviceID, cl_context context, - std::string test_macro_name, - cl_bool& supported) +static int test_feature_macro_subgroups(cl_device_id deviceID, + cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -502,7 +504,7 @@ int test_feature_macro_subgroups(cl_device_id deviceID, cl_context context, compiler_status, supported); } -int test_feature_macro_work_group_collective_functions( +static int test_feature_macro_work_group_collective_functions( cl_device_id deviceID, cl_context context, std::string test_macro_name, cl_bool& supported) { @@ -529,8 +531,9 @@ int test_feature_macro_work_group_collective_functions( compiler_status, supported); } -int test_feature_macro_images(cl_device_id deviceID, cl_context context, - std::string test_macro_name, cl_bool& supported) +static int test_feature_macro_images(cl_device_id deviceID, cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -554,8 +557,9 @@ int test_feature_macro_images(cl_device_id deviceID, cl_context context, compiler_status, supported); } -int test_feature_macro_fp64(cl_device_id deviceID, cl_context context, - std::string test_macro_name, cl_bool& supported) +static int test_feature_macro_fp64(cl_device_id deviceID, cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -580,7 +584,7 @@ int test_feature_macro_fp64(cl_device_id deviceID, cl_context context, compiler_status, supported); } -int test_feature_macro_integer_dot_product_input_4x8bit_packed( +static int test_feature_macro_integer_dot_product_input_4x8bit_packed( cl_device_id deviceID, cl_context context, std::string test_macro_name, cl_bool& supported) { @@ -616,7 +620,7 @@ int test_feature_macro_integer_dot_product_input_4x8bit_packed( compiler_status, supported); } -int test_feature_macro_integer_dot_product_input_4x8bit( +static int test_feature_macro_integer_dot_product_input_4x8bit( cl_device_id deviceID, cl_context context, std::string test_macro_name, cl_bool& supported) { @@ -652,8 +656,9 @@ int test_feature_macro_integer_dot_product_input_4x8bit( compiler_status, supported); } -int test_feature_macro_int64(cl_device_id deviceID, cl_context context, - std::string test_macro_name, cl_bool& supported) +static int test_feature_macro_int64(cl_device_id deviceID, cl_context context, + std::string test_macro_name, + cl_bool& supported) { cl_int error = TEST_FAIL; cl_bool api_status; @@ -722,8 +727,8 @@ int test_feature_macro_int64(cl_device_id deviceID, cl_context context, compiler_status, supported); } -int test_consistency_c_features_list(cl_device_id deviceID, - std::vector vec_to_cmp) +static int test_consistency_c_features_list(cl_device_id deviceID, + std::vector vec_to_cmp) { log_info("\nComparison list of features: CL_DEVICE_OPENCL_C_FEATURES vs " "API/compiler queries.\n"); @@ -792,13 +797,12 @@ int test_consistency_c_features_list(cl_device_id deviceID, #define NEW_FEATURE_MACRO_TEST(feat) \ test_macro_name = "__opencl_c_" #feat; \ - error |= test_feature_macro_##feat(deviceID, context, test_macro_name, \ + error |= test_feature_macro_##feat(device, context, test_macro_name, \ supported); \ if (supported) supported_features_vec.push_back(test_macro_name); -int test_features_macro(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(features_macro, Version(3, 0)) { // Note: Not checking that the feature array is empty for the compiler not @@ -806,7 +810,7 @@ int test_features_macro(cl_device_id deviceID, cl_context context, // support compilation from OpenCL C source, this query may return an empty // array." It "may" return an empty array implies that an implementation // also "may not". - check_compiler_available(deviceID); + check_compiler_available(device); int error = TEST_PASS; cl_bool supported = CL_FALSE; @@ -830,17 +834,16 @@ int test_features_macro(cl_device_id deviceID, cl_context context, NEW_FEATURE_MACRO_TEST(integer_dot_product_input_4x8bit); NEW_FEATURE_MACRO_TEST(integer_dot_product_input_4x8bit_packed); - error |= test_consistency_c_features_list(deviceID, supported_features_vec); + error |= test_consistency_c_features_list(device, supported_features_vec); return error; } // This test checks that a supported feature comes with other required features -int test_features_macro_coupling(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(features_macro_coupling) { OpenCLCFeatures features; - int error = get_device_cl_c_features(deviceID, features); + int error = get_device_cl_c_features(device, features); if (error) { log_error("Couldn't query OpenCL C features for the device!\n"); diff --git a/test_conformance/compiler/test_image_macro.cpp b/test_conformance/compiler/test_image_macro.cpp index 8a5c6707..6e9b0404 100644 --- a/test_conformance/compiler/test_image_macro.cpp +++ b/test_conformance/compiler/test_image_macro.cpp @@ -36,21 +36,22 @@ const char * image_not_supported_source = "kernel void not_enabled(global int * "\r\n } \r\n"; -int test_image_macro(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(image_macro) { cl_bool image_support; char buf[256]; int status; cl_program program; - status = clGetDeviceInfo( deviceID, CL_DEVICE_NAME, sizeof( buf ), buf, NULL ); + status = clGetDeviceInfo(device, CL_DEVICE_NAME, sizeof(buf), buf, NULL); if( status ) { log_error( "getting device info (name): %d\n", status ); exit(-1); } - status = clGetDeviceInfo( deviceID, CL_DEVICE_IMAGE_SUPPORT, sizeof( image_support ), &image_support, NULL ); + status = clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, + sizeof(image_support), &image_support, NULL); if( status ) { log_error( "getting device info (image support): %d\n", status ); @@ -67,7 +68,7 @@ int test_image_macro(cl_device_id deviceID, cl_context context, cl_command_queue return status; } - status = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + status = clBuildProgram(program, 1, &device, NULL, NULL, NULL); if( status ) log_error("CL_DEVICE_IMAGE_SUPPORT is set, __IMAGE_SUPPORT__ macro not set \n"); else @@ -82,7 +83,7 @@ int test_image_macro(cl_device_id deviceID, cl_context context, cl_command_queue return status; } - status = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL ); + status = clBuildProgram(program, 1, &device, NULL, NULL, NULL); if( status ) log_error("CL_DEVICE_IMAGE_SUPPORT not set, __IMAGE_SUPPORT__ macro is set \n"); else @@ -98,4 +99,3 @@ int test_image_macro(cl_device_id deviceID, cl_context context, cl_command_queue return status; } - diff --git a/test_conformance/compiler/test_opencl_c_versions.cpp b/test_conformance/compiler/test_opencl_c_versions.cpp index d750366e..5b24160f 100644 --- a/test_conformance/compiler/test_opencl_c_versions.cpp +++ b/test_conformance/compiler/test_opencl_c_versions.cpp @@ -286,8 +286,7 @@ static int test_CL_DEVICE_OPENCL_C_VERSION_versions(cl_device_id device, return TEST_PASS; } -int test_opencl_c_versions(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(opencl_c_versions) { check_compiler_available(device); diff --git a/test_conformance/compiler/test_pragma_unroll.cpp b/test_conformance/compiler/test_pragma_unroll.cpp index 682faa7f..8e90c22c 100644 --- a/test_conformance/compiler/test_pragma_unroll.cpp +++ b/test_conformance/compiler/test_pragma_unroll.cpp @@ -248,49 +248,60 @@ const char *pragma_unroll_kernels[] = { "}\n", }; -int test_pragma_unroll(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { - const size_t ELEMENT_NUM = 100; - const size_t KERNEL_NUM = 24; +REGISTER_TEST_VERSION(pragma_unroll, Version(2, 0)) +{ + const size_t ELEMENT_NUM = 100; + const size_t KERNEL_NUM = 24; - cl_int error; + cl_int error; - //execute all kernels and check if the results are as expected - for (size_t kernelIdx = 0; kernelIdx < KERNEL_NUM; ++kernelIdx) { - clProgramWrapper program; - clKernelWrapper kernel; - if (create_single_kernel_helper( - context, &program, &kernel, 1, - (const char **)&pragma_unroll_kernels[kernelIdx], "pragma_unroll")) + // execute all kernels and check if the results are as expected + for (size_t kernelIdx = 0; kernelIdx < KERNEL_NUM; ++kernelIdx) { - log_error("The program we attempted to compile was: \n%s\n", - pragma_unroll_kernels[kernelIdx]); - return -1; + clProgramWrapper program; + clKernelWrapper kernel; + if (create_single_kernel_helper( + context, &program, &kernel, 1, + (const char **)&pragma_unroll_kernels[kernelIdx], + "pragma_unroll")) + { + log_error("The program we attempted to compile was: \n%s\n", + pragma_unroll_kernels[kernelIdx]); + return -1; + } + + clMemWrapper buffer = + clCreateBuffer(context, CL_MEM_READ_WRITE, + ELEMENT_NUM * sizeof(cl_uint), NULL, &error); + test_error(error, "clCreateBuffer failed"); + + error = clSetKernelArg(kernel, 0, sizeof(buffer), &buffer); + test_error(error, "clSetKernelArg failed"); + + // only one thread should be enough to verify if kernel is fully + // functional + size_t workSize = 1; + error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &workSize, NULL, + 0, NULL, NULL); + test_error(error, "clEnqueueNDRangeKernel failed"); + + std::vector results(ELEMENT_NUM, 0); + error = clEnqueueReadBuffer(queue, buffer, CL_TRUE, 0, + ELEMENT_NUM * sizeof(cl_uint), &results[0], + 0, NULL, NULL); + test_error(error, "clEnqueueReadBuffer failed"); + + for (size_t i = 0; i < ELEMENT_NUM; ++i) + { + if (results[i] != i) + { + log_error("Kernel %zu returned invalid result. Test: %d, " + "expected: %zu\n", + kernelIdx + 1, results[i], i); + return -1; + } + } } - clMemWrapper buffer = clCreateBuffer(context, CL_MEM_READ_WRITE, ELEMENT_NUM * sizeof(cl_uint), NULL, &error); - test_error(error, "clCreateBuffer failed"); - - error = clSetKernelArg(kernel, 0, sizeof(buffer), &buffer); - test_error(error, "clSetKernelArg failed"); - - //only one thread should be enough to verify if kernel is fully functional - size_t workSize = 1; - error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &workSize, NULL, 0, NULL, NULL); - test_error(error, "clEnqueueNDRangeKernel failed"); - - std::vector results(ELEMENT_NUM, 0); - error = clEnqueueReadBuffer(queue, buffer, CL_TRUE, 0, ELEMENT_NUM * sizeof(cl_uint), &results[0], 0, NULL, NULL); - test_error(error, "clEnqueueReadBuffer failed"); - - for (size_t i = 0; i < ELEMENT_NUM; ++i) { - if (results[i] != i) { - log_error( - "Kernel %zu returned invalid result. Test: %d, expected: %zu\n", - kernelIdx + 1, results[i], i); - return -1; - } - } - } - - return 0; + return 0; } diff --git a/test_conformance/compiler/test_preprocessor.cpp b/test_conformance/compiler/test_preprocessor.cpp index 3c5c2045..12891b2b 100644 --- a/test_conformance/compiler/test_preprocessor.cpp +++ b/test_conformance/compiler/test_preprocessor.cpp @@ -37,85 +37,100 @@ const char *define_kernel_code[] = { "}\n"}; +REGISTER_TEST(preprocessor_define_udef) +{ + cl_int error; + clKernelWrapper kernel; + clProgramWrapper program; + clMemWrapper buffer[3]; + cl_int *srcData, *resultData; + int i; + MTdata d; + + error = create_single_kernel_helper(context, &program, &kernel, 1, + define_kernel_code, "define_test"); + if (error) return -1; + + buffer[0] = clCreateBuffer(context, CL_MEM_READ_WRITE, + num_elements * sizeof(cl_int), NULL, &error); + test_error(error, "clCreateBuffer failed"); + buffer[1] = clCreateBuffer(context, CL_MEM_READ_WRITE, + num_elements * sizeof(cl_int), NULL, &error); + test_error(error, "clCreateBuffer failed"); + buffer[2] = clCreateBuffer(context, CL_MEM_READ_WRITE, + num_elements * sizeof(cl_int), NULL, &error); + test_error(error, "clCreateBuffer failed"); + + srcData = (cl_int *)malloc(sizeof(cl_int) * num_elements); + if (srcData == NULL) + { + log_error("Failed to allocate storage for source data (%d cl_ints).\n", + num_elements); + return -1; + } + + d = init_genrand(gRandomSeed); + for (i = 0; i < num_elements; i++) + srcData[i] = (int)get_random_float(-1024, 1024, d); + free_mtdata(d); + d = NULL; + + resultData = (cl_int *)malloc(sizeof(cl_int) * num_elements); + if (resultData == NULL) + { + free(srcData); + log_error("Failed to allocate storage for result data (%d cl_ints).\n", + num_elements); + return -1; + } + + error = clSetKernelArg(kernel, 0, sizeof(buffer[0]), &buffer[0]); + test_error(error, "clSetKernelArg failed"); + error = clSetKernelArg(kernel, 1, sizeof(buffer[1]), &buffer[1]); + test_error(error, "clSetKernelArg failed"); + error = clSetKernelArg(kernel, 2, sizeof(buffer[2]), &buffer[2]); + test_error(error, "clSetKernelArg failed"); -int test_preprocessor_define_udef(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { + error = clEnqueueWriteBuffer(queue, buffer[0], CL_TRUE, 0, + num_elements * sizeof(cl_int), srcData, 0, + NULL, NULL); + test_error(error, "clEnqueueWriteBuffer failed"); - cl_int error; - clKernelWrapper kernel; - clProgramWrapper program; - clMemWrapper buffer[3]; - cl_int *srcData, *resultData; - int i; - MTdata d; + size_t threads[3] = { (size_t)num_elements, 0, 0 }; + error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, threads, NULL, 0, + NULL, NULL); + test_error(error, "clEnqueueNDRangeKernel failed"); - error = create_single_kernel_helper(context, &program, &kernel, 1, define_kernel_code, "define_test"); - if (error) - return -1; + error = clEnqueueReadBuffer(queue, buffer[1], CL_TRUE, 0, + num_elements * sizeof(cl_int), resultData, 0, + NULL, NULL); + test_error(error, "clEnqueueReadBuffer failed"); - buffer[0] = clCreateBuffer(context, CL_MEM_READ_WRITE, num_elements*sizeof(cl_int), NULL, &error); - test_error( error, "clCreateBuffer failed"); - buffer[1] = clCreateBuffer(context, CL_MEM_READ_WRITE, num_elements*sizeof(cl_int), NULL, &error); - test_error( error, "clCreateBuffer failed"); - buffer[2] = clCreateBuffer(context, CL_MEM_READ_WRITE, num_elements*sizeof(cl_int), NULL, &error); - test_error( error, "clCreateBuffer failed"); + for (i = 0; i < num_elements; i++) + if (resultData[i] != srcData[i] * 2) + { + free(srcData); + free(resultData); + return -1; + } - srcData = (cl_int*)malloc(sizeof(cl_int)*num_elements); - if (srcData == NULL) { - log_error("Failed to allocate storage for source data (%d cl_ints).\n", num_elements); - return -1; - } + error = clEnqueueReadBuffer(queue, buffer[2], CL_TRUE, 0, + num_elements * sizeof(cl_int), resultData, 0, + NULL, NULL); + test_error(error, "clEnqueueReadBuffer failed"); - d = init_genrand( gRandomSeed ); - for (i=0; i Date: Tue, 18 Mar 2025 17:11:43 +0000 Subject: [PATCH 06/11] Migrate generic_address_space suite to the new test registration framework (#2333) Contributes to https://github.com/KhronosGroup/OpenCL-CTS/issues/2181. Signed-off-by: Michael Rizkalla --- .../generic_address_space/advanced_tests.cpp | 67 ++++++++++++------- .../generic_address_space/atomic_tests.cpp | 14 ++-- .../generic_address_space/basic_tests.cpp | 60 ++++++++++------- .../generic_address_space/main.cpp | 66 ++---------------- .../generic_address_space/stress_tests.cpp | 13 ++-- 5 files changed, 97 insertions(+), 123 deletions(-) diff --git a/test_conformance/generic_address_space/advanced_tests.cpp b/test_conformance/generic_address_space/advanced_tests.cpp index 74330e52..b59e0892 100644 --- a/test_conformance/generic_address_space/advanced_tests.cpp +++ b/test_conformance/generic_address_space/advanced_tests.cpp @@ -293,7 +293,8 @@ private: const ExtraKernelArgMemType _extraKernelArgMemType; }; -int test_library_function(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(library_function) +{ const std::string LIBRARY_FUNCTION = common::CONFORMANCE_VERIFY_FENCE + NL NL "bool helperFunction(float *floatp, float val) {" @@ -340,10 +341,11 @@ __kernel void testKernel(__global uint *results) { CAdvancedTest test(LIBRARY_FUNCTION, KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_generic_variable_volatile(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(generic_variable_volatile) +{ std::vector KERNEL_FUNCTIONS; KERNEL_FUNCTIONS.push_back(common::CONFORMANCE_VERIFY_FENCE + @@ -420,10 +422,11 @@ int test_generic_variable_volatile(cl_device_id deviceID, cl_context context, cl CAdvancedTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_generic_variable_const(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(generic_variable_const) +{ std::vector KERNEL_FUNCTIONS; KERNEL_FUNCTIONS.push_back(common::CONFORMANCE_VERIFY_FENCE + @@ -474,10 +477,11 @@ int test_generic_variable_const(cl_device_id deviceID, cl_context context, cl_co CAdvancedTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_generic_variable_gentype(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(generic_variable_gentype) +{ const std::string KERNEL_FUNCTION_TEMPLATE = common::CONFORMANCE_VERIFY_FENCE + NL NL "%s" @@ -531,7 +535,8 @@ int test_generic_variable_gentype(cl_device_id deviceID, cl_context context, cl_ const std::string cl_khr_fp64_pragma = "#pragma OPENCL EXTENSION cl_khr_fp64 : enable"; // Add double floating types if they are supported - if (is_extension_available(deviceID, "cl_khr_fp64")) { + if (is_extension_available(device, "cl_khr_fp64")) + { for (size_t j = 0; j < sizeof(vector_sizes) / sizeof(vector_sizes[0]); j++) { for (size_t k = 0; k < sizeof(address_spaces) / sizeof(address_spaces[0]); k++) { char temp_kernel[1024]; @@ -551,7 +556,8 @@ int test_generic_variable_gentype(cl_device_id deviceID, cl_context context, cl_ const std::string cl_khr_fp16_pragma = "#pragma OPENCL EXTENSION cl_khr_fp16 : enable"; // Add half floating types if they are supported - if (is_extension_available(deviceID, "cl_khr_fp16")) { + if (is_extension_available(device, "cl_khr_fp16")) + { for (size_t j = 0; j < sizeof(vector_sizes) / sizeof(vector_sizes[0]); j++) { for (size_t k = 0; k < sizeof(address_spaces) / sizeof(address_spaces[0]); k++) { char temp_kernel[1024]; @@ -586,7 +592,7 @@ int test_generic_variable_gentype(cl_device_id deviceID, cl_context context, cl_ CAdvancedTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } void create_math_kernels(std::vector& KERNEL_FUNCTIONS) { @@ -919,19 +925,21 @@ void create_vstore_kernels(std::vector& KERNEL_FUNCTIONS, cl_device } } -int test_builtin_functions(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(builtin_functions) +{ std::vector KERNEL_FUNCTIONS; create_math_kernels(KERNEL_FUNCTIONS); - create_vload_kernels(KERNEL_FUNCTIONS, deviceID); - create_vstore_kernels(KERNEL_FUNCTIONS, deviceID); + create_vload_kernels(KERNEL_FUNCTIONS, device); + create_vstore_kernels(KERNEL_FUNCTIONS, device); CAdvancedTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_generic_advanced_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(generic_advanced_casting) +{ std::vector KERNEL_FUNCTIONS; KERNEL_FUNCTIONS.push_back( @@ -980,18 +988,20 @@ int test_generic_advanced_casting(cl_device_id deviceID, cl_context context, cl_ CAdvancedTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_generic_ptr_to_host_mem_svm(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(generic_ptr_to_host_mem_svm) +{ cl_int result = CL_SUCCESS; /* Test SVM capabilities and select matching tests */ cl_device_svm_capabilities caps; - auto version = get_device_cl_version(deviceID); + auto version = get_device_cl_version(device); auto expected_min_version = Version(2, 0); - cl_int error = clGetDeviceInfo(deviceID, CL_DEVICE_SVM_CAPABILITIES, sizeof(caps), &caps, NULL); + cl_int error = clGetDeviceInfo(device, CL_DEVICE_SVM_CAPABILITIES, + sizeof(caps), &caps, NULL); test_error(error, "clGetDeviceInfo(CL_DEVICE_SVM_CAPABILITIES) failed"); if ((version < expected_min_version) @@ -1000,35 +1010,40 @@ int test_generic_ptr_to_host_mem_svm(cl_device_id deviceID, cl_context context, if (caps & CL_DEVICE_SVM_COARSE_GRAIN_BUFFER) { CAdvancedTest test_global_svm_ptr(common::GLOBAL_KERNEL_FUNCTION, ARG_TYPE_COARSE_GRAINED_SVM); - result |= test_global_svm_ptr.Execute(deviceID, context, queue, num_elements); + result |= + test_global_svm_ptr.Execute(device, context, queue, num_elements); } if (caps & CL_DEVICE_SVM_FINE_GRAIN_BUFFER) { CAdvancedTest test_global_svm_ptr(common::GLOBAL_KERNEL_FUNCTION, ARG_TYPE_FINE_GRAINED_BUFFER_SVM); - result |= test_global_svm_ptr.Execute(deviceID, context, queue, num_elements); + result |= + test_global_svm_ptr.Execute(device, context, queue, num_elements); } if (caps & CL_DEVICE_SVM_FINE_GRAIN_SYSTEM) { CAdvancedTest test_global_svm_ptr(common::GLOBAL_KERNEL_FUNCTION, ARG_TYPE_FINE_GRAINED_SYSTEM_SVM); - result |= test_global_svm_ptr.Execute(deviceID, context, queue, num_elements); + result |= + test_global_svm_ptr.Execute(device, context, queue, num_elements); } if (caps & CL_DEVICE_SVM_ATOMICS) { CAdvancedTest test_global_svm_ptr(common::GLOBAL_KERNEL_FUNCTION, ARG_TYPE_ATOMICS_SVM); - result |= test_global_svm_ptr.Execute(deviceID, context, queue, num_elements); + result |= + test_global_svm_ptr.Execute(device, context, queue, num_elements); } return result; } -int test_generic_ptr_to_host_mem(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(generic_ptr_to_host_mem) +{ cl_int result = CL_SUCCESS; CAdvancedTest test_global_ptr(common::GLOBAL_KERNEL_FUNCTION, ARG_TYPE_HOST_PTR); - result |= test_global_ptr.Execute(deviceID, context, queue, num_elements); + result |= test_global_ptr.Execute(device, context, queue, num_elements); CAdvancedTest test_local_ptr(common::LOCAL_KERNEL_FUNCTION, ARG_TYPE_HOST_LOCAL); - result |= test_local_ptr.Execute(deviceID, context, queue, num_elements / 64); + result |= test_local_ptr.Execute(device, context, queue, num_elements / 64); return result; } diff --git a/test_conformance/generic_address_space/atomic_tests.cpp b/test_conformance/generic_address_space/atomic_tests.cpp index a24c6ae2..8a568f0a 100644 --- a/test_conformance/generic_address_space/atomic_tests.cpp +++ b/test_conformance/generic_address_space/atomic_tests.cpp @@ -91,10 +91,9 @@ kernel void testKernel(global atomic_int* globalPtr, local atomic_int* localPtr) )OpenCLC"; } -int test_generic_atomics_invariant(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int) +REGISTER_TEST(generic_atomics_invariant) { - const auto version = get_device_cl_version(deviceID); + const auto version = get_device_cl_version(device); if (version < Version(2, 0)) return TEST_SKIPPED_ITSELF; @@ -108,7 +107,7 @@ int test_generic_atomics_invariant(cl_device_id deviceID, cl_context context, size_t wgSize, retSize; // Attempt to find the simd unit size for the device. - err = clGetKernelWorkGroupInfo(kernel, deviceID, + err = clGetKernelWorkGroupInfo(kernel, device, CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE, sizeof(wgSize), &wgSize, &retSize); test_error(err, "clGetKernelWorkGroupInfo failed"); @@ -154,10 +153,9 @@ int test_generic_atomics_invariant(cl_device_id deviceID, cl_context context, return CL_SUCCESS; } -int test_generic_atomics_variant(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int) +REGISTER_TEST(generic_atomics_variant) { - const auto version = get_device_cl_version(deviceID); + const auto version = get_device_cl_version(device); if (version < Version(2, 0)) return TEST_SKIPPED_ITSELF; @@ -171,7 +169,7 @@ int test_generic_atomics_variant(cl_device_id deviceID, cl_context context, size_t wgSize, retSize; // Attempt to find the simd unit size for the device. - err = clGetKernelWorkGroupInfo(kernel, deviceID, + err = clGetKernelWorkGroupInfo(kernel, device, CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE, sizeof(wgSize), &wgSize, &retSize); test_error(err, "clGetKernelWorkGroupInfo failed"); diff --git a/test_conformance/generic_address_space/basic_tests.cpp b/test_conformance/generic_address_space/basic_tests.cpp index 114fafa0..8e12e6ff 100644 --- a/test_conformance/generic_address_space/basic_tests.cpp +++ b/test_conformance/generic_address_space/basic_tests.cpp @@ -97,7 +97,8 @@ private: const std::vector _kernels; }; -int test_function_get_fence(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(function_get_fence) +{ const std::string KERNEL_FUNCTION = common::CONFORMANCE_VERIFY_FENCE + NL NL "__global int gint = 1;" @@ -142,10 +143,11 @@ int test_function_get_fence(cl_device_id deviceID, cl_context context, cl_comman CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_function_to_address_space(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(function_to_address_space) +{ const std::string KERNEL_FUNCTION = NL NL "__global int gint = 1;" @@ -190,10 +192,11 @@ int test_function_to_address_space(cl_device_id deviceID, cl_context context, cl CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_variable_get_fence(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(variable_get_fence) +{ const std::string KERNEL_FUNCTION = common::CONFORMANCE_VERIFY_FENCE + NL NL "__global int gint = 1;" @@ -223,10 +226,11 @@ int test_variable_get_fence(cl_device_id deviceID, cl_context context, cl_comman CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_variable_to_address_space(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(variable_to_address_space) +{ const std::string KERNEL_FUNCTION = NL NL "__global int gint = 1;" @@ -256,10 +260,11 @@ int test_variable_to_address_space(cl_device_id deviceID, cl_context context, cl CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(casting) +{ std::vector KERNEL_FUNCTIONS; // pointers to global, local or private are implicitly convertible to generic @@ -345,10 +350,11 @@ int test_casting(cl_device_id deviceID, cl_context context, cl_command_queue que CBasicTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_conditional_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(conditional_casting) +{ const std::string KERNEL_FUNCTION = common::CONFORMANCE_VERIFY_FENCE + NL NL "__global int gint = 1;" @@ -376,10 +382,11 @@ int test_conditional_casting(cl_device_id deviceID, cl_context context, cl_comma CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_chain_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(chain_casting) +{ const std::string KERNEL_FUNCTION = common::CONFORMANCE_VERIFY_FENCE + NL NL "__global int gint = 1;" @@ -407,10 +414,11 @@ int test_chain_casting(cl_device_id deviceID, cl_context context, cl_command_que NL; CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_ternary_operator_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(ternary_operator_casting) +{ const std::string KERNEL_FUNCTION = common::CONFORMANCE_VERIFY_FENCE + NL NL "__global int gint = 1;" @@ -435,10 +443,11 @@ int test_ternary_operator_casting(cl_device_id deviceID, cl_context context, cl_ CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_language_struct(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(language_struct) +{ std::vector KERNEL_FUNCTIONS; // implicit private struct @@ -588,10 +597,11 @@ int test_language_struct(cl_device_id deviceID, cl_context context, cl_command_q CBasicTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_language_union(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(language_union) +{ std::vector KERNEL_FUNCTIONS; KERNEL_FUNCTIONS.push_back(common::CONFORMANCE_VERIFY_FENCE + @@ -726,10 +736,11 @@ int test_language_union(cl_device_id deviceID, cl_context context, cl_command_qu CBasicTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_multiple_calls_same_function(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(multiple_calls_same_function) +{ const std::string KERNEL_FUNCTION = NL NL "int shift2(const int *ptr, int arg) {" @@ -759,10 +770,11 @@ int test_multiple_calls_same_function(cl_device_id deviceID, cl_context context, CBasicTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } -int test_compare_pointers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(compare_pointers) +{ std::vector KERNEL_FUNCTIONS; KERNEL_FUNCTIONS.push_back( @@ -884,5 +896,5 @@ int test_compare_pointers(cl_device_id deviceID, cl_context context, cl_command_ CBasicTest test(KERNEL_FUNCTIONS); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } diff --git a/test_conformance/generic_address_space/main.cpp b/test_conformance/generic_address_space/main.cpp index 4d4dea46..cb9de902 100644 --- a/test_conformance/generic_address_space/main.cpp +++ b/test_conformance/generic_address_space/main.cpp @@ -18,66 +18,8 @@ #include // basic tests -extern int test_function_get_fence(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_function_to_address_space(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_variable_get_fence(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_variable_to_address_space(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_conditional_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_chain_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_ternary_operator_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_language_struct(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_language_union(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_multiple_calls_same_function(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_compare_pointers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -// advanced tests -extern int test_library_function(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_generic_variable_volatile(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_generic_variable_const(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_generic_variable_gentype(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_builtin_functions(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_generic_advanced_casting(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_generic_ptr_to_host_mem(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_generic_ptr_to_host_mem_svm(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_max_number_of_params(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -// atomic tests -int test_generic_atomics_invariant(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -int test_generic_atomics_variant(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - -test_definition test_list[] = { - // basic tests - ADD_TEST(function_get_fence), - ADD_TEST(function_to_address_space), - ADD_TEST(variable_get_fence), - ADD_TEST(variable_to_address_space), - ADD_TEST(casting), - ADD_TEST(conditional_casting), - ADD_TEST(chain_casting), - ADD_TEST(ternary_operator_casting), - ADD_TEST(language_struct), - ADD_TEST(language_union), - ADD_TEST(multiple_calls_same_function), - ADD_TEST(compare_pointers), - // advanced tests - ADD_TEST(library_function), - ADD_TEST(generic_variable_volatile), - ADD_TEST(generic_variable_const), - ADD_TEST(generic_variable_gentype), - ADD_TEST(builtin_functions), - ADD_TEST(generic_advanced_casting), - ADD_TEST(generic_ptr_to_host_mem), - ADD_TEST(generic_ptr_to_host_mem_svm), - ADD_TEST(max_number_of_params), - // atomic tests - ADD_TEST(generic_atomics_invariant), - ADD_TEST(generic_atomics_variant), -}; - -const int test_num = ARRAY_SIZE( test_list ); - -test_status InitCL(cl_device_id device) { +test_status InitCL(cl_device_id device) +{ auto version = get_device_cl_version(device); auto expected_min_version = Version(2, 0); @@ -134,5 +76,7 @@ test_status InitCL(cl_device_id device) { int main(int argc, const char *argv[]) { - return runTestHarnessWithCheck(argc, argv, test_num, test_list, false, false, InitCL); + return runTestHarnessWithCheck( + argc, argv, test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), false, false, InitCL); } diff --git a/test_conformance/generic_address_space/stress_tests.cpp b/test_conformance/generic_address_space/stress_tests.cpp index a2143229..374d3cab 100644 --- a/test_conformance/generic_address_space/stress_tests.cpp +++ b/test_conformance/generic_address_space/stress_tests.cpp @@ -99,15 +99,20 @@ private: const std::vector _kernels; }; -int test_max_number_of_params(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) { +REGISTER_TEST(max_number_of_params) +{ cl_int error; size_t deviceMaxParameterSize; - error = clGetDeviceInfo(deviceID, CL_DEVICE_MAX_PARAMETER_SIZE, sizeof(deviceMaxParameterSize), &deviceMaxParameterSize, NULL); + error = clGetDeviceInfo(device, CL_DEVICE_MAX_PARAMETER_SIZE, + sizeof(deviceMaxParameterSize), + &deviceMaxParameterSize, NULL); test_error(error, "clGetDeviceInfo failed"); size_t deviceAddressBits; - error = clGetDeviceInfo(deviceID, CL_DEVICE_ADDRESS_BITS, sizeof(deviceAddressBits), &deviceAddressBits, NULL); + error = + clGetDeviceInfo(device, CL_DEVICE_ADDRESS_BITS, + sizeof(deviceAddressBits), &deviceAddressBits, NULL); test_error(error, "clGetDeviceInfo failed"); size_t maxParams = deviceMaxParameterSize / (deviceAddressBits / 8); @@ -174,5 +179,5 @@ int test_max_number_of_params(cl_device_id deviceID, cl_context context, cl_comm CStressTest test(KERNEL_FUNCTION); - return test.Execute(deviceID, context, queue, num_elements); + return test.Execute(device, context, queue, num_elements); } From ac38405a625f79b4590a02498bf3dcdfd0886964 Mon Sep 17 00:00:00 2001 From: Michael Rizkalla Date: Tue, 18 Mar 2025 17:13:35 +0000 Subject: [PATCH 07/11] Migrate geometrics suite to the new test registration framework (#2335) Contributes to https://github.com/KhronosGroup/OpenCL-CTS/issues/2181. Signed-off-by: Michael Rizkalla --- test_conformance/geometrics/main.cpp | 17 +------ test_conformance/geometrics/procs.h | 35 ------------- test_conformance/geometrics/testBase.h | 18 ++++++- .../geometrics/test_geometrics.cpp | 49 ++++++++++--------- 4 files changed, 46 insertions(+), 73 deletions(-) delete mode 100644 test_conformance/geometrics/procs.h diff --git a/test_conformance/geometrics/main.cpp b/test_conformance/geometrics/main.cpp index 45f2b069..f9d6d5e6 100644 --- a/test_conformance/geometrics/main.cpp +++ b/test_conformance/geometrics/main.cpp @@ -17,27 +17,14 @@ #include #include -#include "procs.h" #include "harness/testHarness.h" #if !defined(_WIN32) #include #endif -test_definition test_list[] = { - ADD_TEST( geom_cross ), - ADD_TEST( geom_dot ), - ADD_TEST( geom_distance ), - ADD_TEST( geom_fast_distance ), - ADD_TEST( geom_length ), - ADD_TEST( geom_fast_length ), - ADD_TEST( geom_normalize ), - ADD_TEST( geom_fast_normalize ), -}; - -const int test_num = ARRAY_SIZE( test_list ); - int main(int argc, const char *argv[]) { - return runTestHarness(argc, argv, test_num, test_list, false, 0); + return runTestHarness(argc, argv, test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), false, 0); } diff --git a/test_conformance/geometrics/procs.h b/test_conformance/geometrics/procs.h deleted file mode 100644 index f9a96ff8..00000000 --- a/test_conformance/geometrics/procs.h +++ /dev/null @@ -1,35 +0,0 @@ -// -// Copyright (c) 2017 The Khronos Group Inc. -// -// Licensed under the Apache License, Version 2.0 (the "License"); -// you may not use this file except in compliance with the License. -// You may obtain a copy of the License at -// -// http://www.apache.org/licenses/LICENSE-2.0 -// -// Unless required by applicable law or agreed to in writing, software -// distributed under the License is distributed on an "AS IS" BASIS, -// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -// See the License for the specific language governing permissions and -// limitations under the License. -// -#include "harness/errorHelpers.h" -#include "harness/kernelHelpers.h" -#include "harness/typeWrappers.h" - -extern int create_program_and_kernel(const char *source, const char *kernel_name, cl_program *program_ret, cl_kernel *kernel_ret); - -extern int test_geom_cross(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_dot(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_distance(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_fast_distance(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_length(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_fast_length(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_normalize(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_geom_fast_normalize(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_geom_cross_double(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements, MTdata d); -extern int test_geom_dot_double(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements, MTdata d); -extern int test_geom_distance_double(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements, MTdata d); -extern int test_geom_length_double(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements, MTdata d); -extern int test_geom_normalize_double(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements, MTdata d); diff --git a/test_conformance/geometrics/testBase.h b/test_conformance/geometrics/testBase.h index 5b49bfd7..dc602a17 100644 --- a/test_conformance/geometrics/testBase.h +++ b/test_conformance/geometrics/testBase.h @@ -16,14 +16,30 @@ #ifndef _testBase_h #define _testBase_h +#include #include "harness/compat.h" +#include "harness/mt19937.h" #include #include #include #include -#include "procs.h" +extern int test_geom_cross_double(cl_device_id deviceID, cl_context context, + cl_command_queue queue, int num_elements, + MTdata d); +extern int test_geom_dot_double(cl_device_id deviceID, cl_context context, + cl_command_queue queue, int num_elements, + MTdata d); +extern int test_geom_distance_double(cl_device_id deviceID, cl_context context, + cl_command_queue queue, int num_elements, + MTdata d); +extern int test_geom_length_double(cl_device_id deviceID, cl_context context, + cl_command_queue queue, int num_elements, + MTdata d); +extern int test_geom_normalize_double(cl_device_id deviceID, cl_context context, + cl_command_queue queue, int num_elements, + MTdata d); #endif // _testBase_h diff --git a/test_conformance/geometrics/test_geometrics.cpp b/test_conformance/geometrics/test_geometrics.cpp index e305026c..2706dc3f 100644 --- a/test_conformance/geometrics/test_geometrics.cpp +++ b/test_conformance/geometrics/test_geometrics.cpp @@ -148,15 +148,13 @@ void cross_product( const float *vecA, const float *vecB, float *outVector, floa } - - -int test_geom_cross(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements ) +REGISTER_TEST(geom_cross) { int vecsize; RandomSeed seed(gRandomSeed); /* Get the default rounding mode */ - cl_device_fp_config defaultRoundingMode = get_default_rounding_mode(deviceID); + cl_device_fp_config defaultRoundingMode = get_default_rounding_mode(device); if( 0 == defaultRoundingMode ) return -1; @@ -269,12 +267,16 @@ int test_geom_cross(cl_device_id deviceID, cl_context context, cl_command_queue } } // for(vecsize=... - if(!is_extension_available(deviceID, "cl_khr_fp64")) { + if (!is_extension_available(device, "cl_khr_fp64")) + { log_info("Extension cl_khr_fp64 not supported; skipping double tests.\n"); return 0; - } else { + } + else + { log_info("Testing doubles...\n"); - return test_geom_cross_double( deviceID, context, queue, num_elements, seed); + return test_geom_cross_double(device, context, queue, num_elements, + seed); } } @@ -497,7 +499,7 @@ double verifyDot( float *srcA, float *srcB, size_t vecSize ) return total; } -int test_geom_dot(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_dot) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; @@ -516,14 +518,14 @@ int test_geom_dot(cl_device_id deviceID, cl_context context, cl_command_queue qu if (retVal) return retVal; - if(!is_extension_available(deviceID, "cl_khr_fp64")) + if (!is_extension_available(device, "cl_khr_fp64")) { log_info("Extension cl_khr_fp64 not supported; skipping double tests.\n"); return 0; } log_info("Testing doubles...\n"); - return test_geom_dot_double( deviceID, context, queue, num_elements, seed); + return test_geom_dot_double(device, context, queue, num_elements, seed); } double verifyFastDistance( float *srcA, float *srcB, size_t vecSize ) @@ -542,7 +544,7 @@ double verifyFastDistance( float *srcA, float *srcB, size_t vecSize ) return sqrt( total ); } -int test_geom_fast_distance(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_fast_distance) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; @@ -589,7 +591,7 @@ double verifyDistance( float *srcA, float *srcB, size_t vecSize ) return sqrt( total ); } -int test_geom_distance(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_distance) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; @@ -616,13 +618,14 @@ int test_geom_distance(cl_device_id deviceID, cl_context context, cl_command_que if (retVal) return retVal; - if(!is_extension_available(deviceID, "cl_khr_fp64")) + if (!is_extension_available(device, "cl_khr_fp64")) { log_info("Extension cl_khr_fp64 not supported; skipping double tests.\n"); return 0; } else { log_info("Testing doubles...\n"); - return test_geom_distance_double( deviceID, context, queue, num_elements, seed); + return test_geom_distance_double(device, context, queue, num_elements, + seed); } } @@ -754,7 +757,7 @@ double verifyLength( float *srcA, size_t vecSize ) return sqrt( total ); } -int test_geom_length(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_length) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; @@ -781,7 +784,7 @@ int test_geom_length(cl_device_id deviceID, cl_context context, cl_command_queue if (retVal) return retVal; - if(!is_extension_available(deviceID, "cl_khr_fp64")) + if (!is_extension_available(device, "cl_khr_fp64")) { log_info("Extension cl_khr_fp64 not supported; skipping double tests.\n"); return 0; @@ -789,7 +792,8 @@ int test_geom_length(cl_device_id deviceID, cl_context context, cl_command_queue else { log_info("Testing doubles...\n"); - return test_geom_length_double( deviceID, context, queue, num_elements, seed); + return test_geom_length_double(device, context, queue, num_elements, + seed); } } @@ -809,7 +813,7 @@ double verifyFastLength( float *srcA, size_t vecSize ) return sqrt( total ); } -int test_geom_fast_length(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_fast_length) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; @@ -1059,7 +1063,7 @@ void verifyNormalize( float *srcA, float *dst, size_t vecSize ) dst[i] = (float)( (double)srcA[i] / value ); } -int test_geom_normalize(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_normalize) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; @@ -1084,18 +1088,19 @@ int test_geom_normalize(cl_device_id deviceID, cl_context context, cl_command_qu if (retVal) return retVal; - if(!is_extension_available(deviceID, "cl_khr_fp64")) + if (!is_extension_available(device, "cl_khr_fp64")) { log_info("Extension cl_khr_fp64 not supported; skipping double tests.\n"); return 0; } else { log_info("Testing doubles...\n"); - return test_geom_normalize_double( deviceID, context, queue, num_elements, seed); + return test_geom_normalize_double(device, context, queue, num_elements, + seed); } } -int test_geom_fast_normalize(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(geom_fast_normalize) { size_t sizes[] = { 1, 2, 3, 4, 0 }; unsigned int size; From e52f6a732b43f8a7b868bd17682d88f75b8de23b Mon Sep 17 00:00:00 2001 From: Ahmed Hesham <117350656+ahesham-arm@users.noreply.github.com> Date: Tue, 18 Mar 2025 17:15:28 +0000 Subject: [PATCH 08/11] Migrate conversions suite to the new test registration framework (#2322) Contributes to #2181. Signed-off-by: Ahmed Hesham --- test_conformance/conversions/test_conversions.cpp | 15 ++++----------- 1 file changed, 4 insertions(+), 11 deletions(-) diff --git a/test_conformance/conversions/test_conversions.cpp b/test_conformance/conversions/test_conversions.cpp index c63572c3..82cb5c2d 100644 --- a/test_conformance/conversions/test_conversions.cpp +++ b/test_conformance/conversions/test_conversions.cpp @@ -92,8 +92,7 @@ char appName[64] = "ctest"; int gMultithread = 1; -int test_conversions(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(conversions) { if (argCount) { @@ -108,13 +107,6 @@ int test_conversions(cl_device_id device, cl_context context, } -test_definition test_list[] = { - ADD_TEST(conversions), -}; - -const int test_num = ARRAY_SIZE(test_list); - - int main(int argc, const char **argv) { int error; @@ -148,8 +140,9 @@ int main(int argc, const char **argv) gMTdata = init_genrand(gRandomSeed); const char *arg[] = { argv[0] }; - int ret = - runTestHarnessWithCheck(1, arg, test_num, test_list, true, 0, InitCL); + int ret = runTestHarnessWithCheck( + 1, arg, test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), true, 0, InitCL); free_mtdata(gMTdata); if (gQueue) From 11db5e5d62f5568ddb287381b79475e1e2538c5e Mon Sep 17 00:00:00 2001 From: Ahmed Hesham <117350656+ahesham-arm@users.noreply.github.com> Date: Tue, 18 Mar 2025 17:22:11 +0000 Subject: [PATCH 09/11] Migrate multiple_device_context suite to the new test registration framework (#2325) Contributes to #2181. Signed-off-by: Ahmed Hesham --- .../multiple_device_context/main.cpp | 30 ++---------------- .../multiple_device_context/procs.h | 31 ------------------- .../multiple_device_context/testBase.h | 16 +++------- .../test_multiple_contexts.cpp | 22 ++++++------- .../test_multiple_devices.cpp | 12 +++---- 5 files changed, 24 insertions(+), 87 deletions(-) delete mode 100644 test_conformance/multiple_device_context/procs.h diff --git a/test_conformance/multiple_device_context/main.cpp b/test_conformance/multiple_device_context/main.cpp index 6e16c244..65777e48 100644 --- a/test_conformance/multiple_device_context/main.cpp +++ b/test_conformance/multiple_device_context/main.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,34 +13,10 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "harness/compat.h" - -#include -#include -#include "procs.h" #include "harness/testHarness.h" -#include "harness/mt19937.h" - -#if !defined(_WIN32) -#include -#endif - -test_definition test_list[] = { - ADD_TEST( context_multiple_contexts_same_device ), - ADD_TEST( context_two_contexts_same_device ), - ADD_TEST( context_three_contexts_same_device ), - ADD_TEST( context_four_contexts_same_device ), - - ADD_TEST( two_devices ), - ADD_TEST( max_devices ), - - ADD_TEST( hundred_queues ), -}; - -const int test_num = ARRAY_SIZE( test_list ); int main(int argc, const char *argv[]) { - return runTestHarness(argc, argv, test_num, test_list, true, 0); + return runTestHarness(argc, argv, test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), true, 0); } - diff --git a/test_conformance/multiple_device_context/procs.h b/test_conformance/multiple_device_context/procs.h deleted file mode 100644 index fb1aae02..00000000 --- a/test_conformance/multiple_device_context/procs.h +++ /dev/null @@ -1,31 +0,0 @@ -// -// Copyright (c) 2017 The Khronos Group Inc. -// -// Licensed under the Apache License, Version 2.0 (the "License"); -// you may not use this file except in compliance with the License. -// You may obtain a copy of the License at -// -// http://www.apache.org/licenses/LICENSE-2.0 -// -// Unless required by applicable law or agreed to in writing, software -// distributed under the License is distributed on an "AS IS" BASIS, -// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -// See the License for the specific language governing permissions and -// limitations under the License. -// -#include "harness/errorHelpers.h" -#include "harness/kernelHelpers.h" -#include "harness/typeWrappers.h" -#include "harness/mt19937.h" - -extern int test_context_multiple_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_context_two_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_context_three_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_context_four_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_two_devices(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_max_devices(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_hundred_queues(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements); - - diff --git a/test_conformance/multiple_device_context/testBase.h b/test_conformance/multiple_device_context/testBase.h index 5b49bfd7..ce4e265c 100644 --- a/test_conformance/multiple_device_context/testBase.h +++ b/test_conformance/multiple_device_context/testBase.h @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -16,16 +16,8 @@ #ifndef _testBase_h #define _testBase_h -#include "harness/compat.h" - -#include -#include -#include -#include - -#include "procs.h" +#include "harness/conversions.h" +#include "harness/testHarness.h" +#include "harness/kernelHelpers.h" #endif // _testBase_h - - - diff --git a/test_conformance/multiple_device_context/test_multiple_contexts.cpp b/test_conformance/multiple_device_context/test_multiple_contexts.cpp index 52ac4682..bcf78a28 100644 --- a/test_conformance/multiple_device_context/test_multiple_contexts.cpp +++ b/test_conformance/multiple_device_context/test_multiple_contexts.cpp @@ -414,8 +414,9 @@ cl_int UseTestItem( const TestItem *item, cl_int *err ) } - -int test_context_multiple_contexts_same_device(cl_device_id deviceID, size_t maxCount, size_t minCount ) +static int test_context_multiple_contexts_same_device(cl_device_id deviceID, + size_t maxCount, + size_t minCount) { size_t i, j; cl_int err = CL_SUCCESS; @@ -525,23 +526,22 @@ exit: // sane limit, currently 200), attempting to use each along the way. We keep track of how many we could make before // a failure occurred. We then free everything and attempt to go do it again a few times. If you are able to make // that many contexts 5 times over, then you pass. -int test_context_multiple_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(context_multiple_contexts_same_device) { - return test_context_multiple_contexts_same_device(deviceID, 200, 1); + return test_context_multiple_contexts_same_device(device, 200, 1); } -int test_context_two_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(context_two_contexts_same_device) { - return test_context_multiple_contexts_same_device( deviceID, 2, 2 ); + return test_context_multiple_contexts_same_device(device, 2, 2); } -int test_context_three_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(context_three_contexts_same_device) { - return test_context_multiple_contexts_same_device( deviceID, 3, 3 ); + return test_context_multiple_contexts_same_device(device, 3, 3); } -int test_context_four_contexts_same_device(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(context_four_contexts_same_device) { - return test_context_multiple_contexts_same_device( deviceID, 4, 4 ); + return test_context_multiple_contexts_same_device(device, 4, 4); } - diff --git a/test_conformance/multiple_device_context/test_multiple_devices.cpp b/test_conformance/multiple_device_context/test_multiple_devices.cpp index 8e45efd0..d0db81db 100644 --- a/test_conformance/multiple_device_context/test_multiple_devices.cpp +++ b/test_conformance/multiple_device_context/test_multiple_devices.cpp @@ -35,7 +35,8 @@ const char *test_kernels[] = { "__kernel void kernelA(__global uint *dst)\n" #define MAX_DEVICES 32 #define MAX_QUEUES 1000 -int test_device_set(size_t deviceCount, size_t queueCount, cl_device_id *devices, int num_elements) +static int test_device_set(size_t deviceCount, size_t queueCount, + cl_device_id *devices, int num_elements) { int error; clContextWrapper context; @@ -187,7 +188,7 @@ int test_device_set(size_t deviceCount, size_t queueCount, cl_device_id *devices return 0; } -int test_two_devices(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(two_devices) { cl_platform_id platform; cl_device_id devices[2]; @@ -215,7 +216,7 @@ int test_two_devices(cl_device_id deviceID, cl_context context, cl_command_queue return test_device_set( 2, 2, devices, num_elements ); } -int test_max_devices(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(max_devices) { cl_platform_id platform; cl_device_id devices[MAX_DEVICES]; @@ -235,8 +236,7 @@ int test_max_devices(cl_device_id deviceID, cl_context context, cl_command_queue return test_device_set( deviceCount, deviceCount, devices, num_elements ); } -int test_hundred_queues(cl_device_id device, cl_context contextIgnore, cl_command_queue queueIgnore, int num_elements) +REGISTER_TEST(hundred_queues) { - return test_device_set( 1, 100, &device, num_elements ); + return test_device_set(1, 100, &device, num_elements); } - From b39eff86c54cefbe6ab7dc0b6c9b59298505ee7d Mon Sep 17 00:00:00 2001 From: Ahmed Hesham <117350656+ahesham-arm@users.noreply.github.com> Date: Tue, 18 Mar 2025 17:25:14 +0000 Subject: [PATCH 10/11] Migrate math_brute_force to the new test registration framework (#2327) Contributes to #2181. Signed-off-by: Ahmed Hesham --- test_conformance/math_brute_force/main.cpp | 165 ++++++++++++++------- 1 file changed, 109 insertions(+), 56 deletions(-) diff --git a/test_conformance/math_brute_force/main.cpp b/test_conformance/math_brute_force/main.cpp index c6a4b5d6..f0f2a4b6 100644 --- a/test_conformance/math_brute_force/main.cpp +++ b/test_conformance/math_brute_force/main.cpp @@ -263,61 +263,112 @@ static int doTest(const char *name) return error; } +#define DO_TEST(name) \ + REGISTER_TEST_VERSION(name, Version(1, 0)) { return doTest(#name); } -#define TEST_LAMBDA(name) \ - [](cl_device_id, cl_context, cl_command_queue, int) { \ - return doTest(#name); \ - } - -// Redefine ADD_TEST to use TEST_LAMBDA. -#undef ADD_TEST -#define ADD_TEST(name) \ - { \ - TEST_LAMBDA(name), #name, Version(1, 0) \ - } - -static test_definition test_list[] = { - ADD_TEST(acos), ADD_TEST(acosh), ADD_TEST(acospi), - ADD_TEST(asin), ADD_TEST(asinh), ADD_TEST(asinpi), - ADD_TEST(atan), ADD_TEST(atanh), ADD_TEST(atanpi), - ADD_TEST(atan2), ADD_TEST(atan2pi), ADD_TEST(cbrt), - ADD_TEST(ceil), ADD_TEST(copysign), ADD_TEST(cos), - ADD_TEST(cosh), ADD_TEST(cospi), ADD_TEST(exp), - ADD_TEST(exp2), ADD_TEST(exp10), ADD_TEST(expm1), - ADD_TEST(fabs), ADD_TEST(fdim), ADD_TEST(floor), - ADD_TEST(fma), ADD_TEST(fmax), ADD_TEST(fmin), - ADD_TEST(fmod), ADD_TEST(fract), ADD_TEST(frexp), - ADD_TEST(hypot), ADD_TEST(ilogb), ADD_TEST(isequal), - ADD_TEST(isfinite), ADD_TEST(isgreater), ADD_TEST(isgreaterequal), - ADD_TEST(isinf), ADD_TEST(isless), ADD_TEST(islessequal), - ADD_TEST(islessgreater), ADD_TEST(isnan), ADD_TEST(isnormal), - ADD_TEST(isnotequal), ADD_TEST(isordered), ADD_TEST(isunordered), - ADD_TEST(ldexp), ADD_TEST(lgamma), ADD_TEST(lgamma_r), - ADD_TEST(log), ADD_TEST(log2), ADD_TEST(log10), - ADD_TEST(log1p), ADD_TEST(logb), ADD_TEST(mad), - ADD_TEST(maxmag), ADD_TEST(minmag), ADD_TEST(modf), - ADD_TEST(nan), ADD_TEST(nextafter), ADD_TEST(pow), - ADD_TEST(pown), ADD_TEST(powr), ADD_TEST(remainder), - ADD_TEST(remquo), ADD_TEST(rint), ADD_TEST(rootn), - ADD_TEST(round), ADD_TEST(rsqrt), ADD_TEST(signbit), - ADD_TEST(sin), ADD_TEST(sincos), ADD_TEST(sinh), - ADD_TEST(sinpi), ADD_TEST(sqrt), ADD_TEST(sqrt_cr), - ADD_TEST(tan), ADD_TEST(tanh), ADD_TEST(tanpi), - ADD_TEST(trunc), ADD_TEST(half_cos), ADD_TEST(half_divide), - ADD_TEST(half_exp), ADD_TEST(half_exp2), ADD_TEST(half_exp10), - ADD_TEST(half_log), ADD_TEST(half_log2), ADD_TEST(half_log10), - ADD_TEST(half_powr), ADD_TEST(half_recip), ADD_TEST(half_rsqrt), - ADD_TEST(half_sin), ADD_TEST(half_sqrt), ADD_TEST(half_tan), - ADD_TEST(add), ADD_TEST(subtract), ADD_TEST(reciprocal), - ADD_TEST(divide), ADD_TEST(divide_cr), ADD_TEST(multiply), - ADD_TEST(assignment), ADD_TEST(not ), ADD_TEST(erf), - ADD_TEST(erfc), -}; - -#undef ADD_TEST -#undef TEST_LAMBDA - -static const int test_num = ARRAY_SIZE(test_list); +DO_TEST(acos) +DO_TEST(acosh) +DO_TEST(acospi) +DO_TEST(asin) +DO_TEST(asinh) +DO_TEST(asinpi) +DO_TEST(atan) +DO_TEST(atanh) +DO_TEST(atanpi) +DO_TEST(atan2) +DO_TEST(atan2pi) +DO_TEST(cbrt) +DO_TEST(ceil) +DO_TEST(copysign) +DO_TEST(cos) +DO_TEST(cosh) +DO_TEST(cospi) +DO_TEST(exp) +DO_TEST(exp2) +DO_TEST(exp10) +DO_TEST(expm1) +DO_TEST(fabs) +DO_TEST(fdim) +DO_TEST(floor) +DO_TEST(fma) +DO_TEST(fmax) +DO_TEST(fmin) +DO_TEST(fmod) +DO_TEST(fract) +DO_TEST(frexp) +DO_TEST(hypot) +DO_TEST(ilogb) +DO_TEST(isequal) +DO_TEST(isfinite) +DO_TEST(isgreater) +DO_TEST(isgreaterequal) +DO_TEST(isinf) +DO_TEST(isless) +DO_TEST(islessequal) +DO_TEST(islessgreater) +DO_TEST(isnan) +DO_TEST(isnormal) +DO_TEST(isnotequal) +DO_TEST(isordered) +DO_TEST(isunordered) +DO_TEST(ldexp) +DO_TEST(lgamma) +DO_TEST(lgamma_r) +DO_TEST(log) +DO_TEST(log2) +DO_TEST(log10) +DO_TEST(log1p) +DO_TEST(logb) +DO_TEST(mad) +DO_TEST(maxmag) +DO_TEST(minmag) +DO_TEST(modf) +DO_TEST(nan) +DO_TEST(nextafter) +DO_TEST(pow) +DO_TEST(pown) +DO_TEST(powr) +DO_TEST(remainder) +DO_TEST(remquo) +DO_TEST(rint) +DO_TEST(rootn) +DO_TEST(round) +DO_TEST(rsqrt) +DO_TEST(signbit) +DO_TEST(sin) +DO_TEST(sincos) +DO_TEST(sinh) +DO_TEST(sinpi) +DO_TEST(sqrt) +DO_TEST(sqrt_cr) +DO_TEST(tan) +DO_TEST(tanh) +DO_TEST(tanpi) +DO_TEST(trunc) +DO_TEST(half_cos) +DO_TEST(half_divide) +DO_TEST(half_exp) +DO_TEST(half_exp2) +DO_TEST(half_exp10) +DO_TEST(half_log) +DO_TEST(half_log2) +DO_TEST(half_log10) +DO_TEST(half_powr) +DO_TEST(half_recip) +DO_TEST(half_rsqrt) +DO_TEST(half_sin) +DO_TEST(half_sqrt) +DO_TEST(half_tan) +DO_TEST(add) +DO_TEST(subtract) +DO_TEST(reciprocal) +DO_TEST(divide) +DO_TEST(divide_cr) +DO_TEST(multiply) +DO_TEST(assignment) +DO_TEST(not ) +DO_TEST(erf) +DO_TEST(erfc) #pragma mark - @@ -355,8 +406,10 @@ int main(int argc, const char *argv[]) FPU_mode_type oldMode; DisableFTZ(&oldMode); - int ret = runTestHarnessWithCheck(gTestNames.size(), gTestNames.data(), - test_num, test_list, true, 0, InitCL); + int ret = runTestHarnessWithCheck( + gTestNames.size(), gTestNames.data(), + test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), true, 0, InitCL); RestoreFPState(&oldMode); From e88e5be93ec14eff0014978b5ca563d8fa1f8608 Mon Sep 17 00:00:00 2001 From: Ahmed Hesham <117350656+ahesham-arm@users.noreply.github.com> Date: Tue, 18 Mar 2025 17:32:58 +0000 Subject: [PATCH 11/11] Migrate basic suite to the new test registration framework (#2316) Contributes to #2181. Signed-off-by: Ahmed Hesham --- test_conformance/basic/main.cpp | 151 +------------ test_conformance/basic/procs.h | 198 ------------------ test_conformance/basic/testBase.h | 26 +++ test_conformance/basic/test_arraycopy.cpp | 11 +- .../basic/test_arrayimagecopy.cpp | 30 ++- .../basic/test_arrayreadwrite.cpp | 10 +- test_conformance/basic/test_astype.cpp | 27 +-- test_conformance/basic/test_async_copy.cpp | 33 ++- test_conformance/basic/test_async_copy2D.cpp | 30 ++- test_conformance/basic/test_async_copy3D.cpp | 32 ++- .../basic/test_async_copy_fence.cpp | 62 +++--- .../basic/test_async_strided_copy.cpp | 25 ++- test_conformance/basic/test_barrier.cpp | 14 +- .../basic/test_basic_parameter_types.cpp | 9 +- .../basic/test_bufferreadwriterect.cpp | 64 +++--- test_conformance/basic/test_constant.cpp | 5 +- .../basic/test_constant_source.cpp | 10 +- .../basic/test_createkernelsinprogram.cpp | 11 +- test_conformance/basic/test_enqueue_map.cpp | 11 +- .../basic/test_enqueued_local_size.cpp | 5 +- test_conformance/basic/test_explicit_s2v.cpp | 16 +- test_conformance/basic/test_fpmath.cpp | 5 +- .../basic/test_get_linear_ids.cpp | 9 +- .../basic/test_global_linear_id.cpp | 5 +- .../basic/test_global_work_offsets.cpp | 9 +- test_conformance/basic/test_hiloeo.cpp | 12 +- test_conformance/basic/test_hostptr.cpp | 11 +- test_conformance/basic/test_if.cpp | 5 +- .../basic/test_image_multipass.cpp | 14 +- test_conformance/basic/test_image_param.cpp | 4 +- test_conformance/basic/test_image_r8.cpp | 5 +- .../basic/test_imagearraycopy.cpp | 30 ++- test_conformance/basic/test_imagecopy.cpp | 9 +- test_conformance/basic/test_imagecopy3d.cpp | 38 ++-- test_conformance/basic/test_imagedim.cpp | 14 +- test_conformance/basic/test_imagenpot.cpp | 24 +-- .../basic/test_imagerandomcopy.cpp | 9 +- .../basic/test_imagereadwrite.cpp | 9 +- .../basic/test_imagereadwrite3d.cpp | 9 +- test_conformance/basic/test_int2fp.cpp | 8 +- test_conformance/basic/test_intmath.cpp | 25 +-- .../test_kernel_call_kernel_function.cpp | 9 +- .../basic/test_kernel_memory_alignment.cpp | 27 +-- test_conformance/basic/test_local.cpp | 10 +- .../basic/test_local_kernel_scope.cpp | 7 +- .../basic/test_local_linear_id.cpp | 7 +- test_conformance/basic/test_loop.cpp | 5 +- .../basic/test_multireadimagemultifmt.cpp | 11 +- .../basic/test_multireadimageonefmt.cpp | 10 +- .../basic/test_numeric_constants.cpp | 28 +-- test_conformance/basic/test_pointercast.cpp | 7 +- test_conformance/basic/test_preprocessors.cpp | 22 +- test_conformance/basic/test_progvar.cpp | 14 +- .../basic/test_queue_priority.cpp | 7 +- test_conformance/basic/test_readimage.cpp | 25 +-- .../basic/test_rw_image_access_qualifier.cpp | 17 +- .../basic/test_simple_image_pitch.cpp | 26 ++- test_conformance/basic/test_sizeof.cpp | 7 +- test_conformance/basic/test_vec_type_hint.cpp | 7 +- .../basic/test_vector_creation.cpp | 8 +- .../basic/test_vector_swizzle.cpp | 6 +- test_conformance/basic/test_vloadstore.cpp | 38 ++-- .../basic/test_work_item_functions.cpp | 20 +- test_conformance/basic/test_writeimage.cpp | 20 +- 64 files changed, 443 insertions(+), 929 deletions(-) delete mode 100644 test_conformance/basic/procs.h create mode 100644 test_conformance/basic/testBase.h diff --git a/test_conformance/basic/main.cpp b/test_conformance/basic/main.cpp index 988bb6f5..f69397dc 100644 --- a/test_conformance/basic/main.cpp +++ b/test_conformance/basic/main.cpp @@ -14,151 +14,12 @@ // limitations under the License. // #include "harness/compat.h" - -#if !defined(_WIN32) -#include -#endif - -#include -#include -#include +#include "harness/deviceInfo.h" +#include "harness/kernelHelpers.h" +#include "harness/testHarness.h" #include -#include "harness/testHarness.h" -#include "procs.h" - -test_definition test_list[] = { - ADD_TEST(hostptr), - ADD_TEST(fpmath), - ADD_TEST(intmath_int), - ADD_TEST(intmath_int2), - ADD_TEST(intmath_int4), - ADD_TEST(intmath_long), - ADD_TEST(intmath_long2), - ADD_TEST(intmath_long4), - ADD_TEST(hiloeo), - ADD_TEST(if), - ADD_TEST(sizeof), - ADD_TEST(loop), - ADD_TEST(pointer_cast), - ADD_TEST(local_arg_def), - ADD_TEST(local_kernel_def), - ADD_TEST(local_kernel_scope), - ADD_TEST(constant), - ADD_TEST(constant_source), - ADD_TEST(readimage), - ADD_TEST(readimage_int16), - ADD_TEST(readimage_fp32), - ADD_TEST(writeimage), - ADD_TEST(writeimage_int16), - ADD_TEST(writeimage_fp32), - ADD_TEST(mri_one), - - ADD_TEST(mri_multiple), - ADD_TEST(image_r8), - ADD_TEST(barrier), - ADD_TEST_VERSION(wg_barrier, Version(2, 0)), - ADD_TEST(int2fp), - ADD_TEST(fp2int), - ADD_TEST(imagereadwrite), - ADD_TEST(imagereadwrite3d), - ADD_TEST(readimage3d), - ADD_TEST(readimage3d_int16), - ADD_TEST(readimage3d_fp32), - ADD_TEST(bufferreadwriterect), - ADD_TEST(arrayreadwrite), - ADD_TEST(arraycopy), - ADD_TEST(imagearraycopy), - ADD_TEST(imagearraycopy3d), - ADD_TEST(imagecopy), - ADD_TEST(imagecopy3d), - ADD_TEST(imagerandomcopy), - ADD_TEST(arrayimagecopy), - ADD_TEST(arrayimagecopy3d), - ADD_TEST(imagenpot), - - ADD_TEST(vload_global), - ADD_TEST(vload_local), - ADD_TEST(vload_constant), - ADD_TEST(vload_private), - ADD_TEST(vstore_global), - ADD_TEST(vstore_local), - ADD_TEST(vstore_private), - - ADD_TEST(createkernelsinprogram), - ADD_TEST(imagedim_pow2), - ADD_TEST(imagedim_non_pow2), - ADD_TEST(image_param), - ADD_TEST(image_multipass_integer_coord), - ADD_TEST(image_multipass_float_coord), - - ADD_TEST(explicit_s2v), - - ADD_TEST(enqueue_map_buffer), - ADD_TEST(enqueue_map_image), - - ADD_TEST(work_item_functions), - ADD_TEST(work_item_functions_out_of_range), - ADD_TEST(work_item_functions_out_of_range_hardcoded), - - ADD_TEST(astype), - - ADD_TEST(async_copy_global_to_local), - ADD_TEST(async_copy_local_to_global), - ADD_TEST(async_strided_copy_global_to_local), - ADD_TEST(async_strided_copy_local_to_global), - ADD_TEST(async_copy_global_to_local2D), - ADD_TEST(async_copy_local_to_global2D), - ADD_TEST(async_copy_global_to_local3D), - ADD_TEST(async_copy_local_to_global3D), - ADD_TEST(async_work_group_copy_fence_import_after_export_aliased_local), - ADD_TEST(async_work_group_copy_fence_import_after_export_aliased_global), - ADD_TEST( - async_work_group_copy_fence_import_after_export_aliased_global_and_local), - ADD_TEST(async_work_group_copy_fence_export_after_import_aliased_local), - ADD_TEST(async_work_group_copy_fence_export_after_import_aliased_global), - ADD_TEST( - async_work_group_copy_fence_export_after_import_aliased_global_and_local), - ADD_TEST(prefetch), - ADD_TEST(kernel_call_kernel_function), - ADD_TEST(host_numeric_constants), - ADD_TEST(kernel_numeric_constants), - ADD_TEST(kernel_limit_constants), - ADD_TEST(kernel_preprocessor_macros), - ADD_TEST(parameter_types), - ADD_TEST(vector_creation), - ADD_TEST(vector_swizzle), - ADD_TEST(vec_type_hint), - ADD_TEST(kernel_memory_alignment_local), - ADD_TEST(kernel_memory_alignment_global), - ADD_TEST(kernel_memory_alignment_constant), - ADD_TEST(kernel_memory_alignment_private), - - ADD_TEST_VERSION(progvar_prog_scope_misc, Version(2, 0)), - ADD_TEST_VERSION(progvar_prog_scope_uninit, Version(2, 0)), - ADD_TEST_VERSION(progvar_prog_scope_init, Version(2, 0)), - ADD_TEST_VERSION(progvar_func_scope, Version(2, 0)), - - ADD_TEST(global_work_offsets), - ADD_TEST(get_global_offset), - - ADD_TEST_VERSION(global_linear_id, Version(2, 0)), - ADD_TEST_VERSION(local_linear_id, Version(2, 0)), - ADD_TEST_VERSION(enqueued_local_size, Version(2, 0)), - - ADD_TEST(simple_read_image_pitch), - ADD_TEST(simple_write_image_pitch), - -#if defined(__APPLE__) - ADD_TEST(queue_priority), -#endif - - ADD_TEST_VERSION(get_linear_ids, Version(2, 0)), - ADD_TEST_VERSION(rw_image_access_qualifier, Version(2, 0)), -}; - -const int test_num = ARRAY_SIZE( test_list ); cl_half_rounding_mode halfRoundingMode = CL_HALF_RTE; test_status InitCL(cl_device_id device) @@ -187,7 +48,7 @@ test_status InitCL(cl_device_id device) int main(int argc, const char *argv[]) { - return runTestHarnessWithCheck(argc, argv, test_num, test_list, false, 0, - InitCL); + return runTestHarnessWithCheck( + argc, argv, test_registry::getInstance().num_tests(), + test_registry::getInstance().definitions(), false, 0, InitCL); } - diff --git a/test_conformance/basic/procs.h b/test_conformance/basic/procs.h deleted file mode 100644 index 69529b1f..00000000 --- a/test_conformance/basic/procs.h +++ /dev/null @@ -1,198 +0,0 @@ -// -// Copyright (c) 2023 The Khronos Group Inc. -// -// Licensed under the Apache License, Version 2.0 (the "License"); -// you may not use this file except in compliance with the License. -// You may obtain a copy of the License at -// -// http://www.apache.org/licenses/LICENSE-2.0 -// -// Unless required by applicable law or agreed to in writing, software -// distributed under the License is distributed on an "AS IS" BASIS, -// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. -// See the License for the specific language governing permissions and -// limitations under the License. -// - -#include "harness/kernelHelpers.h" -#include "harness/testHarness.h" -#include "harness/errorHelpers.h" -#include "harness/typeWrappers.h" -#include "harness/conversions.h" -#include "harness/rounding_mode.h" - -extern int test_hostptr(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_fpmath(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_intmath_int(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_intmath_int2(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_intmath_int4(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_intmath_long(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_intmath_long2(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_intmath_long4(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_hiloeo(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_if(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_sizeof(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_loop(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_pointer_cast(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_local_arg_def(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_local_kernel_def(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_local_kernel_scope(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_constant(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_constant_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_readimage(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_readimage_int16(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_readimage_fp32(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_writeimage(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_writeimage_int16(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_writeimage_fp32(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_mri_one(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_mri_multiple(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_image_r8(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_simplebarrier(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_barrier(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_wg_barrier(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_int2fp(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_fp2int(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_imagearraycopy(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagearraycopy3d(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagereadwrite(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagereadwrite3d(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_readimage3d(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_readimage3d_int16(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_readimage3d_fp32(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_bufferreadwriterect(cl_device_id device, cl_context context, cl_command_queue queue_, int num_elements); -extern int test_imagecopy(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagecopy3d(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagerandomcopy(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_arraycopy(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems); -extern int test_arrayimagecopy(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_arrayimagecopy3d(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagenpot(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_sampler_float(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_sampler_int(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_createkernelsinprogram(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_single_large_allocation(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_multiple_max_allocation(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_arrayreadwrite(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagedim_pow2(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_imagedim_non_pow2(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_image_param(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_image_multipass_integer_coord(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_image_multipass_float_coord(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_vload_global(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vload_local(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vload_constant(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vload_private(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vstore_global(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vstore_local(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vstore_private(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_explicit_s2v(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - -extern int test_enqueue_map_buffer(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_enqueue_map_image(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_work_item_functions(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_work_item_functions_out_of_range(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_work_item_functions_out_of_range_hardcoded( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); - -extern int test_astype(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_native_kernel(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_async_copy_global_to_local(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_async_copy_local_to_global(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_async_strided_copy_global_to_local(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_async_strided_copy_local_to_global(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_async_copy_global_to_local2D(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_async_copy_local_to_global2D(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_async_copy_global_to_local3D(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_async_copy_local_to_global3D(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements); -extern int test_async_work_group_copy_fence_import_after_export_aliased_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_async_work_group_copy_fence_import_after_export_aliased_global( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int -test_async_work_group_copy_fence_import_after_export_aliased_global_and_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_async_work_group_copy_fence_export_after_import_aliased_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_async_work_group_copy_fence_export_after_import_aliased_global( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int -test_async_work_group_copy_fence_export_after_import_aliased_global_and_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements); -extern int test_prefetch(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_host_numeric_constants(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_kernel_numeric_constants(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); -extern int test_kernel_limit_constants(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_kernel_call_kernel_function(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); - -extern int test_parameter_types(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements); -extern int test_vector_creation(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_vector_swizzle(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); -extern int test_vec_type_hint(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements); - - -extern int test_kernel_memory_alignment_local(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_kernel_memory_alignment_global(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_kernel_memory_alignment_constant(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_kernel_memory_alignment_private(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); - -extern int test_progvar_prog_scope_misc(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_progvar_prog_scope_uninit(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_progvar_prog_scope_init(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_progvar_func_scope(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); - -extern int test_global_work_offsets(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_get_global_offset(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); - -extern int test_global_linear_id(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_local_linear_id(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); -extern int test_enqueued_local_size(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ); - -extern int test_simple_read_image_pitch(cl_device_id device, cl_context cl_context_, cl_command_queue q, int num_elements); -extern int test_simple_write_image_pitch(cl_device_id device, cl_context cl_context_, cl_command_queue q, int num_elements); - -#if defined( __APPLE__ ) -extern int test_queue_priority(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements); -#endif - -extern int test_get_linear_ids(cl_device_id device, cl_context cl_context_, cl_command_queue q, int num_elements); -extern int test_rw_image_access_qualifier(cl_device_id device_id, cl_context context, cl_command_queue commands, int num_elements); - diff --git a/test_conformance/basic/testBase.h b/test_conformance/basic/testBase.h new file mode 100644 index 00000000..d7b5af31 --- /dev/null +++ b/test_conformance/basic/testBase.h @@ -0,0 +1,26 @@ +// +// Copyright (c) 2025 The Khronos Group Inc. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. +// +#ifndef _testBase_h +#define _testBase_h + +#include + +#include "harness/conversions.h" +#include "harness/testHarness.h" +#include "harness/typeWrappers.h" +#include "harness/rounding_mode.h" + +#endif // _testBase_h diff --git a/test_conformance/basic/test_arraycopy.cpp b/test_conformance/basic/test_arraycopy.cpp index d9dbcc1b..a981cd02 100644 --- a/test_conformance/basic/test_arraycopy.cpp +++ b/test_conformance/basic/test_arraycopy.cpp @@ -20,8 +20,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" const char *copy_kernel_code = "__kernel void test_copy(__global unsigned int *src, __global unsigned int *dst)\n" @@ -31,14 +30,13 @@ const char *copy_kernel_code = " dst[tid] = src[tid];\n" "}\n"; -int -test_arraycopy(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems) +REGISTER_TEST(arraycopy) { cl_uint *input_ptr, *output_ptr; cl_mem streams[4], results; cl_program program; cl_kernel kernel; - unsigned num_elements = 128 * 1024; + num_elements = 128 * 1024; cl_uint num_copies = 1; size_t delta_offset; unsigned i; @@ -202,6 +200,3 @@ test_arraycopy(cl_device_id device, cl_context context, cl_command_queue queue, return err; } - - - diff --git a/test_conformance/basic/test_arrayimagecopy.cpp b/test_conformance/basic/test_arrayimagecopy.cpp index 91e308a0..8a8f9381 100644 --- a/test_conformance/basic/test_arrayimagecopy.cpp +++ b/test_conformance/basic/test_arrayimagecopy.cpp @@ -23,19 +23,14 @@ #include #include #include - -#include "procs.h" - using test_function_t = int (*)(cl_device_id, cl_context, cl_command_queue, cl_mem_flags, cl_mem_flags, cl_mem_object_type, const cl_image_format *); -int test_arrayimagecopy_single_format(cl_device_id device, cl_context context, - cl_command_queue queue, - cl_mem_flags buffer_flags, - cl_mem_flags image_flags, - cl_mem_object_type image_type, - const cl_image_format *format) +static int test_arrayimagecopy_single_format( + cl_device_id device, cl_context context, cl_command_queue queue, + cl_mem_flags buffer_flags, cl_mem_flags image_flags, + cl_mem_object_type image_type, const cl_image_format *format) { std::unique_ptr bufptr{ nullptr, free }, imgptr{ nullptr, free }; @@ -153,11 +148,12 @@ int test_arrayimagecopy_single_format(cl_device_id device, cl_context context, } -int test_arrayimagecommon(cl_device_id device, cl_context context, - cl_command_queue queue, cl_mem_flags buffer_flags, - cl_mem_flags image_flags, - cl_mem_object_type image_type, - test_function_t test_function) +static int test_arrayimagecommon(cl_device_id device, cl_context context, + cl_command_queue queue, + cl_mem_flags buffer_flags, + cl_mem_flags image_flags, + cl_mem_object_type image_type, + test_function_t test_function) { cl_int err; cl_uint num_formats; @@ -188,8 +184,7 @@ int test_arrayimagecommon(cl_device_id device, cl_context context, return err; } -int test_arrayimagecopy(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(arrayimagecopy) { PASSIVE_REQUIRE_IMAGE_SUPPORT(device) @@ -199,8 +194,7 @@ int test_arrayimagecopy(cl_device_id device, cl_context context, } -int test_arrayimagecopy3d(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(arrayimagecopy3d) { PASSIVE_REQUIRE_3D_IMAGE_SUPPORT(device) diff --git a/test_conformance/basic/test_arrayreadwrite.cpp b/test_conformance/basic/test_arrayreadwrite.cpp index 25e8ed99..ac2c2d7e 100644 --- a/test_conformance/basic/test_arrayreadwrite.cpp +++ b/test_conformance/basic/test_arrayreadwrite.cpp @@ -21,12 +21,9 @@ #include #include +#include "testBase.h" -#include "procs.h" - - -int -test_arrayreadwrite(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(arrayreadwrite) { cl_uint *inptr, *outptr; cl_mem streams[1]; @@ -91,6 +88,3 @@ test_arrayreadwrite(cl_device_id device, cl_context context, cl_command_queue qu return err; } - - - diff --git a/test_conformance/basic/test_astype.cpp b/test_conformance/basic/test_astype.cpp index 45669a7c..e838ba53 100644 --- a/test_conformance/basic/test_astype.cpp +++ b/test_conformance/basic/test_astype.cpp @@ -24,9 +24,6 @@ #include #include #include - -#include "procs.h" - // clang-format off static char extension[128] = { 0 }; @@ -48,9 +45,10 @@ extension, // clang-format on -int test_astype_set( cl_device_id device, cl_context context, cl_command_queue queue, ExplicitType inVecType, ExplicitType outVecType, - unsigned int vecSize, unsigned int outVecSize, - int numElements ) +static int test_astype_set(cl_device_id device, cl_context context, + cl_command_queue queue, ExplicitType inVecType, + ExplicitType outVecType, unsigned int vecSize, + unsigned int outVecSize, int numElements) { int error; @@ -179,7 +177,7 @@ int test_astype_set( cl_device_id device, cl_context context, cl_command_queue q return 0; } -int test_astype(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(astype) { // Note: although casting to different vector element sizes that match the same size (i.e. short2 -> char4) is // legal in OpenCL 1.0, the result is dependent on the device it runs on, which means there's no actual way @@ -236,19 +234,24 @@ int test_astype(cl_device_id device, cl_context context, cl_command_queue queue, { continue; } - error += test_astype_set( device, context, queue, vecTypes[ inTypeIdx ], vecTypes[ outTypeIdx ], vecSizes[ sizeIdx ], vecSizes[outSizeIdx], n_elems ); + error += test_astype_set( + device, context, queue, vecTypes[inTypeIdx], + vecTypes[outTypeIdx], vecSizes[sizeIdx], + vecSizes[outSizeIdx], num_elements); } } if(get_explicit_type_size(vecTypes[inTypeIdx]) == get_explicit_type_size(vecTypes[outTypeIdx])) { // as_type3(vec4) allowed, as_type4(vec3) not allowed - error += test_astype_set( device, context, queue, vecTypes[ inTypeIdx ], vecTypes[ outTypeIdx ], 3, 4, n_elems ); - error += test_astype_set( device, context, queue, vecTypes[ inTypeIdx ], vecTypes[ outTypeIdx ], 4, 3, n_elems ); + error += + test_astype_set(device, context, queue, vecTypes[inTypeIdx], + vecTypes[outTypeIdx], 3, 4, num_elements); + error += + test_astype_set(device, context, queue, vecTypes[inTypeIdx], + vecTypes[outTypeIdx], 4, 3, num_elements); } } } return error; } - - diff --git a/test_conformance/basic/test_async_copy.cpp b/test_conformance/basic/test_async_copy.cpp index bb529bce..c0b08505 100644 --- a/test_conformance/basic/test_async_copy.cpp +++ b/test_conformance/basic/test_async_copy.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,8 +13,6 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "harness/compat.h" - #include #include #include @@ -22,8 +20,7 @@ #include #include -#include "procs.h" -#include "harness/conversions.h" +#include "testBase.h" static const char *async_global_to_local_kernel = "%s\n" // optional pragma string @@ -75,10 +72,9 @@ static const char *prefetch_kernel = "}\n" ; - -int test_copy(cl_device_id deviceID, cl_context context, cl_command_queue queue, const char *kernelCode, - ExplicitType vecType, int vecSize - ) +static int test_copy(cl_device_id deviceID, cl_context context, + cl_command_queue queue, const char *kernelCode, + ExplicitType vecType, int vecSize) { int error; clProgramWrapper program; @@ -238,7 +234,9 @@ int test_copy(cl_device_id deviceID, cl_context context, cl_command_queue queue, return failuresPrinted ? -1 : 0; } -int test_copy_all_types(cl_device_id deviceID, cl_context context, cl_command_queue queue, const char *kernelCode) { +static int test_copy_all_types(cl_device_id deviceID, cl_context context, + cl_command_queue queue, const char *kernelCode) +{ const std::vector vecType = { kChar, kUChar, kShort, kUShort, kInt, kUInt, kLong, kULong, kFloat, kHalf, kDouble }; @@ -271,18 +269,19 @@ int test_copy_all_types(cl_device_id deviceID, cl_context context, cl_command_qu return 0; } -int test_async_copy_global_to_local(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(async_copy_global_to_local) { - return test_copy_all_types( deviceID, context, queue, async_global_to_local_kernel ); + return test_copy_all_types(device, context, queue, + async_global_to_local_kernel); } -int test_async_copy_local_to_global(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(async_copy_local_to_global) { - return test_copy_all_types( deviceID, context, queue, async_local_to_global_kernel ); + return test_copy_all_types(device, context, queue, + async_local_to_global_kernel); } -int test_prefetch(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(prefetch) { - return test_copy_all_types( deviceID, context, queue, prefetch_kernel ); + return test_copy_all_types(device, context, queue, prefetch_kernel); } - diff --git a/test_conformance/basic/test_async_copy2D.cpp b/test_conformance/basic/test_async_copy2D.cpp index 0cbfe779..bdd4bb5e 100644 --- a/test_conformance/basic/test_async_copy2D.cpp +++ b/test_conformance/basic/test_async_copy2D.cpp @@ -13,8 +13,6 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "../../test_common/harness/compat.h" - #include #include #include @@ -22,8 +20,7 @@ #include #include -#include "../../test_common/harness/conversions.h" -#include "procs.h" +#include "testBase.h" static const char *async_global_to_local_kernel2D = R"OpenCLC( #pragma OPENCL EXTENSION cl_khr_extended_async_copies : enable @@ -113,10 +110,11 @@ __kernel void test_fn(const __global VarSizeStruct *src, __global VarSizeStruct }; )OpenCLC"; -int test_copy2D(const cl_device_id deviceID, const cl_context context, - const cl_command_queue queue, const char *const kernelCode, - const size_t elementSize, const int srcMargin, - const int dstMargin, const bool localIsDst) +static int test_copy2D(const cl_device_id deviceID, const cl_context context, + const cl_command_queue queue, + const char *const kernelCode, const size_t elementSize, + const int srcMargin, const int dstMargin, + const bool localIsDst) { int error; @@ -388,9 +386,9 @@ int test_copy2D(const cl_device_id deviceID, const cl_context context, return failuresPrinted ? -1 : 0; } -int test_copy2D_all_types(cl_device_id deviceID, cl_context context, - cl_command_queue queue, const char *kernelCode, - bool localIsDst) +static int test_copy2D_all_types(cl_device_id deviceID, cl_context context, + cl_command_queue queue, const char *kernelCode, + bool localIsDst) { const unsigned int elemSizes[] = { 1, 2, 3, 4, 5, 6, 7, 8, 13, 16, 32, 47, 64 }; @@ -430,16 +428,14 @@ int test_copy2D_all_types(cl_device_id deviceID, cl_context context, return errors ? -1 : 0; } -int test_async_copy_global_to_local2D(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(async_copy_global_to_local2D) { - return test_copy2D_all_types(deviceID, context, queue, + return test_copy2D_all_types(device, context, queue, async_global_to_local_kernel2D, true); } -int test_async_copy_local_to_global2D(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(async_copy_local_to_global2D) { - return test_copy2D_all_types(deviceID, context, queue, + return test_copy2D_all_types(device, context, queue, async_local_to_global_kernel2D, false); } diff --git a/test_conformance/basic/test_async_copy3D.cpp b/test_conformance/basic/test_async_copy3D.cpp index f4641d63..9e5fe01f 100644 --- a/test_conformance/basic/test_async_copy3D.cpp +++ b/test_conformance/basic/test_async_copy3D.cpp @@ -13,8 +13,6 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "../../test_common/harness/compat.h" - #include #include #include @@ -22,8 +20,7 @@ #include #include -#include "../../test_common/harness/conversions.h" -#include "procs.h" +#include "testBase.h" static const char *async_global_to_local_kernel3D = R"OpenCLC( #pragma OPENCL EXTENSION cl_khr_extended_async_copies : enable @@ -124,11 +121,12 @@ __kernel void test_fn(const __global VarSizeStruct *src, __global VarSizeStruct } )OpenCLC"; -int test_copy3D(const cl_device_id deviceID, const cl_context context, - const cl_command_queue queue, const char *const kernelCode, - const size_t elementSize, const int srcLineMargin, - const int dstLineMargin, const int srcPlaneMargin, - const int dstPlaneMargin, const bool localIsDst) +static int test_copy3D(const cl_device_id deviceID, const cl_context context, + const cl_command_queue queue, + const char *const kernelCode, const size_t elementSize, + const int srcLineMargin, const int dstLineMargin, + const int srcPlaneMargin, const int dstPlaneMargin, + const bool localIsDst) { int error; @@ -445,9 +443,9 @@ int test_copy3D(const cl_device_id deviceID, const cl_context context, return failuresPrinted ? -1 : 0; } -int test_copy3D_all_types(cl_device_id deviceID, cl_context context, - cl_command_queue queue, const char *kernelCode, - bool localIsDst) +static int test_copy3D_all_types(cl_device_id deviceID, cl_context context, + cl_command_queue queue, const char *kernelCode, + bool localIsDst) { const unsigned int elemSizes[] = { 1, 2, 3, 4, 5, 6, 7, 8, 13, 16, 32, 47, 64 }; @@ -500,16 +498,14 @@ int test_copy3D_all_types(cl_device_id deviceID, cl_context context, return 0; } -int test_async_copy_global_to_local3D(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(async_copy_global_to_local3D) { - return test_copy3D_all_types(deviceID, context, queue, + return test_copy3D_all_types(device, context, queue, async_global_to_local_kernel3D, true); } -int test_async_copy_local_to_global3D(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(async_copy_local_to_global3D) { - return test_copy3D_all_types(deviceID, context, queue, + return test_copy3D_all_types(device, context, queue, async_local_to_global_kernel3D, false); } diff --git a/test_conformance/basic/test_async_copy_fence.cpp b/test_conformance/basic/test_async_copy_fence.cpp index b8c2e931..d2677a40 100644 --- a/test_conformance/basic/test_async_copy_fence.cpp +++ b/test_conformance/basic/test_async_copy_fence.cpp @@ -13,16 +13,13 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "../../test_common/harness/compat.h" - #include #include #include #include #include -#include "../../test_common/harness/conversions.h" -#include "procs.h" +#include "testBase.h" static const char *import_after_export_aliased_local_kernel = "#pragma OPENCL EXTENSION cl_khr_async_work_group_copy_fence : enable\n" @@ -331,10 +328,11 @@ static const char *export_after_import_aliased_global_and_local_kernel = " }\n" "}\n"; -int test_copy_fence(cl_device_id deviceID, cl_context context, - cl_command_queue queue, const char *kernelCode, - ExplicitType vecType, int vecSize, bool export_after_import, - bool aliased_local_mem, bool aliased_global_mem) +static int test_copy_fence(cl_device_id deviceID, cl_context context, + cl_command_queue queue, const char *kernelCode, + ExplicitType vecType, int vecSize, + bool export_after_import, bool aliased_local_mem, + bool aliased_global_mem) { int error; clProgramWrapper program; @@ -710,10 +708,12 @@ int test_copy_fence(cl_device_id deviceID, cl_context context, return failuresPrinted ? -1 : 0; } -int test_copy_fence_all_types(cl_device_id deviceID, cl_context context, - cl_command_queue queue, const char *kernelCode, - bool export_after_import, bool aliased_local_mem, - bool aliased_global_mem) +static int test_copy_fence_all_types(cl_device_id deviceID, cl_context context, + cl_command_queue queue, + const char *kernelCode, + bool export_after_import, + bool aliased_local_mem, + bool aliased_global_mem) { ExplicitType vecType[] = { kChar, kUChar, kShort, kUShort, kInt, kUInt, kLong, @@ -757,56 +757,46 @@ int test_copy_fence_all_types(cl_device_id deviceID, cl_context context, return 0; } -int test_async_work_group_copy_fence_import_after_export_aliased_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(async_work_group_copy_fence_import_after_export_aliased_local) { - return test_copy_fence_all_types(deviceID, context, queue, + return test_copy_fence_all_types(device, context, queue, import_after_export_aliased_local_kernel, false, true, false); } -int test_async_work_group_copy_fence_import_after_export_aliased_global( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(async_work_group_copy_fence_import_after_export_aliased_global) { - return test_copy_fence_all_types(deviceID, context, queue, + return test_copy_fence_all_types(device, context, queue, import_after_export_aliased_global_kernel, false, false, true); } -int test_async_work_group_copy_fence_import_after_export_aliased_global_and_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST( + async_work_group_copy_fence_import_after_export_aliased_global_and_local) { return test_copy_fence_all_types( - deviceID, context, queue, + device, context, queue, import_after_export_aliased_global_and_local_kernel, false, true, true); } -int test_async_work_group_copy_fence_export_after_import_aliased_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(async_work_group_copy_fence_export_after_import_aliased_local) { - return test_copy_fence_all_types(deviceID, context, queue, + return test_copy_fence_all_types(device, context, queue, export_after_import_aliased_local_kernel, true, true, false); } -int test_async_work_group_copy_fence_export_after_import_aliased_global( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(async_work_group_copy_fence_export_after_import_aliased_global) { - return test_copy_fence_all_types(deviceID, context, queue, + return test_copy_fence_all_types(device, context, queue, export_after_import_aliased_global_kernel, true, false, true); } -int test_async_work_group_copy_fence_export_after_import_aliased_global_and_local( - cl_device_id deviceID, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST( + async_work_group_copy_fence_export_after_import_aliased_global_and_local) { return test_copy_fence_all_types( - deviceID, context, queue, + device, context, queue, export_after_import_aliased_global_and_local_kernel, true, true, true); } diff --git a/test_conformance/basic/test_async_strided_copy.cpp b/test_conformance/basic/test_async_strided_copy.cpp index 932e9b8c..d6c89669 100644 --- a/test_conformance/basic/test_async_strided_copy.cpp +++ b/test_conformance/basic/test_async_strided_copy.cpp @@ -13,8 +13,6 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "harness/compat.h" - #include #include #include @@ -22,8 +20,7 @@ #include #include -#include "procs.h" -#include "harness/conversions.h" +#include "testBase.h" // clang-format off @@ -66,7 +63,9 @@ static const char *async_strided_local_to_global_kernel = // clang-format on -int test_strided_copy(cl_device_id deviceID, cl_context context, cl_command_queue queue, const char *kernelCode, ExplicitType vecType, int vecSize, int stride) +static int test_strided_copy(cl_device_id deviceID, cl_context context, + cl_command_queue queue, const char *kernelCode, + ExplicitType vecType, int vecSize, int stride) { int error; clProgramWrapper program; @@ -233,7 +232,10 @@ int test_strided_copy(cl_device_id deviceID, cl_context context, cl_command_queu return 0; } -int test_strided_copy_all_types(cl_device_id deviceID, cl_context context, cl_command_queue queue, const char *kernelCode) +static int test_strided_copy_all_types(cl_device_id deviceID, + cl_context context, + cl_command_queue queue, + const char *kernelCode) { const std::vector vecType = { kChar, kUChar, kShort, kUShort, kInt, kUInt, kLong, kULong, @@ -272,13 +274,14 @@ int test_strided_copy_all_types(cl_device_id deviceID, cl_context context, cl_co return 0; } -int test_async_strided_copy_global_to_local(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(async_strided_copy_global_to_local) { - return test_strided_copy_all_types( deviceID, context, queue, async_strided_global_to_local_kernel ); + return test_strided_copy_all_types(device, context, queue, + async_strided_global_to_local_kernel); } -int test_async_strided_copy_local_to_global(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(async_strided_copy_local_to_global) { - return test_strided_copy_all_types( deviceID, context, queue, async_strided_local_to_global_kernel ); + return test_strided_copy_all_types(device, context, queue, + async_strided_local_to_global_kernel); } - diff --git a/test_conformance/basic/test_barrier.cpp b/test_conformance/basic/test_barrier.cpp index 6352b42f..9aa17aec 100644 --- a/test_conformance/basic/test_barrier.cpp +++ b/test_conformance/basic/test_barrier.cpp @@ -25,7 +25,7 @@ #include #include -#include "procs.h" +#include "testBase.h" namespace { const char *barrier_kernel_code = R"( @@ -66,9 +66,9 @@ void generate_random_inputs(std::vector &v) std::generate(v.begin(), v.end(), random_generator); } -int test_barrier_common(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements, - std::string barrier_str) +static int test_barrier_common(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements, + std::string barrier_str) { clMemWrapper streams[3]; clProgramWrapper program; @@ -142,14 +142,12 @@ int test_barrier_common(cl_device_id device, cl_context context, } } -int test_barrier(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(barrier) { return test_barrier_common(device, context, queue, num_elements, "barrier"); } -int test_wg_barrier(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(wg_barrier, Version(2, 0)) { return test_barrier_common(device, context, queue, num_elements, "work_group_barrier"); diff --git a/test_conformance/basic/test_basic_parameter_types.cpp b/test_conformance/basic/test_basic_parameter_types.cpp index 6e99d462..277d9647 100644 --- a/test_conformance/basic/test_basic_parameter_types.cpp +++ b/test_conformance/basic/test_basic_parameter_types.cpp @@ -21,7 +21,7 @@ #include #include -#include "procs.h" +#include "testBase.h" const char *kernel_code = R"( __kernel void test_kernel( @@ -46,8 +46,8 @@ __global float%s *result) result[1] = %s(ul); })"; -int test_parameter_types_long(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +static int test_parameter_types_long(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements) { clMemWrapper results; int error; @@ -196,8 +196,7 @@ int test_parameter_types_long(cl_device_id device, cl_context context, return total_errors; } -int test_parameter_types(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(parameter_types) { clMemWrapper results; int error; diff --git a/test_conformance/basic/test_bufferreadwriterect.cpp b/test_conformance/basic/test_bufferreadwriterect.cpp index 0a12bf79..0178200e 100644 --- a/test_conformance/basic/test_bufferreadwriterect.cpp +++ b/test_conformance/basic/test_bufferreadwriterect.cpp @@ -21,7 +21,7 @@ #include #include -#include "procs.h" +#include "testBase.h" #define CL_EXIT_ERROR(cmd,format,...) \ { \ @@ -36,7 +36,7 @@ log_error("\n"); \ typedef unsigned char BufferType; // Globals for test -cl_command_queue queue; +cl_command_queue gQueue; // Width and height of each pair of images. enum { TotalImages = 8 }; @@ -150,13 +150,12 @@ int copy_region(size_t src, size_t soffset[3], size_t sregion[3], size_t dst, si log_info( "Copy overlap reported, skipping copy buffer rect\n" ); return CL_SUCCESS; } else { - if ((err = clEnqueueCopyBufferRect(queue, - buffer[src],buffer[dst], - soffset, doffset, - sregion,/*dregion,*/ - width[src], src_slice_pitch, - width[dst], dst_slice_pitch, - 0, NULL, NULL)) != CL_SUCCESS) + if ((err = clEnqueueCopyBufferRect( + gQueue, buffer[src], buffer[dst], soffset, doffset, + sregion, /*dregion,*/ + width[src], src_slice_pitch, width[dst], dst_slice_pitch, 0, + NULL, NULL)) + != CL_SUCCESS) { CL_EXIT_ERROR(err, "clEnqueueCopyBufferRect failed between %u and %u",(unsigned)src,(unsigned)dst); } @@ -253,15 +252,12 @@ int read_verify_region(size_t src, size_t soffset[3], size_t sregion[3], size_t size_t dst_slice_pitch = (width[dst]*height[dst] != 1) ? width[dst]*height[dst] : 0; // Copy the source region of the cl buffer, to the destination region of the temporary buffer. - CL_EXIT_ERROR(clEnqueueReadBufferRect(queue, - buffer[src], - CL_TRUE, - soffset,doffset, - sregion, - width[src], src_slice_pitch, - width[dst], dst_slice_pitch, - tmp_buffer, - 0, NULL, NULL), "clEnqueueCopyBufferRect failed between %u and %u",(unsigned)src,(unsigned)dst); + CL_EXIT_ERROR(clEnqueueReadBufferRect( + gQueue, buffer[src], CL_TRUE, soffset, doffset, sregion, + width[src], src_slice_pitch, width[dst], dst_slice_pitch, + tmp_buffer, 0, NULL, NULL), + "clEnqueueCopyBufferRect failed between %u and %u", + (unsigned)src, (unsigned)dst); return verify_region(tmp_buffer,src,soffset,sregion,dst,doffset); } @@ -276,7 +272,9 @@ int map_verify_region(size_t src) { // Copy the source region of the cl buffer, to the destination region of the temporary buffer. cl_int err; - BufferType* mapped = (BufferType*)clEnqueueMapBuffer(queue,buffer[src],CL_TRUE,CL_MAP_READ,0,size_bytes,0,NULL,NULL,&err); + BufferType* mapped = (BufferType*)clEnqueueMapBuffer( + gQueue, buffer[src], CL_TRUE, CL_MAP_READ, 0, size_bytes, 0, NULL, NULL, + &err); CL_EXIT_ERROR(err, "clEnqueueMapBuffer failed for buffer %u",(unsigned)src); size_t soffset[] = { 0, 0, 0 }; @@ -284,8 +282,9 @@ int map_verify_region(size_t src) { int ret = verify_region(mapped,src,soffset,sregion,src,soffset); - CL_EXIT_ERROR(clEnqueueUnmapMemObject(queue,buffer[src],mapped,0,NULL,NULL), - "clEnqueueUnmapMemObject failed for buffer %u",(unsigned)src); + CL_EXIT_ERROR( + clEnqueueUnmapMemObject(gQueue, buffer[src], mapped, 0, NULL, NULL), + "clEnqueueUnmapMemObject failed for buffer %u", (unsigned)src); return ret; } @@ -301,15 +300,12 @@ int write_region(size_t src, size_t soffset[3], size_t sregion[3], size_t dst, s size_t dst_slice_pitch = (width[dst]*height[dst] != 1) ? width[dst]*height[dst] : 0; // Copy the source region of the cl buffer, to the destination region of the temporary buffer. - CL_EXIT_ERROR(clEnqueueWriteBufferRect(queue, - buffer[dst], - CL_TRUE, - doffset,soffset, - /*sregion,*/dregion, - width[dst], dst_slice_pitch, - width[src], src_slice_pitch, - tmp_buffer, - 0, NULL, NULL), "clEnqueueWriteBufferRect failed between %u and %u",(unsigned)src,(unsigned)dst); + CL_EXIT_ERROR(clEnqueueWriteBufferRect( + gQueue, buffer[dst], CL_TRUE, doffset, soffset, + /*sregion,*/ dregion, width[dst], dst_slice_pitch, + width[src], src_slice_pitch, tmp_buffer, 0, NULL, NULL), + "clEnqueueWriteBufferRect failed between %u and %u", + (unsigned)src, (unsigned)dst); // Copy from the temporary buffer to the host buffer. size_t spitch = width[src]; @@ -345,10 +341,9 @@ void CL_CALLBACK mem_obj_destructor_callback( cl_mem, void *data ) } // This is the main test function for the conformance test. -int -test_bufferreadwriterect(cl_device_id device, cl_context context, cl_command_queue queue_, int num_elements) +REGISTER_TEST(bufferreadwriterect) { - queue = queue_; + gQueue = queue; cl_int err; // Initialize the random number generator. @@ -564,6 +559,3 @@ test_bufferreadwriterect(cl_device_id device, cl_context context, cl_command_que return err; } - - - diff --git a/test_conformance/basic/test_constant.cpp b/test_conformance/basic/test_constant.cpp index e9f41848..b7d7790c 100644 --- a/test_conformance/basic/test_constant.cpp +++ b/test_conformance/basic/test_constant.cpp @@ -24,7 +24,7 @@ #include #include -#include "procs.h" +#include "testBase.h" namespace { const char* constant_kernel_code = R"( @@ -99,8 +99,7 @@ template void generate_random_inputs(std::vector& v) } } -int test_constant(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(constant) { clMemWrapper streams[3]; clProgramWrapper program; diff --git a/test_conformance/basic/test_constant_source.cpp b/test_conformance/basic/test_constant_source.cpp index d6099e7c..22ecbb98 100644 --- a/test_conformance/basic/test_constant_source.cpp +++ b/test_conformance/basic/test_constant_source.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" const char *constant_source_kernel_code[] = { "__constant int outVal = 42;\n" @@ -44,7 +43,7 @@ const char *constant_source_kernel_code[] = { " }\n" "}\n" }; -int test_constant_source(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(constant_source) { clProgramWrapper program; clKernelWrapper kernel; @@ -94,8 +93,3 @@ int test_constant_source(cl_device_id device, cl_context context, cl_command_que return 0; } - - - - - diff --git a/test_conformance/basic/test_createkernelsinprogram.cpp b/test_conformance/basic/test_createkernelsinprogram.cpp index 77eba3f9..39e62dce 100644 --- a/test_conformance/basic/test_createkernelsinprogram.cpp +++ b/test_conformance/basic/test_createkernelsinprogram.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" const char *sample_single_kernel = { "__kernel void sample_test(__global float *src, __global int *dst)\n" @@ -50,8 +49,7 @@ const char *sample_double_kernel = { "}\n"}; -int -test_createkernelsinprogram(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(createkernelsinprogram) { cl_program program; cl_kernel kernel[2]; @@ -98,8 +96,3 @@ test_createkernelsinprogram(cl_device_id device, cl_context context, cl_command_ return err; } - - - - - diff --git a/test_conformance/basic/test_enqueue_map.cpp b/test_conformance/basic/test_enqueue_map.cpp index c2ea24ef..c13cebe4 100644 --- a/test_conformance/basic/test_enqueue_map.cpp +++ b/test_conformance/basic/test_enqueue_map.cpp @@ -20,9 +20,6 @@ #include #include #include - - -#include "procs.h" #include "harness/conversions.h" #include "harness/typeWrappers.h" @@ -44,8 +41,7 @@ const char *flag_set_names[] = { }; // clang-format on -int test_enqueue_map_buffer(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(enqueue_map_buffer) { int error; const size_t bufferSize = 256 * 256; @@ -141,15 +137,14 @@ int test_enqueue_map_buffer(cl_device_id deviceID, cl_context context, return 0; } -int test_enqueue_map_image(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(enqueue_map_image) { int error; cl_image_format format = { CL_RGBA, CL_UNSIGNED_INT32 }; const size_t imageSize = 256; const size_t imageDataSize = imageSize * imageSize * 4 * sizeof(cl_uint); - PASSIVE_REQUIRE_IMAGE_SUPPORT(deviceID) + PASSIVE_REQUIRE_IMAGE_SUPPORT(device) BufferOwningPtr hostPtrData{ malloc(imageDataSize) }; BufferOwningPtr referenceData{ malloc(imageDataSize) }; diff --git a/test_conformance/basic/test_enqueued_local_size.cpp b/test_conformance/basic/test_enqueued_local_size.cpp index ea95df68..47123120 100644 --- a/test_conformance/basic/test_enqueued_local_size.cpp +++ b/test_conformance/basic/test_enqueued_local_size.cpp @@ -24,7 +24,7 @@ #include -#include "procs.h" +#include "testBase.h" static const char *enqueued_local_size_2d_code = R"( __kernel void test_enqueued_local_size_2d(global int *dst) @@ -65,8 +65,7 @@ static int verify_enqueued_local_size(int *result, size_t *expected, int n) } -int test_enqueued_local_size(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(enqueued_local_size, Version(2, 0)) { clMemWrapper stream; clProgramWrapper program[2]; diff --git a/test_conformance/basic/test_explicit_s2v.cpp b/test_conformance/basic/test_explicit_s2v.cpp index a061f9bb..a5ae452f 100644 --- a/test_conformance/basic/test_explicit_s2v.cpp +++ b/test_conformance/basic/test_explicit_s2v.cpp @@ -25,8 +25,6 @@ using std::isnan; #include #include - -#include "procs.h" #include "harness/conversions.h" #include "harness/typeWrappers.h" @@ -114,10 +112,11 @@ bool IsHalfNaN(cl_half v) return (h_exp == 0x1F && h_mant != 0); } -int test_explicit_s2v_function(cl_context context, cl_command_queue queue, - cl_kernel kernel, ExplicitType srcType, - unsigned int count, ExplicitType destType, - unsigned int vecSize, void *inputData) +static int test_explicit_s2v_function(cl_context context, + cl_command_queue queue, cl_kernel kernel, + ExplicitType srcType, unsigned int count, + ExplicitType destType, + unsigned int vecSize, void *inputData) { int error; clMemWrapper streams[2]; @@ -387,12 +386,11 @@ protected: } // anonymous namespace -int test_explicit_s2v(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(explicit_s2v) { try { - TypesIterator(deviceID, context, queue); + TypesIterator(device, context, queue); } catch (const std::runtime_error &e) { log_error("%s", e.what()); diff --git a/test_conformance/basic/test_fpmath.cpp b/test_conformance/basic/test_fpmath.cpp index fc7b9108..c39a2fec 100644 --- a/test_conformance/basic/test_fpmath.cpp +++ b/test_conformance/basic/test_fpmath.cpp @@ -31,7 +31,7 @@ #include #include -#include "procs.h" +#include "testBase.h" extern cl_half_rounding_mode halfRoundingMode; @@ -380,8 +380,7 @@ protected: } // anonymous namespace -int test_fpmath(cl_device_id device, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(fpmath) { try { diff --git a/test_conformance/basic/test_get_linear_ids.cpp b/test_conformance/basic/test_get_linear_ids.cpp index ee7dfb2f..b8c1a3c3 100644 --- a/test_conformance/basic/test_get_linear_ids.cpp +++ b/test_conformance/basic/test_get_linear_ids.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,9 +13,10 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "procs.h" #include +#include "testBase.h" + static const char *linear_ids_source[1] = { "__kernel void test_linear_ids(__global int2 *out)\n" "{\n" @@ -46,8 +47,7 @@ static const char *linear_ids_source[1] = { #define MAX_3D 16 #define MAX_OFFSET 100000 -int -test_get_linear_ids(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(get_linear_ids, Version(2, 0)) { clProgramWrapper program; clKernelWrapper kernel; @@ -193,4 +193,3 @@ test_get_linear_ids(cl_device_id device, cl_context context, cl_command_queue qu free_mtdata(seed); return 0; } - diff --git a/test_conformance/basic/test_global_linear_id.cpp b/test_conformance/basic/test_global_linear_id.cpp index ccd8ce27..de462f23 100644 --- a/test_conformance/basic/test_global_linear_id.cpp +++ b/test_conformance/basic/test_global_linear_id.cpp @@ -25,7 +25,7 @@ #include #include -#include "procs.h" +#include "testBase.h" namespace { const char *global_linear_id_2d_code = R"( @@ -64,8 +64,7 @@ int verify_global_linear_id(std::vector &result, int n) } } -int test_global_linear_id(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(global_linear_id, Version(2, 0)) { clProgramWrapper program[2]; clKernelWrapper kernel[2]; diff --git a/test_conformance/basic/test_global_work_offsets.cpp b/test_conformance/basic/test_global_work_offsets.cpp index af4d1cb6..557de703 100644 --- a/test_conformance/basic/test_global_work_offsets.cpp +++ b/test_conformance/basic/test_global_work_offsets.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,9 +13,9 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "procs.h" #include +#include "testBase.h" const char *work_offset_test[] = { "__kernel void test( __global int * outputID_A, \n" @@ -122,7 +122,7 @@ int check_results( size_t threads[], size_t offsets[], cl_int outputA[], cl_int return ( missed | multiple | errored | corrected ); } -int test_global_work_offsets(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(global_work_offsets) { clProgramWrapper program; clKernelWrapper kernel; @@ -222,7 +222,7 @@ const char *get_offset_test[] = { "}\n" }; -int test_get_global_offset(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(get_global_offset) { clProgramWrapper program; clKernelWrapper kernel; @@ -300,4 +300,3 @@ int test_get_global_offset(cl_device_id deviceID, cl_context context, cl_command // All done! return 0; } - diff --git a/test_conformance/basic/test_hiloeo.cpp b/test_conformance/basic/test_hiloeo.cpp index 4e921a6e..7b5fe4f8 100644 --- a/test_conformance/basic/test_hiloeo.cpp +++ b/test_conformance/basic/test_hiloeo.cpp @@ -21,7 +21,7 @@ #include #include -#include "procs.h" +#include "testBase.h" int hi_offset( int index, int vectorSize) { return index + vectorSize / 2; } int lo_offset( int index, int vectorSize) { return index; } @@ -48,7 +48,7 @@ static const char *vector_size_names[] = { "", "2", "3", "4", "8", "16"}; static const size_t kSizes[] = { 1, 1, 2, 2, 4, 4, 8, 8, 2, 4, 8 }; static int CheckResults( void *in, void *out, size_t elementCount, int type, int vectorSize, int operatorToUse ); -int test_hiloeo(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems) +REGISTER_TEST(hiloeo) { int err; int hasDouble = is_extension_available( device, "cl_khr_fp64" ); @@ -60,12 +60,12 @@ int test_hiloeo(cl_device_id device, cl_context context, cl_command_queue queue, int expressionMode; int numExpressionModes = 2; - size_t length = sizeof(cl_int) * 4 * n_elems; + size_t length = sizeof(cl_int) * 4 * num_elements; - std::vector input_ptr(4 * n_elems); - std::vector output_ptr(4 * n_elems); + std::vector input_ptr(4 * num_elements); + std::vector output_ptr(4 * num_elements); - for (cl_uint i = 0; i < 4 * (cl_uint)n_elems; i++) + for (cl_uint i = 0; i < 4 * (cl_uint)num_elements; i++) input_ptr[i] = genrand_int32(d); for( type = 0; type < sizeof( test_str_names ) / sizeof( test_str_names[0] ); type++ ) diff --git a/test_conformance/basic/test_hostptr.cpp b/test_conformance/basic/test_hostptr.cpp index dee78675..e58b636e 100644 --- a/test_conformance/basic/test_hostptr.cpp +++ b/test_conformance/basic/test_hostptr.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" const char *hostptr_kernel_code = "__kernel void test_hostptr(__global float *srcA, __global float *srcB, __global float *dst)\n" @@ -92,8 +91,7 @@ verify_rgba8_image(unsigned char *image, unsigned char *outptr, int w, int h) return 0; } -int -test_hostptr(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(hostptr) { cl_float *input_ptr[2], *output_ptr; cl_program program; @@ -282,8 +280,3 @@ test_hostptr(cl_device_id device, cl_context context, cl_command_queue queue, in return err; } - - - - - diff --git a/test_conformance/basic/test_if.cpp b/test_conformance/basic/test_if.cpp index f2a8fa82..a0640879 100644 --- a/test_conformance/basic/test_if.cpp +++ b/test_conformance/basic/test_if.cpp @@ -24,7 +24,7 @@ #include #include -#include "procs.h" +#include "testBase.h" namespace { const char *conditional_kernel_code = R"( @@ -88,8 +88,7 @@ void generate_random_inputs(std::vector &v) std::generate(v.begin(), v.end(), random_generator); } } -int test_if(cl_device_id device, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(if) { clMemWrapper streams[2]; clProgramWrapper program; diff --git a/test_conformance/basic/test_image_multipass.cpp b/test_conformance/basic/test_image_multipass.cpp index 7f51665c..5d8ae993 100644 --- a/test_conformance/basic/test_image_multipass.cpp +++ b/test_conformance/basic/test_image_multipass.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static const char *image_to_image_kernel_integer_coord_code = "\n" @@ -143,8 +142,7 @@ verify_byte_image(unsigned char *image, unsigned char *outptr, int w, int h, int return 0; } -int -test_image_multipass_integer_coord(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(image_multipass_integer_coord) { int img_width = 512; int img_height = 512; @@ -397,8 +395,7 @@ test_image_multipass_integer_coord(cl_device_id device, cl_context context, cl_c return err; } -int -test_image_multipass_float_coord(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(image_multipass_float_coord) { int img_width = 512; int img_height = 512; @@ -637,8 +634,3 @@ test_image_multipass_float_coord(cl_device_id device, cl_context context, cl_com return err; } - - - - - diff --git a/test_conformance/basic/test_image_param.cpp b/test_conformance/basic/test_image_param.cpp index 3efc4c94..a5b1da71 100644 --- a/test_conformance/basic/test_image_param.cpp +++ b/test_conformance/basic/test_image_param.cpp @@ -21,8 +21,6 @@ #include #include - -#include "procs.h" #include "harness/typeWrappers.h" #include "harness/imageHelpers.h" #include "harness/conversions.h" @@ -161,7 +159,7 @@ int validate_results( size_t width, size_t height, cl_image_format &format, char return 0; } -int test_image_param(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(image_param) { size_t sizes[] = { 64, 100, 128, 250, 512 }; cl_image_format formats[] = { { CL_RGBA, CL_UNORM_INT8 }, { CL_RGBA, CL_UNORM_INT16 }, { CL_RGBA, CL_FLOAT }, { CL_BGRA, CL_UNORM_INT8 } }; diff --git a/test_conformance/basic/test_image_r8.cpp b/test_conformance/basic/test_image_r8.cpp index 2dca1611..3eb44507 100644 --- a/test_conformance/basic/test_image_r8.cpp +++ b/test_conformance/basic/test_image_r8.cpp @@ -24,7 +24,7 @@ #include #include -#include "procs.h" +#include "testBase.h" namespace { const char *r_uint8_kernel_code = R"( @@ -52,8 +52,7 @@ void generate_random_inputs(std::vector &v) } } -int test_image_r8(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(image_r8) { clMemWrapper streams[2]; clProgramWrapper program; diff --git a/test_conformance/basic/test_imagearraycopy.cpp b/test_conformance/basic/test_imagearraycopy.cpp index 0411c789..7f177ff4 100644 --- a/test_conformance/basic/test_imagearraycopy.cpp +++ b/test_conformance/basic/test_imagearraycopy.cpp @@ -23,19 +23,14 @@ #include #include #include - -#include "procs.h" - using test_function_t = int (*)(cl_device_id, cl_context, cl_command_queue, cl_mem_flags, cl_mem_flags, cl_mem_object_type, const cl_image_format *); -int test_imagearraycopy_single_format(cl_device_id device, cl_context context, - cl_command_queue queue, - cl_mem_flags image_flags, - cl_mem_flags buffer_flags, - cl_mem_object_type image_type, - const cl_image_format *format) +static int test_imagearraycopy_single_format( + cl_device_id device, cl_context context, cl_command_queue queue, + cl_mem_flags image_flags, cl_mem_flags buffer_flags, + cl_mem_object_type image_type, const cl_image_format *format) { std::unique_ptr bufptr{ nullptr, free }, imgptr{ nullptr, free }; @@ -153,11 +148,12 @@ int test_imagearraycopy_single_format(cl_device_id device, cl_context context, return err; } -int test_imagearraycommon(cl_device_id device, cl_context context, - cl_command_queue queue, cl_mem_flags image_flags, - cl_mem_flags buffer_flags, - cl_mem_object_type image_type, - test_function_t test_function) +static int test_imagearraycommon(cl_device_id device, cl_context context, + cl_command_queue queue, + cl_mem_flags image_flags, + cl_mem_flags buffer_flags, + cl_mem_object_type image_type, + test_function_t test_function) { cl_int err; cl_uint num_formats; @@ -188,8 +184,7 @@ int test_imagearraycommon(cl_device_id device, cl_context context, return err; } -int test_imagearraycopy(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(imagearraycopy) { PASSIVE_REQUIRE_IMAGE_SUPPORT(device) @@ -199,8 +194,7 @@ int test_imagearraycopy(cl_device_id device, cl_context context, } -int test_imagearraycopy3d(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(imagearraycopy3d) { PASSIVE_REQUIRE_3D_IMAGE_SUPPORT(device) diff --git a/test_conformance/basic/test_imagecopy.cpp b/test_conformance/basic/test_imagecopy.cpp index bcb9fef9..ffcd1b6e 100644 --- a/test_conformance/basic/test_imagecopy.cpp +++ b/test_conformance/basic/test_imagecopy.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static unsigned char * generate_rgba8_image(int w, int h, MTdata d) @@ -105,8 +104,7 @@ verify_rgbafp_image(float *image, float *outptr, int w, int h) } -int -test_imagecopy(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(imagecopy) { cl_image_format img_format; unsigned char *rgba8_inptr, *rgba8_outptr; @@ -236,6 +234,3 @@ test_imagecopy(cl_device_id device, cl_context context, cl_command_queue queue, return err; } - - - diff --git a/test_conformance/basic/test_imagecopy3d.cpp b/test_conformance/basic/test_imagecopy3d.cpp index 39ca7e67..33d9b2bd 100644 --- a/test_conformance/basic/test_imagecopy3d.cpp +++ b/test_conformance/basic/test_imagecopy3d.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static unsigned char * generate_uint8_image(unsigned num_elements, MTdata d) @@ -105,8 +104,7 @@ verify_float_image(float *image, float *outptr, unsigned num_elements) } -int -test_imagecopy3d(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements_ignored) +REGISTER_TEST(imagecopy3d) { cl_image_format img_format; unsigned char *rgba8_inptr, *rgba8_outptr; @@ -118,20 +116,21 @@ test_imagecopy3d(cl_device_id device, cl_context context, cl_command_queue queue int img_depth = 64; int i; cl_int err; - unsigned num_elements = img_width * img_height * img_depth * 4; + unsigned num_elems = img_width * img_height * img_depth * 4; MTdata d; PASSIVE_REQUIRE_3D_IMAGE_SUPPORT( device ) d = init_genrand( gRandomSeed ); - rgba8_inptr = (unsigned char *)generate_uint8_image(num_elements, d); - rgba16_inptr = (unsigned short *)generate_uint16_image(num_elements, d); - rgbafp_inptr = (float *)generate_float_image(num_elements, d); + rgba8_inptr = (unsigned char *)generate_uint8_image(num_elems, d); + rgba16_inptr = (unsigned short *)generate_uint16_image(num_elems, d); + rgbafp_inptr = (float *)generate_float_image(num_elems, d); free_mtdata(d); d = NULL; - rgba8_outptr = (unsigned char*)malloc(sizeof(unsigned char) * num_elements); - rgba16_outptr = (unsigned short*)malloc(sizeof(unsigned short) * num_elements); - rgbafp_outptr = (float*)malloc(sizeof(float) * num_elements); + rgba8_outptr = (unsigned char *)malloc(sizeof(unsigned char) * num_elems); + rgba16_outptr = + (unsigned short *)malloc(sizeof(unsigned short) * num_elems); + rgbafp_outptr = (float *)malloc(sizeof(float) * num_elems); img_format.image_channel_order = CL_RGBA; img_format.image_channel_data_type = CL_UNORM_INT8; @@ -202,16 +201,18 @@ test_imagecopy3d(cl_device_id device, cl_context context, cl_command_queue queue switch (i) { case 0: - err = verify_uint8_image(rgba8_inptr, rgba8_outptr, num_elements); - if (err) log_error("Failed uint8\n"); + err = verify_uint8_image(rgba8_inptr, rgba8_outptr, num_elems); + if (err) log_error("Failed uint8\n"); break; case 1: - err = verify_uint16_image(rgba16_inptr, rgba16_outptr, num_elements); - if (err) log_error("Failed uint16\n"); + err = + verify_uint16_image(rgba16_inptr, rgba16_outptr, num_elems); + if (err) log_error("Failed uint16\n"); break; case 2: - err = verify_float_image(rgbafp_inptr, rgbafp_outptr, num_elements); - if (err) log_error("Failed float\n"); + err = + verify_float_image(rgbafp_inptr, rgbafp_outptr, num_elems); + if (err) log_error("Failed float\n"); break; } @@ -233,6 +234,3 @@ test_imagecopy3d(cl_device_id device, cl_context context, cl_command_queue queue return err; } - - - diff --git a/test_conformance/basic/test_imagedim.cpp b/test_conformance/basic/test_imagedim.cpp index 2107d4c0..cb0b5702 100644 --- a/test_conformance/basic/test_imagedim.cpp +++ b/test_conformance/basic/test_imagedim.cpp @@ -24,7 +24,7 @@ #include #include -#include "procs.h" +#include "testBase.h" namespace { const char *image_dim_kernel_code = R"( @@ -105,9 +105,9 @@ int get_max_image_dimensions(cl_device_id device, size_t &max_img_width, return err; } -int test_imagedim_common(cl_context context, cl_command_queue queue, - cl_kernel kernel, size_t *local_threads, - size_t img_width, size_t img_height) +static int test_imagedim_common(cl_context context, cl_command_queue queue, + cl_kernel kernel, size_t *local_threads, + size_t img_width, size_t img_height) { int err; @@ -173,8 +173,7 @@ int test_imagedim_common(cl_context context, cl_command_queue queue, } } -int test_imagedim_pow2(cl_device_id device, cl_context context, - cl_command_queue queue, int n_elems) +REGISTER_TEST(imagedim_pow2) { clProgramWrapper program; clKernelWrapper kernel; @@ -211,8 +210,7 @@ int test_imagedim_pow2(cl_device_id device, cl_context context, } -int test_imagedim_non_pow2(cl_device_id device, cl_context context, - cl_command_queue queue, int n_elems) +REGISTER_TEST(imagedim_non_pow2) { clProgramWrapper program; clKernelWrapper kernel; diff --git a/test_conformance/basic/test_imagenpot.cpp b/test_conformance/basic/test_imagenpot.cpp index baa5b2e4..1e2c213e 100644 --- a/test_conformance/basic/test_imagenpot.cpp +++ b/test_conformance/basic/test_imagenpot.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static const char *rgba8888_kernel_code = "\n" @@ -74,8 +73,7 @@ verify_rgba8888_image(unsigned char *src, unsigned char *dst, int w, int h) int img_width_selection[] = { 97, 111, 322, 479 }; int img_height_selection[] = { 149, 222, 754, 385 }; -int -test_imagenpot(cl_device_id device_id, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(imagenpot) { cl_mem streams[2]; cl_image_format img_format; @@ -91,10 +89,11 @@ test_imagenpot(cl_device_id device_id, cl_context context, cl_command_queue queu size_t max_local_workgroup_size[3]; MTdata d; - PASSIVE_REQUIRE_IMAGE_SUPPORT( device_id ) + PASSIVE_REQUIRE_IMAGE_SUPPORT(device) cl_device_type device_type; - err = clGetDeviceInfo(device_id, CL_DEVICE_TYPE, sizeof(device_type), &device_type, NULL); + err = clGetDeviceInfo(device, CL_DEVICE_TYPE, sizeof(device_type), + &device_type, NULL); if (err) { log_error("Failed to get device type: %d\n",err); return -1; @@ -163,10 +162,14 @@ test_imagenpot(cl_device_id device_id, cl_context context, cl_command_queue queu return -1; } - err = clGetKernelWorkGroupInfo(kernel, device_id, CL_KERNEL_WORK_GROUP_SIZE, sizeof(local_workgroup_size), &local_workgroup_size, NULL); + err = clGetKernelWorkGroupInfo( + kernel, device, CL_KERNEL_WORK_GROUP_SIZE, + sizeof(local_workgroup_size), &local_workgroup_size, NULL); test_error(err, "clGetKernelWorkGroupInfo for CL_KERNEL_WORK_GROUP_SIZE failed"); - err = clGetDeviceInfo(device_id, CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(max_local_workgroup_size), max_local_workgroup_size, NULL); + err = clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_ITEM_SIZES, + sizeof(max_local_workgroup_size), + max_local_workgroup_size, NULL); test_error(err, "clGetDeviceInfo failed for CL_DEVICE_MAX_WORK_ITEM_SIZES"); // Pick the minimum of the device and the kernel @@ -214,8 +217,3 @@ test_imagenpot(cl_device_id device_id, cl_context context, cl_command_queue queu return err; } - - - - - diff --git a/test_conformance/basic/test_imagerandomcopy.cpp b/test_conformance/basic/test_imagerandomcopy.cpp index c3355de3..79e6b749 100644 --- a/test_conformance/basic/test_imagerandomcopy.cpp +++ b/test_conformance/basic/test_imagerandomcopy.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static unsigned char * generate_rgba8_image(int w, int h, MTdata d) @@ -117,8 +116,7 @@ verify_rgbafp_image(float *image, float *outptr, int x, int y, int w, int h, int #define NUM_COPIES 10 static const char *test_str_names[] = { "CL_RGBA CL_UNORM_INT8", "CL_RGBA CL_UNORM_INT16", "CL_RGBA CL_FLOAT" }; -int -test_imagerandomcopy(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(imagerandomcopy) { cl_image_format img_format; unsigned char *rgba8_inptr, *rgba8_outptr; @@ -271,6 +269,3 @@ test_imagerandomcopy(cl_device_id device, cl_context context, cl_command_queue q return err; } - - - diff --git a/test_conformance/basic/test_imagereadwrite.cpp b/test_conformance/basic/test_imagereadwrite.cpp index fee73ea3..dd0929a0 100644 --- a/test_conformance/basic/test_imagereadwrite.cpp +++ b/test_conformance/basic/test_imagereadwrite.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static unsigned char * generate_rgba8_image(int w, int h, MTdata d) @@ -188,8 +187,7 @@ verify_rgbafp_image(float *image, float *outptr, int w, int h) } -int -test_imagereadwrite(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(imagereadwrite) { cl_image_format img_format; unsigned char *rgba8_inptr, *rgba8_outptr; @@ -422,6 +420,3 @@ test_imagereadwrite(cl_device_id device, cl_context context, cl_command_queue qu return err; } - - - diff --git a/test_conformance/basic/test_imagereadwrite3d.cpp b/test_conformance/basic/test_imagereadwrite3d.cpp index d52f16d0..1d7351a9 100644 --- a/test_conformance/basic/test_imagereadwrite3d.cpp +++ b/test_conformance/basic/test_imagereadwrite3d.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static unsigned char * generate_rgba8_image(int w, int h, int d, MTdata mtData) @@ -196,8 +195,7 @@ verify_rgbafp_image(float *image, float *outptr, int w, int h, int d) } -int -test_imagereadwrite3d(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(imagereadwrite3d) { cl_image_format img_format; unsigned char *rgba8_inptr, *rgba8_outptr; @@ -419,6 +417,3 @@ test_imagereadwrite3d(cl_device_id device, cl_context context, cl_command_queue return err; } - - - diff --git a/test_conformance/basic/test_int2fp.cpp b/test_conformance/basic/test_int2fp.cpp index dd5cc9a1..c709abe8 100644 --- a/test_conformance/basic/test_int2fp.cpp +++ b/test_conformance/basic/test_int2fp.cpp @@ -29,7 +29,7 @@ #include #include -#include "procs.h" +#include "testBase.h" extern cl_half_rounding_mode halfRoundingMode; @@ -294,8 +294,7 @@ protected: } -int test_int2fp(cl_device_id device, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(int2fp) { try { @@ -309,8 +308,7 @@ int test_int2fp(cl_device_id device, cl_context context, cl_command_queue queue, return TEST_PASS; } -int test_fp2int(cl_device_id device, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(fp2int) { try { diff --git a/test_conformance/basic/test_intmath.cpp b/test_conformance/basic/test_intmath.cpp index 5a4e9c2a..21c83655 100644 --- a/test_conformance/basic/test_intmath.cpp +++ b/test_conformance/basic/test_intmath.cpp @@ -18,7 +18,7 @@ #include #include -#include "procs.h" +#include "testBase.h" template struct TestDef { @@ -28,8 +28,9 @@ template struct TestDef }; template -int test_intmath(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements, std::string typestr) +static int test_intmath(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements, + std::string typestr) { TestDef tests[] = { // Test addition @@ -197,43 +198,37 @@ int test_intmath(cl_device_id device, cl_context context, return TEST_PASS; } -int test_intmath_int(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(intmath_int) { return test_intmath(device, context, queue, num_elements, "uint"); } -int test_intmath_int2(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(intmath_int2) { return test_intmath(device, context, queue, num_elements, "uint2"); } -int test_intmath_int4(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(intmath_int4) { return test_intmath(device, context, queue, num_elements, "uint4"); } -int test_intmath_long(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(intmath_long) { return test_intmath(device, context, queue, num_elements, "ulong"); } -int test_intmath_long2(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(intmath_long2) { return test_intmath(device, context, queue, num_elements, "ulong2"); } -int test_intmath_long4(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(intmath_long4) { return test_intmath(device, context, queue, num_elements, "ulong4"); diff --git a/test_conformance/basic/test_kernel_call_kernel_function.cpp b/test_conformance/basic/test_kernel_call_kernel_function.cpp index 80fea55f..0669ee24 100644 --- a/test_conformance/basic/test_kernel_call_kernel_function.cpp +++ b/test_conformance/basic/test_kernel_call_kernel_function.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,7 +13,7 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "procs.h" +#include "testBase.h" const char *kernel_call_kernel_code[] = { "void test_function_to_call(__global int *output, __global int *input, int where);\n" @@ -57,8 +57,7 @@ const char *kernel_call_kernel_code[] = { }; - -int test_kernel_call_kernel_function(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(kernel_call_kernel_function) { num_elements = 256; @@ -249,5 +248,3 @@ int test_kernel_call_kernel_function(cl_device_id deviceID, cl_context context, return errors; } - - diff --git a/test_conformance/basic/test_kernel_memory_alignment.cpp b/test_conformance/basic/test_kernel_memory_alignment.cpp index 869ce2ec..f361c621 100644 --- a/test_conformance/basic/test_kernel_memory_alignment.cpp +++ b/test_conformance/basic/test_kernel_memory_alignment.cpp @@ -17,7 +17,6 @@ #include #endif -#include "procs.h" #include "harness/conversions.h" #include "harness/typeWrappers.h" #include "harness/errorHelpers.h" @@ -148,7 +147,9 @@ const char * get_explicit_address_name( AddressSpaces address ) } -int test_kernel_memory_alignment(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems, AddressSpaces address ) +static int test_kernel_memory_alignment(cl_device_id device, cl_context context, + cl_command_queue queue, int n_elems, + AddressSpaces address) { const char *constant_kernel; const char *parameter_kernel; @@ -495,17 +496,19 @@ int test_kernel_memory_alignment(cl_device_id device, cl_context context, cl_com } -int test_kernel_memory_alignment_local(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(kernel_memory_alignment_local) { - return test_kernel_memory_alignment( device, context, queue, n_elems, kLocal ); + return test_kernel_memory_alignment(device, context, queue, num_elements, + kLocal); } -int test_kernel_memory_alignment_global(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(kernel_memory_alignment_global) { - return test_kernel_memory_alignment( device, context, queue, n_elems, kGlobal ); + return test_kernel_memory_alignment(device, context, queue, num_elements, + kGlobal); } -int test_kernel_memory_alignment_constant(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(kernel_memory_alignment_constant) { // There is a class of approved OpenCL 1.0 conformant devices out there that in some circumstances // are unable to meaningfully take (or more precisely use) the address of constant data by virtue @@ -559,12 +562,12 @@ int test_kernel_memory_alignment_constant(cl_device_id device, cl_context contex free(version_string); // Everyone else is to be ground mercilessly under the wheels of progress - return test_kernel_memory_alignment( device, context, queue, n_elems, kConstant ); + return test_kernel_memory_alignment(device, context, queue, num_elements, + kConstant); } -int test_kernel_memory_alignment_private(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(kernel_memory_alignment_private) { - return test_kernel_memory_alignment( device, context, queue, n_elems, kPrivate ); + return test_kernel_memory_alignment(device, context, queue, num_elements, + kPrivate); } - - diff --git a/test_conformance/basic/test_local.cpp b/test_conformance/basic/test_local.cpp index f0ffe81e..2125ec65 100644 --- a/test_conformance/basic/test_local.cpp +++ b/test_conformance/basic/test_local.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" const char *barrier_with_localmem_kernel_code[] = { "__kernel void compute_sum_with_localmem(__global int *a, int n, __local int *tmp_sum, __global int *sum)\n" @@ -115,7 +114,7 @@ verify_sum(int *inptr, int *outptr, int n) return 0; } -int test_local_arg_def(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(local_arg_def) { cl_mem streams[2]; cl_program program; @@ -231,7 +230,7 @@ int test_local_arg_def(cl_device_id device, cl_context context, cl_command_queue return err; } -int test_local_kernel_def(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(local_kernel_def) { cl_mem streams[2]; cl_program program; @@ -363,6 +362,3 @@ int test_local_kernel_def(cl_device_id device, cl_context context, cl_command_qu return err; } - - - diff --git a/test_conformance/basic/test_local_kernel_scope.cpp b/test_conformance/basic/test_local_kernel_scope.cpp index 8c23c1c3..5a3613e2 100644 --- a/test_conformance/basic/test_local_kernel_scope.cpp +++ b/test_conformance/basic/test_local_kernel_scope.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" #define MAX_LOCAL_STORAGE_SIZE 256 #define MAX_LOCAL_STORAGE_SIZE_STRING "256" @@ -56,7 +55,7 @@ const char *kernelSource[] = { "}\n" }; -int test_local_kernel_scope(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(local_kernel_scope) { cl_int error; clProgramWrapper program; @@ -138,5 +137,3 @@ int test_local_kernel_scope(cl_device_id device, cl_context context, cl_command_ free(outputData); return 0; } - - diff --git a/test_conformance/basic/test_local_linear_id.cpp b/test_conformance/basic/test_local_linear_id.cpp index 279bd713..5b717c4a 100644 --- a/test_conformance/basic/test_local_linear_id.cpp +++ b/test_conformance/basic/test_local_linear_id.cpp @@ -20,9 +20,9 @@ #include #include #include -#include "harness/rounding_mode.h" -#include "procs.h" +#include "testBase.h" +#include "harness/rounding_mode.h" static const char *local_linear_id_1d_code = "__kernel void test_local_linear_id_1d(global int *dst)\n" @@ -63,8 +63,7 @@ verify_local_linear_id(int *result, int n) } -int -test_local_linear_id(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(local_linear_id, Version(2, 0)) { cl_mem streams; cl_program program[2]; diff --git a/test_conformance/basic/test_loop.cpp b/test_conformance/basic/test_loop.cpp index 1c9acd1a..96f07473 100644 --- a/test_conformance/basic/test_loop.cpp +++ b/test_conformance/basic/test_loop.cpp @@ -23,7 +23,7 @@ #include -#include "procs.h" +#include "testBase.h" namespace { const char *loop_kernel_code = R"( @@ -68,8 +68,7 @@ int verify_loop(std::vector inptr, std::vector loopindx, return 0; } } -int test_loop(cl_device_id device, cl_context context, cl_command_queue queue, - int num_elements) +REGISTER_TEST(loop) { clMemWrapper streams[4]; clProgramWrapper program; diff --git a/test_conformance/basic/test_multireadimagemultifmt.cpp b/test_conformance/basic/test_multireadimagemultifmt.cpp index 7fe58d3c..b92daf88 100644 --- a/test_conformance/basic/test_multireadimagemultifmt.cpp +++ b/test_conformance/basic/test_multireadimagemultifmt.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static const char *multireadimage_kernel_code = "__kernel void test_multireadimage(read_only image2d_t img0, read_only image2d_t img1, \n" @@ -110,8 +109,7 @@ verify_multireadimage(void *image[], float *outptr, int w, int h) } -int -test_mri_multiple(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(mri_multiple) { cl_mem streams[4]; cl_image_format img_format; @@ -229,8 +227,3 @@ test_mri_multiple(cl_device_id device, cl_context context, cl_command_queue queu return err; } - - - - - diff --git a/test_conformance/basic/test_multireadimageonefmt.cpp b/test_conformance/basic/test_multireadimageonefmt.cpp index c230e67a..1d0b5b8d 100644 --- a/test_conformance/basic/test_multireadimageonefmt.cpp +++ b/test_conformance/basic/test_multireadimageonefmt.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static const char *multireadimage_kernel_code = "__kernel void test_multireadimage(int n, int m, sampler_t sampler, \n" @@ -93,7 +92,7 @@ verify_multireadimage(void *image[], int num_images, float *outptr, int w, int h } -int test_mri_one(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(mri_one) { cl_mem streams[8]; cl_image_format img_format; @@ -190,8 +189,3 @@ int test_mri_one(cl_device_id device, cl_context context, cl_command_queue queue return err; } - - - - - diff --git a/test_conformance/basic/test_numeric_constants.cpp b/test_conformance/basic/test_numeric_constants.cpp index 83687ee3..d88c8172 100644 --- a/test_conformance/basic/test_numeric_constants.cpp +++ b/test_conformance/basic/test_numeric_constants.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,7 +13,7 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "procs.h" +#include "testBase.h" #define TEST_VALUE_POSITIVE( string_name, name, value ) \ { \ @@ -55,7 +55,7 @@ log_info("\t" string_name ": " #name " = %a (%17.21g)\n", value, value); \ } \ } -int test_host_numeric_constants(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(host_numeric_constants) { int errors = 0; TEST_VALUE_EQUAL_LITERAL( "CL_CHAR_BIT", CL_CHAR_BIT, 8) @@ -215,7 +215,7 @@ const char *kernel_double[] = { }; -int test_kernel_numeric_constants(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(kernel_numeric_constants) { int error, errors = 0; // clProgramWrapper program; @@ -307,7 +307,7 @@ int test_kernel_numeric_constants(cl_device_id deviceID, cl_context context, cl_ TEST_VALUE_EQUAL( "M_SQRT1_2_F", float_out[15], CL_M_SQRT1_2_F ) // We need to check these values against what we know is supported on the device - if( checkForImageSupport( deviceID ) == 0 ) + if (checkForImageSupport(device) == 0) { // has images // If images are supported, the constant should have been defined to the value 1 if( int_out[18] == 0xf00baa ) @@ -383,7 +383,8 @@ int test_kernel_numeric_constants(cl_device_id deviceID, cl_context context, cl_ /** DOUBLEs **/ - if(!is_extension_available(deviceID, "cl_khr_fp64")) { + if (!is_extension_available(device, "cl_khr_fp64")) + { log_info("Extension cl_khr_fp64 not supported; skipping double tests.\n"); } else @@ -529,7 +530,7 @@ const char *kernel_constant_double_limits[] = { #define TEST_FLOAT_ASSERTION( a, msg, f ) if( !( a ) ) { log_error( "ERROR: Float constant failed requirement: %s (bitwise value is 0x%8.8x)\n", msg, *( (uint32_t *)&f ) ); return -1; } #define TEST_DOUBLE_ASSERTION( a, msg, f ) if( !( a ) ) { log_error( "ERROR: Double constant failed requirement: %s (bitwise value is 0x%16.16llx)\n", msg, *( (uint64_t *)&f ) ); return -1; } -int test_kernel_limit_constants(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(kernel_limit_constants) { int error; size_t threads[] = {1,1,1}; @@ -583,7 +584,8 @@ int test_kernel_limit_constants(cl_device_id deviceID, cl_context context, cl_co // Stage 2: INFINITY and NAN char profileStr[128] = ""; - error = clGetDeviceInfo( deviceID, CL_DEVICE_PROFILE, sizeof( profileStr ), &profileStr, NULL ); + error = clGetDeviceInfo(device, CL_DEVICE_PROFILE, sizeof(profileStr), + &profileStr, NULL); test_error( error, "Unable to run INFINITY/NAN tests (unable to get CL_DEVICE_PROFILE" ); bool testInfNan = true; @@ -591,7 +593,8 @@ int test_kernel_limit_constants(cl_device_id deviceID, cl_context context, cl_co { // We test if we're not an embedded profile, OR if the inf/nan flag in the config is set cl_device_fp_config single = 0; - error = clGetDeviceInfo( deviceID, CL_DEVICE_SINGLE_FP_CONFIG, sizeof( single ), &single, NULL ); + error = clGetDeviceInfo(device, CL_DEVICE_SINGLE_FP_CONFIG, + sizeof(single), &single, NULL); test_error( error, "Unable to run INFINITY/NAN tests (unable to get FP_CONFIG bits)" ); if( ( single & CL_FP_INF_NAN ) == 0 ) @@ -666,12 +669,13 @@ int test_kernel_limit_constants(cl_device_id deviceID, cl_context context, cl_co } // Stage 3: limits on HUGE_VAL (double) - if( !is_extension_available( deviceID, "cl_khr_fp64" ) ) + if (!is_extension_available(device, "cl_khr_fp64")) log_info( "Note: Skipping double HUGE_VAL tests (doubles unsupported on device)\n" ); else { cl_device_fp_config config = 0; - error = clGetDeviceInfo( deviceID, CL_DEVICE_DOUBLE_FP_CONFIG, sizeof( config ), &config, NULL ); + error = clGetDeviceInfo(device, CL_DEVICE_DOUBLE_FP_CONFIG, + sizeof(config), &config, NULL); test_error( error, "Unable to run INFINITY/NAN tests (unable to get double FP_CONFIG bits)" ); if( ( config & CL_FP_INF_NAN ) == 0 ) @@ -716,5 +720,3 @@ int test_kernel_limit_constants(cl_device_id deviceID, cl_context context, cl_co return 0; } - - diff --git a/test_conformance/basic/test_pointercast.cpp b/test_conformance/basic/test_pointercast.cpp index a3993d52..e738c087 100644 --- a/test_conformance/basic/test_pointercast.cpp +++ b/test_conformance/basic/test_pointercast.cpp @@ -21,8 +21,7 @@ #include #include - -#include "procs.h" +#include "testBase.h" static const char *pointer_cast_kernel_code = "__kernel void test_pointer_cast(__global unsigned char *src, __global unsigned int *dst)\n" @@ -57,7 +56,7 @@ verify_pointer_cast(unsigned char *inptr, unsigned int *outptr, int n) return 0; } -int test_pointer_cast(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(pointer_cast) { cl_mem streams[2]; unsigned char *input_ptr; @@ -137,5 +136,3 @@ int test_pointer_cast(cl_device_id device, cl_context context, cl_command_queue return err; } - - diff --git a/test_conformance/basic/test_preprocessors.cpp b/test_conformance/basic/test_preprocessors.cpp index 994b62f2..610bd14c 100644 --- a/test_conformance/basic/test_preprocessors.cpp +++ b/test_conformance/basic/test_preprocessors.cpp @@ -1,6 +1,6 @@ // // Copyright (c) 2017 The Khronos Group Inc. -// +// // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. // You may obtain a copy of the License at @@ -13,9 +13,10 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "procs.h" #include +#include "testBase.h" + // Test __FILE__, __LINE__, __OPENCL_VERSION__, __OPENCL_C_VERSION__, __ENDIAN_LITTLE__, __ROUNDING_MODE__, __IMAGE_SUPPORT__, __FAST_RELAXED_MATH__ // __kernel_exec @@ -83,7 +84,7 @@ const char *preprocessor_test = { "}\n" }; -int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(kernel_preprocessor_macros) { clProgramWrapper program; clKernelWrapper kernel; @@ -156,7 +157,7 @@ int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, c /////// Check the integer results // We need to check these values against what we know is supported on the device - if( checkForImageSupport( deviceID ) == 0 ) + if (checkForImageSupport(device) == 0) { // If images are supported, the constant should have been defined to the value 1 if( results[ 0 ] == 0xf00baa ) @@ -182,7 +183,9 @@ int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, c // __ENDIAN_LITTLE__ is similar to __IMAGE_SUPPORT__: 1 if it's true, undefined if it isn't cl_bool deviceIsLittleEndian; - error = clGetDeviceInfo( deviceID, CL_DEVICE_ENDIAN_LITTLE, sizeof( deviceIsLittleEndian ), &deviceIsLittleEndian, NULL ); + error = clGetDeviceInfo(device, CL_DEVICE_ENDIAN_LITTLE, + sizeof(deviceIsLittleEndian), &deviceIsLittleEndian, + NULL); test_error( error, "Unable to get endian property of device to validate against" ); if( deviceIsLittleEndian ) @@ -216,7 +219,7 @@ int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, c // The OpenCL version reported by the macro reports the feature level supported by the compiler. Since // this doesn't directly match any property we can query, we just check to see if it's a sane value - auto device_cl_version = get_device_cl_version(deviceID); + auto device_cl_version = get_device_cl_version(device); int device_cl_version_int = device_cl_version.to_uint() * 10; if ((results[2] < 100) || (results[2] > device_cl_version_int)) { @@ -241,11 +244,11 @@ int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, c int cl_c_minor_version = (results[3] / 10) % 10; if ((results[3] < 100) || (!device_supports_cl_c_version( - deviceID, + device, Version{ (cl_uint)cl_c_major_version, (cl_uint)cl_c_minor_version }))) { - auto device_version = get_device_cl_c_version(deviceID); + auto device_version = get_device_cl_c_version(device); log_error( "ERROR: Kernel preprocessor __OPENCL_C_VERSION__ does not make " "sense w.r.t. device's version string! " @@ -337,7 +340,7 @@ int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, c test_error( error, "Unable to create test program" ); // Try compiling - error = clBuildProgram( programB, 1, &deviceID, "-cl-fast-relaxed-math", NULL, NULL ); + error = clBuildProgram( programB, 1, &device, "-cl-fast-relaxed-math", NULL, NULL ); test_error( error, "Unable to build program" ); // Create a kernel again to run against @@ -373,4 +376,3 @@ int test_kernel_preprocessor_macros(cl_device_id deviceID, cl_context context, c return 0; } - diff --git a/test_conformance/basic/test_progvar.cpp b/test_conformance/basic/test_progvar.cpp index a18925c3..15b4df43 100644 --- a/test_conformance/basic/test_progvar.cpp +++ b/test_conformance/basic/test_progvar.cpp @@ -50,8 +50,6 @@ #include "harness/errorHelpers.h" #include "harness/featureHelpers.h" #include "harness/mt19937.h" -#include "procs.h" - //////////////////// // Device capabilities @@ -2006,8 +2004,7 @@ static cl_int should_skip(cl_device_id device, cl_bool& skip) // Test support for variables at program scope. Miscellaneous -int test_progvar_prog_scope_misc(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(progvar_prog_scope_misc, Version(2, 0)) { cl_bool skip{ CL_FALSE }; auto error = should_skip(device, skip); @@ -2038,8 +2035,7 @@ int test_progvar_prog_scope_misc(cl_device_id device, cl_context context, // Test support for variables at program scope. Unitialized data -int test_progvar_prog_scope_uninit(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(progvar_prog_scope_uninit, Version(2, 0)) { cl_bool skip{ CL_FALSE }; auto error = should_skip(device, skip); @@ -2068,8 +2064,7 @@ int test_progvar_prog_scope_uninit(cl_device_id device, cl_context context, } // Test support for variables at program scope. Initialized data. -int test_progvar_prog_scope_init(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(progvar_prog_scope_init, Version(2, 0)) { cl_bool skip{ CL_FALSE }; auto error = should_skip(device, skip); @@ -2098,8 +2093,7 @@ int test_progvar_prog_scope_init(cl_device_id device, cl_context context, // A simple test for support of static variables inside a kernel. -int test_progvar_func_scope(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST_VERSION(progvar_func_scope, Version(2, 0)) { cl_bool skip{ CL_FALSE }; auto error = should_skip(device, skip); diff --git a/test_conformance/basic/test_queue_priority.cpp b/test_conformance/basic/test_queue_priority.cpp index ff6283cd..2eb0dd2d 100644 --- a/test_conformance/basic/test_queue_priority.cpp +++ b/test_conformance/basic/test_queue_priority.cpp @@ -20,9 +20,9 @@ #include #include #include -#include "harness/rounding_mode.h" -#include "procs.h" +#include "testBase.h" +#include "harness/rounding_mode.h" static const char *fpadd_kernel_code = "__kernel void test_fpadd(__global float *srcA, __global float *srcB, __global float *dst)\n" @@ -131,7 +131,7 @@ verify_fpmul(float *inptrA, float *inptrB, float *outptr, int n, int fileNum) #if defined( __APPLE__ ) -int test_queue_priority(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(queue_priority) { int err; int command_queue_priority = 0; @@ -342,6 +342,5 @@ int test_queue_priority(cl_device_id device, cl_context context, cl_command_queu } - #endif diff --git a/test_conformance/basic/test_readimage.cpp b/test_conformance/basic/test_readimage.cpp index 0aa70525..c23888e6 100644 --- a/test_conformance/basic/test_readimage.cpp +++ b/test_conformance/basic/test_readimage.cpp @@ -26,7 +26,7 @@ #include #include -#include "procs.h" +#include "testBase.h" #define TEST_IMAGE_WIDTH_2D (512) #define TEST_IMAGE_HEIGHT_2D (512) @@ -204,8 +204,9 @@ cl_mem create_image_xd(cl_context context, cl_mem_flags flags, } template -int test_readimage(cl_device_id device, cl_context context, - cl_command_queue queue, const cl_image_format *img_format) +static int test_readimage(cl_device_id device, cl_context context, + cl_command_queue queue, + const cl_image_format *img_format) { clMemWrapper streams[2]; clProgramWrapper program; @@ -301,8 +302,7 @@ bool check_format(cl_device_id device, cl_context context, } } -int test_readimage(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(readimage) { const cl_image_format format[] = { { CL_RGBA, CL_UNORM_INT8 }, { CL_BGRA, CL_UNORM_INT8 } }; @@ -319,24 +319,21 @@ int test_readimage(cl_device_id device, cl_context context, return err; } -int test_readimage_int16(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(readimage_int16) { const cl_image_format format = { CL_RGBA, CL_UNORM_INT16 }; return test_readimage(device, context, queue, &format); } -int test_readimage_fp32(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(readimage_fp32) { const cl_image_format format = { CL_RGBA, CL_FLOAT }; return test_readimage(device, context, queue, &format); } -int test_readimage3d(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(readimage3d) { const cl_image_format format[] = { { CL_RGBA, CL_UNORM_INT8 }, { CL_BGRA, CL_UNORM_INT8 } }; @@ -355,8 +352,7 @@ int test_readimage3d(cl_device_id device, cl_context context, return err; } -int test_readimage3d_int16(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(readimage3d_int16) { const cl_image_format format = { CL_RGBA, CL_UNORM_INT16 }; @@ -365,8 +361,7 @@ int test_readimage3d_int16(cl_device_id device, cl_context context, return test_readimage(device, context, queue, &format); } -int test_readimage3d_fp32(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(readimage3d_fp32) { const cl_image_format format = { CL_RGBA, CL_FLOAT }; diff --git a/test_conformance/basic/test_rw_image_access_qualifier.cpp b/test_conformance/basic/test_rw_image_access_qualifier.cpp index b06c82a5..b8dc1630 100644 --- a/test_conformance/basic/test_rw_image_access_qualifier.cpp +++ b/test_conformance/basic/test_rw_image_access_qualifier.cpp @@ -18,7 +18,7 @@ #include #include -#include "procs.h" +#include "testBase.h" #include "harness/clImageHelper.h" static const char* rw_kernel_code = @@ -41,10 +41,10 @@ static const char* rw_kernel_code = "}\n"; -int test_rw_image_access_qualifier(cl_device_id device_id, cl_context context, cl_command_queue commands, int num_elements) +REGISTER_TEST_VERSION(rw_image_access_qualifier, Version(2, 0)) { // This test should be skipped if images are not supported. - if (checkForImageSupport(device_id)) + if (checkForImageSupport(device)) { return TEST_SKIPPED_ITSELF; } @@ -53,11 +53,11 @@ int test_rw_image_access_qualifier(cl_device_id device_id, cl_context context, c // or 2.X device if the device supports images. In OpenCL-3.0 // read-write images are optional. This test is already being skipped // for 1.X devices. - if (get_device_cl_version(device_id) >= Version(3, 0)) + if (get_device_cl_version(device) >= Version(3, 0)) { cl_uint are_rw_images_supported{}; test_error( - clGetDeviceInfo(device_id, CL_DEVICE_MAX_READ_WRITE_IMAGE_ARGS, + clGetDeviceInfo(device, CL_DEVICE_MAX_READ_WRITE_IMAGE_ARGS, sizeof(are_rw_images_supported), &are_rw_images_supported, nullptr), "clGetDeviceInfo failed for CL_DEVICE_MAX_READ_IMAGE_ARGS\n"); @@ -155,15 +155,14 @@ int test_rw_image_access_qualifier(cl_device_id device_id, cl_context context, c err = CL_SUCCESS; unsigned int num_iter = 1; for(i = 0; i < num_iter; i++) { - err |= clEnqueueNDRangeKernel(commands, kernel, dim_count, - NULL, global_dim, local_dim, - 0, NULL, NULL); + err |= clEnqueueNDRangeKernel(queue, kernel, dim_count, NULL, + global_dim, local_dim, 0, NULL, NULL); } /* Read back the results from the device to verify the output */ const size_t origin[3] = {0, 0, 0}; const size_t region[3] = {size_x, size_y, 1}; - err |= clEnqueueReadImage(commands, src_image, CL_TRUE, origin, region, 0, 0, + err |= clEnqueueReadImage(queue, src_image, CL_TRUE, origin, region, 0, 0, output, 0, NULL, NULL); if (err != CL_SUCCESS) { log_error("Error: clEnqueueReadBuffer failed\n"); diff --git a/test_conformance/basic/test_simple_image_pitch.cpp b/test_conformance/basic/test_simple_image_pitch.cpp index 83df1699..facf0edd 100644 --- a/test_conformance/basic/test_simple_image_pitch.cpp +++ b/test_conformance/basic/test_simple_image_pitch.cpp @@ -21,9 +21,9 @@ #include #include -#include "procs.h" +#include "testBase.h" -int test_simple_read_image_pitch(cl_device_id device, cl_context cl_context_, cl_command_queue q, int num_elements) +REGISTER_TEST(simple_read_image_pitch) { cl_int err = CL_SUCCESS; @@ -51,7 +51,9 @@ int test_simple_read_image_pitch(cl_device_id device, cl_context cl_context_, cl desc.image_width = imageW; desc.image_height = imageH; - cl_mem image = clCreateImage(cl_context_, CL_MEM_COPY_HOST_PTR|CL_MEM_READ_WRITE, &fmt, &desc, host_image, &err); + cl_mem image = + clCreateImage(context, CL_MEM_COPY_HOST_PTR | CL_MEM_READ_WRITE, &fmt, + &desc, host_image, &err); test_error(err,"clCreateImage"); char* host_buffer = (char*)malloc(buffer_bytes); @@ -61,7 +63,8 @@ int test_simple_read_image_pitch(cl_device_id device, cl_context cl_context_, cl size_t origin[] = { 0, 0, 0 }; size_t region[] = { imageW, imageH, 1 }; - err = clEnqueueReadImage(q, image, CL_TRUE, origin, region, bufferW, 0, host_buffer, 0, NULL, NULL); + err = clEnqueueReadImage(queue, image, CL_TRUE, origin, region, bufferW, 0, + host_buffer, 0, NULL, NULL); test_error(err,"clEnqueueReadImage"); size_t errors = 0; @@ -88,7 +91,7 @@ int test_simple_read_image_pitch(cl_device_id device, cl_context cl_context_, cl return errors == 0 ? TEST_PASS : TEST_FAIL; } -int test_simple_write_image_pitch(cl_device_id device, cl_context cl_context_, cl_command_queue q, int num_elements) +REGISTER_TEST(simple_write_image_pitch) { cl_int err = CL_SUCCESS; @@ -116,7 +119,9 @@ int test_simple_write_image_pitch(cl_device_id device, cl_context cl_context_, c desc.image_width = imageW; desc.image_height = imageH; - cl_mem image = clCreateImage(cl_context_, CL_MEM_COPY_HOST_PTR|CL_MEM_READ_WRITE, &fmt, &desc, host_image, &err); + cl_mem image = + clCreateImage(context, CL_MEM_COPY_HOST_PTR | CL_MEM_READ_WRITE, &fmt, + &desc, host_image, &err); test_error(err,"clCreateImage"); char* host_buffer = (char*)malloc(buffer_bytes); @@ -126,11 +131,14 @@ int test_simple_write_image_pitch(cl_device_id device, cl_context cl_context_, c size_t origin[] = { 0, 0, 0 }; size_t region[] = { imageW, imageH, 1 }; - err = clEnqueueWriteImage(q, image, CL_TRUE, origin, region, bufferW, 0, host_buffer, 0, NULL, NULL); + err = clEnqueueWriteImage(queue, image, CL_TRUE, origin, region, bufferW, 0, + host_buffer, 0, NULL, NULL); test_error(err,"clEnqueueWriteImage"); size_t mapped_pitch = 0; - char* mapped_image = (char*)clEnqueueMapImage(q, image, CL_TRUE, CL_MAP_READ, origin, region, &mapped_pitch, NULL, 0, NULL, NULL, &err); + char* mapped_image = (char*)clEnqueueMapImage( + queue, image, CL_TRUE, CL_MAP_READ, origin, region, &mapped_pitch, NULL, + 0, NULL, NULL, &err); test_error(err,"clEnqueueMapImage"); size_t errors = 0; @@ -145,7 +153,7 @@ int test_simple_write_image_pitch(cl_device_id device, cl_context cl_context_, c } } - err = clEnqueueUnmapMemObject(q, image, (void *)mapped_image, 0, 0, 0); + err = clEnqueueUnmapMemObject(queue, image, (void*)mapped_image, 0, 0, 0); test_error(err,"clEnqueueUnmapMemObject"); test_error(clReleaseMemObject(image),"clReleaseMemObject"); diff --git a/test_conformance/basic/test_sizeof.cpp b/test_conformance/basic/test_sizeof.cpp index b6ea89f7..8827727e 100644 --- a/test_conformance/basic/test_sizeof.cpp +++ b/test_conformance/basic/test_sizeof.cpp @@ -20,9 +20,8 @@ #include #include #include -#include "procs.h" - +#include "testBase.h" cl_int get_type_size( cl_context context, cl_command_queue queue, const char *type, cl_ulong *size, cl_device_id device ) { @@ -127,7 +126,7 @@ const char *other_types[] = static int IsPowerOfTwo( cl_ulong x ){ return 0 == (x & (x-1)); } -int test_sizeof(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(sizeof) { size_t i, j; cl_ulong test; @@ -376,5 +375,3 @@ int test_sizeof(cl_device_id device, cl_context context, cl_command_queue queue, return err; } - - diff --git a/test_conformance/basic/test_vec_type_hint.cpp b/test_conformance/basic/test_vec_type_hint.cpp index 0ba105db..f2fced67 100644 --- a/test_conformance/basic/test_vec_type_hint.cpp +++ b/test_conformance/basic/test_vec_type_hint.cpp @@ -21,7 +21,6 @@ #include #include -#include "procs.h" #include "harness/conversions.h" #include "harness/typeWrappers.h" @@ -36,7 +35,7 @@ static const char *sample_kernel = { "}\n" }; -int test_vec_type_hint(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements) +REGISTER_TEST(vec_type_hint) { int error; int vec_type_index, vec_size_index; @@ -51,14 +50,14 @@ int test_vec_type_hint(cl_device_id deviceID, cl_context context, cl_command_que { if (vecType[vec_type_index] == kHalf - && !is_extension_available(deviceID, "cl_khr_fp16")) + && !is_extension_available(device, "cl_khr_fp16")) { log_info( "Extension cl_khr_fp16 not supported; skipping half tests.\n"); continue; } else if (vecType[vec_type_index] == kDouble - && !is_extension_available(deviceID, "cl_khr_fp64")) + && !is_extension_available(device, "cl_khr_fp64")) { log_info( "Extension cl_khr_fp64 not supported; skipping double tests.\n"); diff --git a/test_conformance/basic/test_vector_creation.cpp b/test_conformance/basic/test_vector_creation.cpp index 79c97f7d..641823a0 100644 --- a/test_conformance/basic/test_vector_creation.cpp +++ b/test_conformance/basic/test_vector_creation.cpp @@ -13,7 +13,6 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "procs.h" #include "harness/conversions.h" #include "harness/typeWrappers.h" #include "harness/errorHelpers.h" @@ -244,8 +243,7 @@ int create_kernel(ExplicitType type, int output_size, char *program, } -int test_vector_creation(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(vector_creation) { const std::vector vecType = { kChar, kUChar, kShort, kUShort, kInt, kUInt, kLong, kULong, @@ -274,7 +272,7 @@ int test_vector_creation(cl_device_id deviceID, cl_context context, } else if (vecType[type_index] == kDouble) { - if (!is_extension_available(deviceID, "cl_khr_fp64")) + if (!is_extension_available(device, "cl_khr_fp64")) { log_info("Extension cl_khr_fp64 not supported; skipping double " "tests.\n"); @@ -285,7 +283,7 @@ int test_vector_creation(cl_device_id deviceID, cl_context context, } else if (vecType[type_index] == kHalf) { - if (!is_extension_available(deviceID, "cl_khr_fp16")) + if (!is_extension_available(device, "cl_khr_fp16")) { log_info("Extension cl_khr_fp16 not supported; skipping half " "tests.\n"); diff --git a/test_conformance/basic/test_vector_swizzle.cpp b/test_conformance/basic/test_vector_swizzle.cpp index 6bdf651d..ad3a3959 100644 --- a/test_conformance/basic/test_vector_swizzle.cpp +++ b/test_conformance/basic/test_vector_swizzle.cpp @@ -19,8 +19,7 @@ #include #include -#include "procs.h" -#include "harness/testHarness.h" +#include "testBase.h" static std::string pragma_extension; @@ -692,8 +691,7 @@ static int test_type(const char* type_name, cl_device_id device, | test_vectype(type_name, device, context, queue); } -int test_vector_swizzle(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(vector_swizzle) { int hasDouble = is_extension_available(device, "cl_khr_fp64"); int hasHalf = is_extension_available(device, "cl_khr_fp16"); diff --git a/test_conformance/basic/test_vloadstore.cpp b/test_conformance/basic/test_vloadstore.cpp index e076c081..e66c9e56 100644 --- a/test_conformance/basic/test_vloadstore.cpp +++ b/test_conformance/basic/test_vloadstore.cpp @@ -23,8 +23,6 @@ #include #include - -#include "procs.h" #include "harness/conversions.h" #include "harness/errorHelpers.h" #include "harness/stringHelpers.h" @@ -111,9 +109,10 @@ typedef void (*create_program_fn)(std::string &, size_t, ExplicitType, size_t, typedef int (*test_fn)(cl_device_id, cl_context, cl_command_queue, ExplicitType, unsigned int, create_program_fn, size_t); -int test_vload(cl_device_id device, cl_context context, cl_command_queue queue, - ExplicitType type, unsigned int vecSize, - create_program_fn createFn, size_t bufferSize) +static int test_vload(cl_device_id device, cl_context context, + cl_command_queue queue, ExplicitType type, + unsigned int vecSize, create_program_fn createFn, + size_t bufferSize) { clProgramWrapper program; clKernelWrapper kernel; @@ -278,8 +277,9 @@ int test_vload(cl_device_id device, cl_context context, cl_command_queue queue, } template -int test_vset(cl_device_id device, cl_context context, cl_command_queue queue, - create_program_fn createFn, size_t bufferSize) +static int test_vset(cl_device_id device, cl_context context, + cl_command_queue queue, create_program_fn createFn, + size_t bufferSize) { std::vector vecType = { kChar, kUChar, kShort, kUShort, kInt, kUInt, kLong, kULong, @@ -342,7 +342,7 @@ void create_global_load_code(std::string &destBuffer, size_t inBufferSize, typeName, (int)inVectorSize, (int)inVectorSize); } -int test_vload_global(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vload_global) { return test_vset(device, context, queue, create_global_load_code, 10240); @@ -372,7 +372,7 @@ void create_local_load_code(std::string &destBuffer, size_t inBufferSize, (int)inVectorSize, (int)inVectorSize, typeName); } -int test_vload_local(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vload_local) { // Determine the max size of a local buffer that we can test against cl_ulong localSize; @@ -408,7 +408,7 @@ void create_constant_load_code(std::string &destBuffer, size_t inBufferSize, typeName, (int)inVectorSize, (int)inVectorSize); } -int test_vload_constant(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vload_constant) { // Determine the max size of a local buffer that we can test against cl_ulong maxSize; @@ -448,7 +448,7 @@ void create_private_load_code(std::string &destBuffer, size_t inBufferSize, (int)inVectorSize, (int)inVectorSize, typeName); } -int test_vload_private(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vload_private) { // We have no idea how much actual private storage is available, so just pick a reasonable value, // which is that we can fit at least two 16-element long, which is 2*8 bytes * 16 = 256 bytes @@ -459,9 +459,10 @@ int test_vload_private(cl_device_id device, cl_context context, cl_command_queue /////////////////////////////////////////////////////////////////////////////////////////////////////////////////////// #pragma mark -------------------- vstore harness -------------------------- -int test_vstore(cl_device_id device, cl_context context, cl_command_queue queue, - ExplicitType type, unsigned int vecSize, - create_program_fn createFn, size_t bufferSize) +static int test_vstore(cl_device_id device, cl_context context, + cl_command_queue queue, ExplicitType type, + unsigned int vecSize, create_program_fn createFn, + size_t bufferSize) { clProgramWrapper program; clKernelWrapper kernel; @@ -716,7 +717,7 @@ void create_global_store_code(std::string &destBuffer, size_t inBufferSize, } } -int test_vstore_global(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vstore_global) { return test_vset(device, context, queue, create_global_store_code, 10240); @@ -805,7 +806,7 @@ void create_local_store_code(std::string &destBuffer, size_t inBufferSize, } } -int test_vstore_local(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vstore_local) { // Determine the max size of a local buffer that we can test against cl_ulong localSize; @@ -890,13 +891,10 @@ void create_private_store_code(std::string &destBuffer, size_t inBufferSize, } } -int test_vstore_private(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems ) +REGISTER_TEST(vstore_private) { // We have no idea how much actual private storage is available, so just pick a reasonable value, // which is that we can fit at least two 16-element long, which is 2*8 bytes * 16 = 256 bytes return test_vset(device, context, queue, create_private_store_code, 256); } - - - diff --git a/test_conformance/basic/test_work_item_functions.cpp b/test_conformance/basic/test_work_item_functions.cpp index ce427dc3..046640b3 100644 --- a/test_conformance/basic/test_work_item_functions.cpp +++ b/test_conformance/basic/test_work_item_functions.cpp @@ -23,7 +23,6 @@ #include #include -#include "procs.h" #include "harness/conversions.h" #include "harness/typeWrappers.h" @@ -551,29 +550,22 @@ struct TestWorkItemFnsOutOfRange } // anonymous namespace -int test_work_item_functions(cl_device_id deviceID, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(work_item_functions) { - TestWorkItemFns fnct(deviceID, context, queue); + TestWorkItemFns fnct(device, context, queue); return fnct.Run(); } -int test_work_item_functions_out_of_range(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(work_item_functions_out_of_range) { - TestWorkItemFnsOutOfRange fnct(deviceID, context, queue, + TestWorkItemFnsOutOfRange fnct(device, context, queue, outOfRangeWorkItemKernelCode); return fnct.Run(); } -int test_work_item_functions_out_of_range_hardcoded(cl_device_id deviceID, - cl_context context, - cl_command_queue queue, - int num_elements) +REGISTER_TEST(work_item_functions_out_of_range_hardcoded) { - TestWorkItemFnsOutOfRange fnct(deviceID, context, queue, + TestWorkItemFnsOutOfRange fnct(device, context, queue, outOfRangeWorkItemHardcodedKernelCode); return fnct.Run(); } diff --git a/test_conformance/basic/test_writeimage.cpp b/test_conformance/basic/test_writeimage.cpp index a2847e27..b509e54c 100644 --- a/test_conformance/basic/test_writeimage.cpp +++ b/test_conformance/basic/test_writeimage.cpp @@ -22,14 +22,11 @@ #include #include - -#include "procs.h" - #include #include #include -#include "procs.h" +#include "testBase.h" namespace { const char *kernel_source = R"( @@ -120,9 +117,9 @@ const char *get_mem_flag_name(cl_mem_flags flags) } template -int test_writeimage(cl_device_id device, cl_context context, - cl_command_queue queue, const cl_image_format *img_format, - cl_mem_flags img_flags) +static int +test_writeimage(cl_device_id device, cl_context context, cl_command_queue queue, + const cl_image_format *img_format, cl_mem_flags img_flags) { clMemWrapper streams[2]; clProgramWrapper program; @@ -207,8 +204,7 @@ bool check_format(cl_device_id device, cl_context context, &img_format); } } -int test_writeimage(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(writeimage) { int err = 0; const cl_image_format format[] = { { CL_RGBA, CL_UNORM_INT8 }, @@ -230,8 +226,7 @@ int test_writeimage(cl_device_id device, cl_context context, return err; } -int test_writeimage_int16(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(writeimage_int16) { int err = 0; const cl_image_format format = { CL_RGBA, CL_UNORM_INT16 }; @@ -245,8 +240,7 @@ int test_writeimage_int16(cl_device_id device, cl_context context, return err; } -int test_writeimage_fp32(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) +REGISTER_TEST(writeimage_fp32) { int err = 0; const cl_image_format format = { CL_RGBA, CL_FLOAT };