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