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