ahmed/OpenCL-CTS
1
0
Fork 0
mirror of https://github.com/KhronosGroup/OpenCL-CTS.git synced 2026-03-19 06:09:01 +00:00
Code Issues Packages Projects Releases Wiki Activity

Merge branch 'main' into cl_khr_unified_svm

Browse Source
This commit is contained in:
Ben Ashbaugh
2025-03-16 14:12:26 -07:00
parent e96b510aa0 bb604702a5
commit debc466d40
57 changed files with 460 additions and 1244 deletions
Download Patch File Download Diff File Expand all files Collapse all files

1
CMakeLists.txt
View File

@@ -18,6 +18,7 @@ add_definitions(-DCL_USE_DEPRECATED_OPENCL_1_2_APIS=1)
add_definitions(-DCL_USE_DEPRECATED_OPENCL_1_1_APIS=1)
add_definitions(-DCL_USE_DEPRECATED_OPENCL_1_0_APIS=1)
add_definitions(-DCL_NO_EXTENSION_PROTOTYPES)
add_definitions(-DCL_ENABLE_BETA_EXTENSIONS)
option(USE_CL_EXPERIMENTAL "Use Experimental definitions" OFF)
if(USE_CL_EXPERIMENTAL)

5
test_conformance/computeinfo/conforming_version.cpp
View File

@@ -18,11 +18,10 @@
#include "harness/testHarness.h"
#include "harness/deviceInfo.h"
int test_conformance_version(cl_device_id deviceID, cl_context context,
cl_command_queue ignoreQueue, int num_elements)
REGISTER_TEST_VERSION(conformance_version, Version(3, 0))
{
std::string version_string{ get_device_info_string(
deviceID, CL_DEVICE_LATEST_CONFORMANCE_VERSION_PASSED) };
device, CL_DEVICE_LATEST_CONFORMANCE_VERSION_PASSED) };
// Latest conformance version passed should match vYYYY-MM-DD-XX, where XX
// is a number

15
test_conformance/computeinfo/device_uuid.cpp
View File

@@ -99,10 +99,9 @@ static bool get_uuid(const cl_device_id device, const cl_device_info info,
return true;
}
int test_device_uuid(cl_device_id deviceID, cl_context context,
cl_command_queue ignoreQueue, int num_elements)
REGISTER_TEST(device_uuid)
{
if (!is_extension_available(deviceID, "cl_khr_device_uuid"))
if (!is_extension_available(device, "cl_khr_device_uuid"))
{
log_info("cl_khr_device_uuid not supported. Skipping test...\n");
return TEST_SKIPPED_ITSELF;
@@ -112,7 +111,7 @@ int test_device_uuid(cl_device_id deviceID, cl_context context,
/* CL_DEVICE_UUID_KHR */
uuid device_uuid;
bool success = get_uuid(deviceID, CL_DEVICE_UUID_KHR, device_uuid);
bool success = get_uuid(device, CL_DEVICE_UUID_KHR, device_uuid);
if (!success)
{
log_error("Error getting device UUID\n");
@@ -127,7 +126,7 @@ int test_device_uuid(cl_device_id deviceID, cl_context context,
/* CL_DRIVER_UUID_KHR */
uuid driver_uuid;
success = get_uuid(deviceID, CL_DRIVER_UUID_KHR, driver_uuid);
success = get_uuid(device, CL_DRIVER_UUID_KHR, driver_uuid);
if (!success)
{
log_error("Error getting driver UUID\n");
@@ -144,7 +143,7 @@ int test_device_uuid(cl_device_id deviceID, cl_context context,
/* CL_DEVICE_LUID_VALID_KHR */
cl_bool device_luid_valid{};
cl_int err = clGetDeviceInfo(deviceID, CL_DEVICE_LUID_VALID_KHR,
cl_int err = clGetDeviceInfo(device, CL_DEVICE_LUID_VALID_KHR,
sizeof(device_luid_valid), &device_luid_valid,
&size_ret);
if (!check_device_info_returns(err, size_ret, sizeof(device_luid_valid)))
@@ -162,7 +161,7 @@ int test_device_uuid(cl_device_id deviceID, cl_context context,
/* CL_DEVICE_LUID_KHR */
luid device_luid;
success =
get_uuid(deviceID, CL_DEVICE_LUID_KHR, device_luid, device_luid_valid);
get_uuid(device, CL_DEVICE_LUID_KHR, device_luid, device_luid_valid);
if (!success)
{
log_error("Error getting device LUID\n");
@@ -178,7 +177,7 @@ int test_device_uuid(cl_device_id deviceID, cl_context context,
/* CL_DEVICE_NODE_MASK_KHR */
cl_uint device_node_mask{};
err =
clGetDeviceInfo(deviceID, CL_DEVICE_NODE_MASK_KHR,
clGetDeviceInfo(device, CL_DEVICE_NODE_MASK_KHR,
sizeof(device_node_mask), &device_node_mask, &size_ret);
if (!check_device_info_returns(err, size_ret, sizeof(device_node_mask)))
{

17
test_conformance/computeinfo/extended_versioning.cpp
View File

@@ -724,11 +724,10 @@ static_assert(sizeof(cl_name_version) == sizeof(cl_name_version_khr),
static_assert(CL_MAKE_VERSION(1, 2, 3) == CL_MAKE_VERSION_KHR(1, 2, 3),
"CL_MAKE_VERSION mismatch");
int test_extended_versioning(cl_device_id deviceID, cl_context context,
cl_command_queue ignoreQueue, int num_elements)
REGISTER_TEST(extended_versioning)
{
bool ext = is_extension_available(deviceID, "cl_khr_extended_versioning");
bool core = get_device_cl_version(deviceID) >= Version(3, 0);
bool ext = is_extension_available(device, "cl_khr_extended_versioning");
bool core = get_device_cl_version(device) >= Version(3, 0);
if (!ext && !core)
{
@@ -736,17 +735,17 @@ int test_extended_versioning(cl_device_id deviceID, cl_context context,
}
cl_platform_id platform;
cl_int err = clGetDeviceInfo(deviceID, CL_DEVICE_PLATFORM, sizeof(platform),
cl_int err = clGetDeviceInfo(device, CL_DEVICE_PLATFORM, sizeof(platform),
&platform, nullptr);
test_error(err, "clGetDeviceInfo failed\n");
int total_errors = 0;
total_errors += test_extended_versioning_platform_version(platform);
total_errors += test_extended_versioning_platform_extensions(platform);
total_errors += test_extended_versioning_device_versions(ext, deviceID);
total_errors += test_extended_versioning_device_extensions(deviceID);
total_errors += test_extended_versioning_device_il(deviceID);
total_errors += test_extended_versioning_device_built_in_kernels(deviceID);
total_errors += test_extended_versioning_device_versions(ext, device);
total_errors += test_extended_versioning_device_extensions(device);
total_errors += test_extended_versioning_device_il(device);
total_errors += test_extended_versioning_device_built_in_kernels(device);
return total_errors;
}

24
test_conformance/computeinfo/main.cpp
View File

@@ -1251,8 +1251,7 @@ int getPlatformCapabilities(cl_platform_id platform)
return total_errors;
}
int test_computeinfo(cl_device_id deviceID, cl_context context,
cl_command_queue ignoreQueue, int num_elements)
REGISTER_TEST(computeinfo)
{
int err;
int total_errors = 0;
@@ -1411,23 +1410,6 @@ int test_computeinfo(cl_device_id deviceID, cl_context context,
return total_errors;
}
extern int test_extended_versioning(cl_device_id, cl_context, cl_command_queue,
int);
extern int test_device_uuid(cl_device_id, cl_context, cl_command_queue, int);
extern int test_conformance_version(cl_device_id, cl_context, cl_command_queue,
int);
extern int test_pci_bus_info(cl_device_id, cl_context, cl_command_queue, int);
test_definition test_list[] = {
ADD_TEST(computeinfo),
ADD_TEST(extended_versioning),
ADD_TEST(device_uuid),
ADD_TEST_VERSION(conformance_version, Version(3, 0)),
ADD_TEST(pci_bus_info),
};
const int test_num = ARRAY_SIZE(test_list);
int main(int argc, const char** argv)
{
const char** argList = (const char**)calloc(argc, sizeof(char*));
@@ -1453,7 +1435,9 @@ int main(int argc, const char** argv)
}
}
int error = runTestHarness(argCount, argList, test_num, test_list, true, 0);
int error = runTestHarness(
argCount, argList, test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), true, 0);
free(argList);

11
test_conformance/computeinfo/pci_bus_info.cpp
View File

@@ -21,10 +21,9 @@
#include "harness/testHarness.h"
#include "harness/deviceInfo.h"
int test_pci_bus_info(cl_device_id deviceID, cl_context context,
cl_command_queue ignoreQueue, int num_elements)
REGISTER_TEST(pci_bus_info)
{
if (!is_extension_available(deviceID, "cl_khr_pci_bus_info"))
if (!is_extension_available(device, "cl_khr_pci_bus_info"))
{
log_info("cl_khr_pci_bus_info not supported. Skipping test...\n");
return TEST_SKIPPED_ITSELF;
@@ -35,14 +34,14 @@ int test_pci_bus_info(cl_device_id deviceID, cl_context context,
cl_device_pci_bus_info_khr info;
size_t size_ret;
error = clGetDeviceInfo(deviceID, CL_DEVICE_PCI_BUS_INFO_KHR, 0, NULL,
&size_ret);
error =
clGetDeviceInfo(device, CL_DEVICE_PCI_BUS_INFO_KHR, 0, NULL, &size_ret);
test_error(error, "Unable to query CL_DEVICE_PCI_BUS_INFO_KHR size");
test_assert_error(
size_ret == sizeof(info),
"Query for CL_DEVICE_PCI_BUS_INFO_KHR returned an unexpected size");
error = clGetDeviceInfo(deviceID, CL_DEVICE_PCI_BUS_INFO_KHR, sizeof(info),
error = clGetDeviceInfo(device, CL_DEVICE_PCI_BUS_INFO_KHR, sizeof(info),
&info, NULL);
test_error(error, "Unable to query CL_DEVICE_PCI_BUS_INFO_KHR");

109
test_conformance/contractions/contractions.cpp
View File

@@ -218,106 +218,37 @@ float ppc_mul(float a, float b)
}
#endif
int test_contractions_float_0(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(0);
}
REGISTER_TEST(contractions_float_0) { return RunTest(0); }
int test_contractions_float_1(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(1);
}
REGISTER_TEST(contractions_float_1) { return RunTest(1); }
int test_contractions_float_2(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(2);
}
REGISTER_TEST(contractions_float_2) { return RunTest(2); }
int test_contractions_float_3(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(3);
}
REGISTER_TEST(contractions_float_3) { return RunTest(3); }
int test_contractions_float_4(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(4);
}
REGISTER_TEST(contractions_float_4) { return RunTest(4); }
int test_contractions_float_5(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(5);
}
REGISTER_TEST(contractions_float_5) { return RunTest(5); }
int test_contractions_float_6(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(6);
}
REGISTER_TEST(contractions_float_6) { return RunTest(6); }
int test_contractions_float_7(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest(7);
}
REGISTER_TEST(contractions_float_7) { return RunTest(7); }
int test_contractions_double_0(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(0);
}
REGISTER_TEST(contractions_double_0) { return RunTest_Double(0); }
int test_contractions_double_1(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(1);
}
REGISTER_TEST(contractions_double_1) { return RunTest_Double(1); }
int test_contractions_double_2(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(2);
}
REGISTER_TEST(contractions_double_2) { return RunTest_Double(2); }
int test_contractions_double_3(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(3);
}
REGISTER_TEST(contractions_double_3) { return RunTest_Double(3); }
int test_contractions_double_4(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(4);
}
REGISTER_TEST(contractions_double_4) { return RunTest_Double(4); }
int test_contractions_double_5(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(5);
}
REGISTER_TEST(contractions_double_5) { return RunTest_Double(5); }
int test_contractions_double_6(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(6);
}
REGISTER_TEST(contractions_double_6) { return RunTest_Double(6); }
int test_contractions_double_7(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
{
return RunTest_Double(7);
}
test_definition test_list[] = {
ADD_TEST( contractions_float_0 ),
ADD_TEST( contractions_float_1 ),
ADD_TEST( contractions_float_2 ),
ADD_TEST( contractions_float_3 ),
ADD_TEST( contractions_float_4 ),
ADD_TEST( contractions_float_5 ),
ADD_TEST( contractions_float_6 ),
ADD_TEST( contractions_float_7 ),
ADD_TEST( contractions_double_0 ),
ADD_TEST( contractions_double_1 ),
ADD_TEST( contractions_double_2 ),
ADD_TEST( contractions_double_3 ),
ADD_TEST( contractions_double_4 ),
ADD_TEST( contractions_double_5 ),
ADD_TEST( contractions_double_6 ),
ADD_TEST( contractions_double_7 ),
};
const int test_num = ARRAY_SIZE( test_list );
REGISTER_TEST(contractions_double_7) { return RunTest_Double(7); }
int main( int argc, const char **argv )
{
@@ -331,7 +262,9 @@ int main( int argc, const char **argv )
if( !error )
{
error = runTestHarnessWithCheck( gArgCount, gArgList, test_num, test_list, true, 0, InitCL );
error = runTestHarnessWithCheck(
gArgCount, gArgList, test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), true, 0, InitCL);
}
if( gQueue )
@@ -450,9 +383,9 @@ static void PrintUsage( void )
vlog( "\t\t-sNUMBER set random seed.\n");
vlog( "\n" );
vlog( "\tTest names:\n" );
for( int i = 0; i < test_num; i++ )
for (size_t i = 0; i < test_registry::getInstance().num_tests(); i++)
{
vlog( "\t\t%s\n", test_list[i].name );
vlog("\t\t%s\n", test_registry::getInstance().definitions()[i].name);
}
}

3
test_conformance/device_execution/device_info.cpp
View File

@@ -20,7 +20,6 @@
#include "harness/testHarness.h"
#include "harness/typeWrappers.h"
#include "procs.h"
#include "utils.h"
static const cl_uint MIN_DEVICE_PREFFERED_QUEUE_SIZE = 16 * 1024;
@@ -29,7 +28,7 @@ static const cl_uint MAX_DEVICE_EMBEDDED_QUEUE_SIZE = 64 * 1024;
#ifdef CL_VERSION_2_0
int test_device_info(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(device_info)
{
cl_int err_ret;
int embedded = 0;

4
test_conformance/device_execution/device_queue.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
static int check_device_queue(cl_device_id device, cl_context context, cl_command_queue queue, cl_uint size)
@@ -97,7 +96,7 @@ static int check_device_queues(cl_device_id device, cl_context context, cl_uint
return res;
}
int test_device_queue(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(device_queue)
{
cl_int err_ret, res = 0;
size_t ret_len;
@@ -187,4 +186,3 @@ int test_device_queue(cl_device_id device, cl_context context, cl_command_queue
return res;
}

4
test_conformance/device_execution/enqueue_block.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -595,7 +594,7 @@ static int check_kernel_results(cl_int* results, cl_int len)
return -1;
}
int test_enqueue_block(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_block)
{
cl_uint i;
cl_int n, err_ret, res = 0;
@@ -660,7 +659,6 @@ int test_enqueue_block(cl_device_id device, cl_context context, cl_command_queue
}
#endif

4
test_conformance/device_execution/enqueue_flags.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -671,7 +670,7 @@ static const kernel_src sources_enqueue_block_flags[] =
static const size_t num_enqueue_block_flags = arr_size(sources_enqueue_block_flags);
int test_enqueue_flags(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_flags)
{
cl_uint i;
cl_int err_ret, res = 0;
@@ -759,5 +758,4 @@ int test_enqueue_flags(cl_device_id device, cl_context context, cl_command_queue
}
#endif

4
test_conformance/device_execution/enqueue_multi_queue.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -95,7 +94,7 @@ static int check_kernel_results(cl_int* results, cl_int len)
return -1;
}
int test_enqueue_multi_queue(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_multi_queue)
{
cl_uint i;
cl_int k, err_ret, res = 0;
@@ -196,7 +195,6 @@ int test_enqueue_multi_queue(cl_device_id device, cl_context context, cl_command
}
#endif

3
test_conformance/device_execution/enqueue_ndrange.cpp
View File

@@ -21,7 +21,6 @@
#include <algorithm>
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -609,7 +608,7 @@ static int check_kernel_results(cl_int* results, cl_int len, std::vector<cl_uint
return -1;
}
int test_enqueue_ndrange(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_ndrange)
{
MTdata d;
cl_uint i;

4
test_conformance/device_execution/enqueue_profiling.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -43,8 +42,7 @@ static const char* enqueue_multi_level = R"(
block_fn(res, level);
})";
int test_enqueue_profiling(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_profiling)
{
cl_int err_ret, res = 0;
clCommandQueueWrapper dev_queue;

3
test_conformance/device_execution/enqueue_wg_size.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -1637,7 +1636,7 @@ static const kernel_src_check sources_enqueue_wg_size[] =
{ KERNEL(enqueue_mix_wg_size_all_diff), check_all_diff_mix }
};
int test_enqueue_wg_size(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_wg_size)
{
MTdata d;
cl_uint i, k;

3
test_conformance/device_execution/execute_block.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -1010,7 +1009,7 @@ static int check_kernel_results(cl_int* results, cl_int len)
return -1;
}
int test_execute_block(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(execute_block)
{
size_t i;
size_t ret_len;

5
test_conformance/device_execution/host_multi_queue.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -129,7 +128,7 @@ static const kernel_src sources_multi_queue_block[] =
static const size_t num_kernels_multi_queue_block = arr_size(sources_multi_queue_block);
int test_host_multi_queue(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(host_multi_queue)
{
cl_uint i;
cl_int err_ret, res = 0;
@@ -228,7 +227,5 @@ int test_host_multi_queue(cl_device_id device, cl_context context, cl_command_qu
}
#endif

3
test_conformance/device_execution/host_queue_order.cpp
View File

@@ -21,7 +21,6 @@
#include <algorithm>
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -90,7 +89,7 @@ static int check_kernel_results(cl_int* results, cl_int len)
/*
Test checks kernel block execution order in case of two different kernels with enqueue block submitted to one ordered host queue.
*/
int test_host_queue_order(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(host_queue_order)
{
cl_int k, err_ret, res = 0;
clCommandQueueWrapper dev_queue;

16
test_conformance/device_execution/main.cpp
View File

@@ -23,7 +23,6 @@
#include "harness/testHarness.h"
#include "harness/parseParameters.h"
#include "utils.h"
#include "procs.h"
std::string gKernelName;
int gWimpyMode = 0;
@@ -57,17 +56,6 @@ test_status InitCL(cl_device_id device) {
return TEST_PASS;
}
test_definition test_list[] = {
ADD_TEST(device_info), ADD_TEST(device_queue),
ADD_TEST(execute_block), ADD_TEST(enqueue_block),
ADD_TEST(enqueue_nested_blocks), ADD_TEST(enqueue_wg_size),
ADD_TEST(enqueue_flags), ADD_TEST(enqueue_multi_queue),
ADD_TEST(host_multi_queue), ADD_TEST(enqueue_ndrange),
ADD_TEST(host_queue_order), ADD_TEST(enqueue_profiling),
};
const int test_num = ARRAY_SIZE( test_list );
int main(int argc, const char *argv[])
{
argc = parseCustomParam(argc, argv);
@@ -98,5 +86,7 @@ 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);
}

3
test_conformance/device_execution/nested_blocks.cpp
View File

@@ -20,7 +20,6 @@
#include <vector>
#include "procs.h"
#include "utils.h"
#include <time.h>
@@ -303,7 +302,7 @@ static const kernel_src_check sources_nested_blocks[] =
{ KERNEL(enqueue_nested_blocks_all_diff), check_all_diff }
};
int test_enqueue_nested_blocks(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(enqueue_nested_blocks)
{
cl_uint i, k;
cl_int err_ret, res = 0;

34
test_conformance/device_execution/procs.h
View File

@@ -1,34 +0,0 @@
//
// Copyright (c) 2017 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "harness/testHarness.h"
extern int test_device_info(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_device_queue(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_execute_block(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_block(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_nested_blocks(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_wg_size(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_flags(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_multi_queue(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_host_multi_queue(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_ndrange(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_host_queue_order(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);
extern int test_enqueue_profiling(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_execution_stress(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements);

51
test_conformance/events/main.cpp
View File

@@ -13,57 +13,10 @@
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "harness/compat.h"
#include <stdio.h>
#include <string.h>
#include "procs.h"
#include "harness/testHarness.h"
#if !defined(_WIN32)
#include <unistd.h>
#endif
test_definition test_list[] = {
ADD_TEST(event_get_execute_status),
ADD_TEST(event_get_write_array_status),
ADD_TEST(event_get_read_array_status),
ADD_TEST(event_get_info),
ADD_TEST(event_wait_for_execute),
ADD_TEST(event_wait_for_array),
ADD_TEST(event_flush),
ADD_TEST(event_finish_execute),
ADD_TEST(event_finish_array),
ADD_TEST(event_release_before_done),
ADD_TEST(event_enqueue_marker),
#ifdef CL_VERSION_1_2
ADD_TEST(event_enqueue_marker_with_event_list),
ADD_TEST(event_enqueue_barrier_with_event_list),
#endif
ADD_TEST(out_of_order_event_waitlist_single_queue),
ADD_TEST(out_of_order_event_waitlist_multi_queue),
ADD_TEST(out_of_order_event_waitlist_multi_queue_multi_device),
ADD_TEST(out_of_order_event_enqueue_wait_for_events_single_queue),
ADD_TEST(out_of_order_event_enqueue_wait_for_events_multi_queue),
ADD_TEST(
out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device),
ADD_TEST(out_of_order_event_enqueue_marker_single_queue),
ADD_TEST(out_of_order_event_enqueue_marker_multi_queue),
ADD_TEST(out_of_order_event_enqueue_marker_multi_queue_multi_device),
ADD_TEST(out_of_order_event_enqueue_barrier_single_queue),
ADD_TEST(waitlists),
ADD_TEST(userevents),
ADD_TEST(callbacks),
ADD_TEST(callbacks_simultaneous),
ADD_TEST(userevents_multithreaded),
ADD_TEST(callback_on_error_simple),
ADD_TEST(callback_on_error_enqueue_command)
};
const int test_num = ARRAY_SIZE(test_list);
int main(int argc, const char *argv[])
{
return runTestHarness(argc, argv, test_num, test_list, false, 0);
return runTestHarness(argc, argv, test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), false, 0);
}

126
test_conformance/events/procs.h
View File

@@ -1,126 +0,0 @@
//
// Copyright (c) 2017 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "harness/errorHelpers.h"
#include "harness/kernelHelpers.h"
#include "harness/typeWrappers.h"
#include "harness/clImageHelper.h"
extern float random_float(float low, float high);
extern float calculate_ulperror(float a, float b);
extern int test_event_get_execute_status(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_event_get_write_array_status(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_event_get_read_array_status(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_event_get_info(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_event_wait_for_execute(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_event_wait_for_array(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_event_flush(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_event_finish_execute(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_event_finish_array(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_event_release_before_done(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_event_enqueue_marker(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
#ifdef CL_VERSION_1_2
extern int test_event_enqueue_marker_with_event_list(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_event_enqueue_barrier_with_event_list(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
#endif
extern int test_out_of_order_event_waitlist_single_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_waitlist_multi_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_waitlist_multi_queue_multi_device(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_enqueue_wait_for_events_single_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_enqueue_wait_for_events_multi_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int
test_out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_enqueue_barrier_single_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_enqueue_marker_single_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_enqueue_marker_multi_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_out_of_order_event_enqueue_marker_multi_queue_multi_device(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements);
extern int test_waitlists(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_userevents(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_callbacks(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_callbacks_simultaneous(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_userevents_multithreaded(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_callback_on_error_simple(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_callback_on_error_enqueue_command(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);

9
test_conformance/events/testBase.h
View File

@@ -17,12 +17,7 @@
#define _testBase_h
#include "harness/compat.h"
#include <stdio.h>
#include <string.h>
#include <sys/types.h>
#include <sys/stat.h>
#include "procs.h"
#include "harness/testHarness.h"
#include "harness/typeWrappers.h"
#endif // _testBase_h

33
test_conformance/events/test_callbacks.cpp
View File

@@ -82,8 +82,9 @@ void CL_CALLBACK combuf_event_callback_function(cl_event event,
*pdata = true;
}
int test_callback_event_single(cl_device_id device, cl_context context,
cl_command_queue queue, Action *actionToTest)
static int test_callback_event_single(cl_device_id device, cl_context context,
cl_command_queue queue,
Action *actionToTest)
{
// Note: we don't use the waiting feature here. We just want to verify that
// we get a callback called when the given event finishes
@@ -163,15 +164,14 @@ int test_callback_event_single(cl_device_id device, cl_context context,
{ \
name##Action action; \
log_info("-- Testing " #name "...\n"); \
if ((error = test_callback_event_single(deviceID, context, queue, \
&action)) \
if ((error = \
test_callback_event_single(device, context, queue, &action)) \
!= CL_SUCCESS) \
retVal++; \
clFinish(queue); \
}
int test_callbacks(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(callbacks)
{
cl_int error;
int retVal = 0;
@@ -185,7 +185,7 @@ int test_callbacks(cl_device_id deviceID, cl_context context,
TEST_ACTION(MapBuffer)
TEST_ACTION(UnmapBuffer)
if (checkForImageSupport(deviceID) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
if (checkForImageSupport(device) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
{
log_info("\nNote: device does not support images. Skipping remainder "
"of callback tests...\n");
@@ -199,7 +199,7 @@ int test_callbacks(cl_device_id deviceID, cl_context context,
TEST_ACTION(CopyBufferTo2DImage)
TEST_ACTION(MapImage)
if (checkFor3DImageSupport(deviceID) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
if (checkFor3DImageSupport(device) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
log_info("\nNote: device does not support 3D images. Skipping "
"remainder of waitlist tests...\n");
else
@@ -237,8 +237,7 @@ void CL_CALLBACK simultaneous_event_callback_function(cl_event event,
ThreadPool_AtomicAdd(&sSimultaneousCount, 1);
}
int test_callbacks_simultaneous(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(callbacks_simultaneous)
{
cl_int error;
@@ -255,7 +254,7 @@ int test_callbacks_simultaneous(cl_device_id deviceID, cl_context context,
actions[index++] = new MapBufferAction();
actions[index++] = new UnmapBufferAction();
if (checkForImageSupport(deviceID) != CL_IMAGE_FORMAT_NOT_SUPPORTED)
if (checkForImageSupport(device) != CL_IMAGE_FORMAT_NOT_SUPPORTED)
{
actions[index++] = new ReadImage2DAction();
actions[index++] = new WriteImage2DAction();
@@ -264,7 +263,7 @@ int test_callbacks_simultaneous(cl_device_id deviceID, cl_context context,
actions[index++] = new CopyBufferTo2DImageAction();
actions[index++] = new MapImageAction();
if (checkFor3DImageSupport(deviceID) != CL_IMAGE_FORMAT_NOT_SUPPORTED)
if (checkFor3DImageSupport(device) != CL_IMAGE_FORMAT_NOT_SUPPORTED)
{
actions[index++] = new ReadImage3DAction();
actions[index++] = new WriteImage3DAction();
@@ -282,7 +281,7 @@ int test_callbacks_simultaneous(cl_device_id deviceID, cl_context context,
log_info("\tSetting up test events...\n");
for (index = 0; actions[index] != NULL; index++)
{
error = actions[index]->Setup(deviceID, context, queue);
error = actions[index]->Setup(device, context, queue);
test_error(error, "Unable to set up test action");
sSimultaneousFlags[index] = false;
}
@@ -384,8 +383,7 @@ int test_callbacks_simultaneous(cl_device_id deviceID, cl_context context,
return -1;
}
int test_callback_on_error_simple(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(callback_on_error_simple)
{
cl_int error = CL_SUCCESS;
clEventWrapper user_event = clCreateUserEvent(context, &error);
@@ -412,10 +410,7 @@ int test_callback_on_error_simple(cl_device_id deviceID, cl_context context,
return CL_SUCCESS;
}
int test_callback_on_error_enqueue_command(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(callback_on_error_enqueue_command)
{
cl_int error = CL_SUCCESS;
bool confirmation = false;

116
test_conformance/events/test_event_dependencies.cpp
View File

@@ -44,7 +44,7 @@ const char *write_kernels[] = {
them (only for single queue). If neither are set, nothing is done to prevent
them from executing in the wrong order. This can be used for verification.
*/
int test_event_enqueue_wait_for_events_run_test(
static int test_event_enqueue_wait_for_events_run_test(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements, int two_queues, int two_devices,
int test_enqueue_wait_for_events, int test_barrier, int use_waitlist,
@@ -473,10 +473,10 @@ int test_event_enqueue_wait_for_events_run_test(
return failed;
}
int test(cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements, int two_queues, int two_devices,
int test_enqueue_wait_for_events, int test_barrier, int use_waitlists,
int use_marker)
static int run_test(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements, int two_queues,
int two_devices, int test_enqueue_wait_for_events,
int test_barrier, int use_waitlists, int use_marker)
{
if (!checkDeviceForQueueSupport(deviceID,
CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE))
@@ -548,10 +548,7 @@ int test(cl_device_id deviceID, cl_context context, cl_command_queue queue,
}
int test_out_of_order_event_waitlist_single_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_waitlist_single_queue)
{
int two_queues = 0;
int two_devices = 0;
@@ -559,15 +556,12 @@ int test_out_of_order_event_waitlist_single_queue(cl_device_id deviceID,
int test_barrier = 0;
int use_waitlists = 1;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_waitlist_multi_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_waitlist_multi_queue)
{
int two_queues = 1;
int two_devices = 0;
@@ -575,14 +569,12 @@ int test_out_of_order_event_waitlist_multi_queue(cl_device_id deviceID,
int test_barrier = 0;
int use_waitlists = 1;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_waitlist_multi_queue_multi_device(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_waitlist_multi_queue_multi_device)
{
int two_queues = 1;
int two_devices = 1;
@@ -590,15 +582,13 @@ int test_out_of_order_event_waitlist_multi_queue_multi_device(
int test_barrier = 0;
int use_waitlists = 1;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_wait_for_events_single_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_enqueue_wait_for_events_single_queue)
{
int two_queues = 0;
int two_devices = 0;
@@ -606,14 +596,12 @@ int test_out_of_order_event_enqueue_wait_for_events_single_queue(
int test_barrier = 0;
int use_waitlists = 0;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_wait_for_events_multi_queue(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_enqueue_wait_for_events_multi_queue)
{
int two_queues = 1;
int two_devices = 0;
@@ -621,15 +609,14 @@ int test_out_of_order_event_enqueue_wait_for_events_multi_queue(
int test_barrier = 0;
int use_waitlists = 0;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements)
REGISTER_TEST(
out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device)
{
int two_queues = 1;
int two_devices = 1;
@@ -637,16 +624,13 @@ int test_out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device(
int test_barrier = 0;
int use_waitlists = 0;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_barrier_single_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_enqueue_barrier_single_queue)
{
int two_queues = 0;
int two_devices = 0;
@@ -654,16 +638,13 @@ int test_out_of_order_event_enqueue_barrier_single_queue(cl_device_id deviceID,
int test_barrier = 1;
int use_waitlists = 0;
int use_marker = 0;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_marker_single_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_enqueue_marker_single_queue)
{
int two_queues = 0;
int two_devices = 0;
@@ -671,15 +652,12 @@ int test_out_of_order_event_enqueue_marker_single_queue(cl_device_id deviceID,
int test_barrier = 0;
int use_waitlists = 0;
int use_marker = 1;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_marker_multi_queue(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_enqueue_marker_multi_queue)
{
int two_queues = 1;
int two_devices = 0;
@@ -687,15 +665,13 @@ int test_out_of_order_event_enqueue_marker_multi_queue(cl_device_id deviceID,
int test_barrier = 0;
int use_waitlists = 0;
int use_marker = 1;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}
int test_out_of_order_event_enqueue_marker_multi_queue_multi_device(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
int num_elements)
REGISTER_TEST(out_of_order_event_enqueue_marker_multi_queue_multi_device)
{
int two_queues = 1;
int two_devices = 1;
@@ -703,7 +679,7 @@ int test_out_of_order_event_enqueue_marker_multi_queue_multi_device(
int test_barrier = 0;
int use_waitlists = 0;
int use_marker = 1;
return test(deviceID, context, queue, num_elements, two_queues, two_devices,
test_enqueue_wait_for_events, test_barrier, use_waitlists,
use_marker);
return run_test(device, context, queue, num_elements, two_queues,
two_devices, test_enqueue_wait_for_events, test_barrier,
use_waitlists, use_marker);
}

43
test_conformance/events/test_events.cpp
View File

@@ -102,8 +102,7 @@ const char *IGetStatusString(cl_int status)
}
/* Note: tests clGetEventStatus and clReleaseEvent (implicitly) */
int test_event_get_execute_status(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_get_execute_status)
{
cl_int status;
SETUP_EVENT(context, queue);
@@ -128,8 +127,7 @@ int test_event_get_execute_status(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_get_info(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_get_info)
{
SETUP_EVENT(context, queue);
@@ -204,8 +202,7 @@ int test_event_get_info(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_get_write_array_status(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_get_write_array_status)
{
cl_mem stream;
cl_float testArray[1024 * 32];
@@ -246,8 +243,7 @@ int test_event_get_write_array_status(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_get_read_array_status(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_get_read_array_status)
{
cl_mem stream;
cl_float testArray[1024 * 32];
@@ -305,8 +301,7 @@ int test_event_get_read_array_status(cl_device_id deviceID, cl_context context,
/* clGetEventStatus not implemented yet */
int test_event_wait_for_execute(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_wait_for_execute)
{
cl_int status;
SETUP_EVENT(context, queue);
@@ -331,8 +326,7 @@ int test_event_wait_for_execute(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_wait_for_array(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_wait_for_array)
{
cl_mem streams[2];
cl_float readArray[1024 * 32];
@@ -419,8 +413,7 @@ int test_event_wait_for_array(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_flush(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_flush)
{
cl_int status;
SETUP_EVENT(context, queue);
@@ -472,8 +465,7 @@ int test_event_flush(cl_device_id deviceID, cl_context context,
}
int test_event_finish_execute(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_finish_execute)
{
cl_int status;
SETUP_EVENT(context, queue);
@@ -498,8 +490,7 @@ int test_event_finish_execute(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_finish_array(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_finish_array)
{
cl_mem streams[2];
cl_float readArray[1024 * 32];
@@ -589,8 +580,7 @@ int test_event_finish_array(cl_device_id deviceID, cl_context context,
#define NUM_EVENT_RUNS 100
int test_event_release_before_done(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_release_before_done)
{
// Create a kernel to run
clProgramWrapper program;
@@ -680,8 +670,7 @@ int test_event_release_before_done(cl_device_id deviceID, cl_context context,
return 0;
}
int test_event_enqueue_marker(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(event_enqueue_marker)
{
cl_int status;
SETUP_EVENT(context, queue);
@@ -718,10 +707,7 @@ int test_event_enqueue_marker(cl_device_id deviceID, cl_context context,
}
#ifdef CL_VERSION_1_2
int test_event_enqueue_marker_with_event_list(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(event_enqueue_marker_with_event_list)
{
SETUP_EVENT(context, queue);
cl_event event_list[3] = { NULL, NULL, NULL };
@@ -771,10 +757,7 @@ int test_event_enqueue_marker_with_event_list(cl_device_id deviceID,
return 0;
}
int test_event_enqueue_barrier_with_event_list(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements)
REGISTER_TEST(event_enqueue_barrier_with_event_list)
{
SETUP_EVENT(context, queue);
cl_event event_list[3] = { NULL, NULL, NULL };

3
test_conformance/events/test_userevents.cpp
View File

@@ -78,8 +78,7 @@ enum
MaxDevices = 8
};
int test_userevents(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(userevents)
{
cl_int err;

5
test_conformance/events/test_userevents_multithreaded.cpp
View File

@@ -30,8 +30,7 @@ void trigger_user_event(cl_event *event)
clSetUserEventStatus(*event, CL_COMPLETE);
}
int test_userevents_multithreaded(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(userevents_multithreaded)
{
cl_int error;
@@ -50,7 +49,7 @@ int test_userevents_multithreaded(cl_device_id deviceID, cl_context context,
for (int i = 0; actions[i] != NULL; i++)
{
error = actions[i]->Setup(deviceID, context, queue);
error = actions[i]->Setup(device, context, queue);
test_error(error, "Unable to set up test action");
error = actions[i]->Execute(queue, 1, &gateEvent, &actionEvents[i]);

28
test_conformance/events/test_waitlists.cpp
View File

@@ -21,8 +21,9 @@ extern const char *IGetStatusString(cl_int status);
#define PRINT_OPS 0
int test_waitlist(cl_device_id device, cl_context context,
cl_command_queue queue, Action *actionToTest, bool multiple)
static int test_waitlist(cl_device_id device, cl_context context,
cl_command_queue queue, Action *actionToTest,
bool multiple)
{
NDRangeKernelAction actions[2];
clEventWrapper events[3];
@@ -314,38 +315,39 @@ int test_waitlist(cl_device_id device, cl_context context,
{ \
name##Action action; \
log_info("-- Testing " #name " (waiting on 1 event)...\n"); \
if ((error = test_waitlist(deviceID, context, queue, &action, false)) \
if ((error = \
test_waitlist(device, context, test_queue, &action, false)) \
!= CL_SUCCESS) \
retVal++; \
clFinish(queue); \
clFinish(test_queue); \
} \
if (error \
== CL_SUCCESS) /* Only run multiples test if single test passed */ \
{ \
name##Action action; \
log_info("-- Testing " #name " (waiting on 2 events)...\n"); \
if ((error = test_waitlist(deviceID, context, queue, &action, true)) \
if ((error = \
test_waitlist(device, context, test_queue, &action, true)) \
!= CL_SUCCESS) \
retVal++; \
clFinish(queue); \
clFinish(test_queue); \
}
int test_waitlists(cl_device_id deviceID, cl_context context,
cl_command_queue oldQueue, int num_elements)
REGISTER_TEST(waitlists)
{
cl_int error;
int retVal = 0;
cl_command_queue_properties props = CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE;
if (!checkDeviceForQueueSupport(deviceID, props))
if (!checkDeviceForQueueSupport(device, props))
{
log_info("WARNING: Device does not support out-of-order exec mode; "
"skipping test.\n");
return 0;
}
clCommandQueueWrapper queue =
clCreateCommandQueue(context, deviceID, props, &error);
clCommandQueueWrapper test_queue =
clCreateCommandQueue(context, device, props, &error);
test_error(error, "Unable to create out-of-order queue");
log_info("\n");
@@ -357,7 +359,7 @@ int test_waitlists(cl_device_id deviceID, cl_context context,
TEST_ACTION(MapBuffer)
TEST_ACTION(UnmapBuffer)
if (checkForImageSupport(deviceID) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
if (checkForImageSupport(device) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
{
log_info("\nNote: device does not support images. Skipping remainder "
"of waitlist tests...\n");
@@ -371,7 +373,7 @@ int test_waitlists(cl_device_id deviceID, cl_context context,
TEST_ACTION(CopyBufferTo2DImage)
TEST_ACTION(MapImage)
if (checkFor3DImageSupport(deviceID) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
if (checkFor3DImageSupport(device) == CL_IMAGE_FORMAT_NOT_SUPPORTED)
log_info("Device does not support 3D images. Skipping remainder of "
"waitlist tests...\n");
else

4
test_conformance/half/Test_roundTrip.cpp
View File

@@ -22,7 +22,7 @@
#include "tests.h"
#include "harness/testHarness.h"
int test_roundTrip( cl_device_id device, cl_context context, cl_command_queue queue, int num_elements )
REGISTER_TEST(roundTrip)
{
int vectorSize, error;
uint64_t i, j;
@@ -400,5 +400,3 @@ exit:
return error;
}

11
test_conformance/half/Test_vLoadHalf.cpp
View File

@@ -604,13 +604,6 @@ exit:
return error;
}
int test_vload_half( cl_device_id device, cl_context context, cl_command_queue queue, int num_elements )
{
return Test_vLoadHalf_private( device, false );
}
int test_vloada_half( cl_device_id device, cl_context context, cl_command_queue queue, int num_elements )
{
return Test_vLoadHalf_private( device, true );
}
REGISTER_TEST(vload_half) { return Test_vLoadHalf_private(device, false); }
REGISTER_TEST(vloada_half) { return Test_vLoadHalf_private(device, true); }

58
test_conformance/half/Test_vStoreHalf.cpp
View File

@@ -247,89 +247,79 @@ static cl_half double2half_rtn(double f)
return cl_half_from_double(f, CL_HALF_RTN);
}
int test_vstore_half(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstore_half)
{
switch (get_default_rounding_mode(deviceID))
switch (get_default_rounding_mode(device))
{
case CL_FP_ROUND_TO_ZERO:
return Test_vStoreHalf_private(deviceID, float2half_rtz,
return Test_vStoreHalf_private(device, float2half_rtz,
double2half_rte, "");
case 0: return -1;
default:
return Test_vStoreHalf_private(deviceID, float2half_rte,
return Test_vStoreHalf_private(device, float2half_rte,
double2half_rte, "");
}
}
int test_vstore_half_rte(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstore_half_rte)
{
return Test_vStoreHalf_private(deviceID, float2half_rte, double2half_rte,
return Test_vStoreHalf_private(device, float2half_rte, double2half_rte,
"_rte");
}
int test_vstore_half_rtz(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstore_half_rtz)
{
return Test_vStoreHalf_private(deviceID, float2half_rtz, double2half_rtz,
return Test_vStoreHalf_private(device, float2half_rtz, double2half_rtz,
"_rtz");
}
int test_vstore_half_rtp(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstore_half_rtp)
{
return Test_vStoreHalf_private(deviceID, float2half_rtp, double2half_rtp,
return Test_vStoreHalf_private(device, float2half_rtp, double2half_rtp,
"_rtp");
}
int test_vstore_half_rtn(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstore_half_rtn)
{
return Test_vStoreHalf_private(deviceID, float2half_rtn, double2half_rtn,
return Test_vStoreHalf_private(device, float2half_rtn, double2half_rtn,
"_rtn");
}
int test_vstorea_half(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstorea_half)
{
switch (get_default_rounding_mode(deviceID))
switch (get_default_rounding_mode(device))
{
case CL_FP_ROUND_TO_ZERO:
return Test_vStoreaHalf_private(deviceID, float2half_rtz,
return Test_vStoreaHalf_private(device, float2half_rtz,
double2half_rte, "");
case 0: return -1;
default:
return Test_vStoreaHalf_private(deviceID, float2half_rte,
return Test_vStoreaHalf_private(device, float2half_rte,
double2half_rte, "");
}
}
int test_vstorea_half_rte(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstorea_half_rte)
{
return Test_vStoreaHalf_private(deviceID, float2half_rte, double2half_rte,
return Test_vStoreaHalf_private(device, float2half_rte, double2half_rte,
"_rte");
}
int test_vstorea_half_rtz(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstorea_half_rtz)
{
return Test_vStoreaHalf_private(deviceID, float2half_rtz, double2half_rtz,
return Test_vStoreaHalf_private(device, float2half_rtz, double2half_rtz,
"_rtz");
}
int test_vstorea_half_rtp(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstorea_half_rtp)
{
return Test_vStoreaHalf_private(deviceID, float2half_rtp, double2half_rtp,
return Test_vStoreaHalf_private(device, float2half_rtp, double2half_rtp,
"_rtp");
}
int test_vstorea_half_rtn(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(vstorea_half_rtn)
{
return Test_vStoreaHalf_private(deviceID, float2half_rtn, double2half_rtn,
return Test_vStoreaHalf_private(device, float2half_rtn, double2half_rtn,
"_rtn");
}

26
test_conformance/half/main.cpp
View File

@@ -53,24 +53,6 @@ int g_arrVecSizes[kVectorSizeCount+kStrangeVectorSizeCount];
int g_arrVecAligns[kLargestVectorSize+1];
static int arrStrangeVecSizes[kStrangeVectorSizeCount] = {3};
test_definition test_list[] = {
ADD_TEST( vload_half ),
ADD_TEST( vloada_half ),
ADD_TEST( vstore_half ),
ADD_TEST( vstorea_half ),
ADD_TEST( vstore_half_rte ),
ADD_TEST( vstorea_half_rte ),
ADD_TEST( vstore_half_rtz ),
ADD_TEST( vstorea_half_rtz ),
ADD_TEST( vstore_half_rtp ),
ADD_TEST( vstorea_half_rtp ),
ADD_TEST( vstore_half_rtn ),
ADD_TEST( vstorea_half_rtn ),
ADD_TEST( roundTrip ),
};
const int test_num = ARRAY_SIZE( test_list );
int main (int argc, const char **argv )
{
int error;
@@ -109,7 +91,9 @@ int main (int argc, const char **argv )
}
fflush( stdout );
error = runTestHarnessWithCheck( argCount, argList, test_num, test_list, true, 0, InitCL );
error = runTestHarnessWithCheck(
argCount, argList, test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), true, 0, InitCL);
exit:
if(gQueue)
@@ -248,8 +232,8 @@ static void PrintUsage( void )
"1-12, default factor(%u)\n",
gWimpyReductionFactor);
vlog("\t\t-h\tHelp\n");
for (int i = 0; i < test_num; i++)
for (size_t i = 0; i < test_registry::getInstance().num_tests(); i++)
{
vlog("\t\t%s\n", test_list[i].name );
vlog("\t\t%s\n", test_registry::getInstance().definitions()[i].name);
}
}

2
test_conformance/mem_host_flags/checker.h
View File

@@ -21,7 +21,7 @@
#include <stdio.h>
#include <string.h>
#include "procs.h"
#include "testBase.h"
#include "C_host_memory_block.h"
#define TEST_VALUE 5

19
test_conformance/mem_host_flags/main.cpp
View File

@@ -22,30 +22,17 @@
#include <CL/cl.h>
#endif
#include "procs.h"
#include "harness/testHarness.h"
#include "harness/imageHelpers.h"
#if !defined(_WIN32)
#include <unistd.h>
#endif
test_definition test_list[] = {
ADD_TEST(mem_host_read_only_buffer),
ADD_TEST(mem_host_read_only_subbuffer),
ADD_TEST(mem_host_write_only_buffer),
ADD_TEST(mem_host_write_only_subbuffer),
ADD_TEST(mem_host_no_access_buffer),
ADD_TEST(mem_host_no_access_subbuffer),
ADD_TEST(mem_host_read_only_image),
ADD_TEST(mem_host_write_only_image),
ADD_TEST(mem_host_no_access_image),
};
const int test_num = ARRAY_SIZE(test_list);
int main(int argc, const char *argv[])
{
log_info("1st part, non gl-sharing objects...\n");
gTestRounding = true;
return runTestHarness(argc, argv, test_num, test_list, false, 0);
return runTestHarness(argc, argv, test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), false, 0);
}

105
test_conformance/mem_host_flags/mem_host_buffer.cpp
View File

@@ -20,20 +20,18 @@
#include <sys/types.h>
#include <sys/stat.h>
#include "procs.h"
#include "checker_mem_host_read_only.hpp"
#include "checker_mem_host_write_only.hpp"
#include "checker_mem_host_no_access.hpp"
static int test_mem_host_read_only_buffer_RW(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_read_only<TEST_ELEMENT_TYPE> checker(deviceID,
context, queue);
cBuffer_check_mem_host_read_only<TEST_ELEMENT_TYPE> checker(device, context,
queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
cl_int err;
@@ -55,14 +53,14 @@ static int test_mem_host_read_only_buffer_RW(
}
static int test_mem_host_read_only_buffer_RW_Rect(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_read_only<TEST_ELEMENT_TYPE> checker(deviceID,
context, queue);
cBuffer_check_mem_host_read_only<TEST_ELEMENT_TYPE> checker(device, context,
queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
cl_int err;
@@ -84,14 +82,14 @@ static int test_mem_host_read_only_buffer_RW_Rect(
}
static int test_mem_host_read_only_buffer_RW_Mapping(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_read_only<TEST_ELEMENT_TYPE> checker(deviceID,
context, queue);
cBuffer_check_mem_host_read_only<TEST_ELEMENT_TYPE> checker(device, context,
queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
cl_int err;
@@ -112,8 +110,7 @@ static int test_mem_host_read_only_buffer_RW_Mapping(
return err;
}
int test_mem_host_read_only_buffer(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_read_only_buffer)
{
cl_mem_flags buffer_mem_flags[2] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_READ_ONLY,
@@ -128,17 +125,17 @@ int test_mem_host_read_only_buffer(cl_device_id deviceID, cl_context context,
{
err = test_mem_host_read_only_buffer_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[k], 0,
device, context, queue, blocking[i], buffer_mem_flags[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_read_only_buffer_RW_Rect(
deviceID, context, queue, blocking[i], buffer_mem_flags[k], 0,
device, context, queue, blocking[i], buffer_mem_flags[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_read_only_buffer_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[k], 0,
device, context, queue, blocking[i], buffer_mem_flags[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
}
@@ -146,8 +143,7 @@ int test_mem_host_read_only_buffer(cl_device_id deviceID, cl_context context,
return err;
}
int test_mem_host_read_only_subbuffer(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_read_only_subbuffer)
{
cl_mem_flags parent_buffer_mem_flags[1] = { CL_MEM_READ_WRITE
| CL_MEM_USE_HOST_PTR
@@ -169,17 +165,17 @@ int test_mem_host_read_only_subbuffer(cl_device_id deviceID, cl_context context,
for (int i = 0; i < 2; i++)
{
err = test_mem_host_read_only_buffer_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
parent_buffer_mem_flags[p], _Sub_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_read_only_buffer_RW_Rect(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
parent_buffer_mem_flags[p], _Sub_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_read_only_buffer_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
parent_buffer_mem_flags[p], _Sub_BUFFER);
test_error(err, __FUNCTION__);
}
@@ -191,14 +187,14 @@ int test_mem_host_read_only_subbuffer(cl_device_id deviceID, cl_context context,
//=============================== Write only
static cl_int test_mem_host_write_only_buffer_RW(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_write_only<TEST_ELEMENT_TYPE> checker(
deviceID, context, queue);
device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -221,14 +217,14 @@ static cl_int test_mem_host_write_only_buffer_RW(
}
static cl_int test_mem_host_write_only_buffer_RW_Rect(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_write_only<TEST_ELEMENT_TYPE> checker(
deviceID, context, queue);
device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
cl_int err;
@@ -250,14 +246,14 @@ static cl_int test_mem_host_write_only_buffer_RW_Rect(
}
static cl_int test_mem_host_write_only_buffer_RW_Mapping(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_write_only<TEST_ELEMENT_TYPE> checker(
deviceID, context, queue);
device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
cl_int err;
@@ -278,8 +274,7 @@ static cl_int test_mem_host_write_only_buffer_RW_Mapping(
return err;
}
int test_mem_host_write_only_buffer(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_write_only_buffer)
{
cl_mem_flags buffer_mem_flags[2] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_WRITE_ONLY,
@@ -293,17 +288,17 @@ int test_mem_host_write_only_buffer(cl_device_id deviceID, cl_context context,
for (int i = 0; i < 2; i++)
{
err = test_mem_host_write_only_buffer_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[k], 0,
device, context, queue, blocking[i], buffer_mem_flags[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_write_only_buffer_RW_Rect(
deviceID, context, queue, blocking[i], buffer_mem_flags[k], 0,
device, context, queue, blocking[i], buffer_mem_flags[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_write_only_buffer_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[k], 0,
device, context, queue, blocking[i], buffer_mem_flags[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
}
@@ -311,9 +306,7 @@ int test_mem_host_write_only_buffer(cl_device_id deviceID, cl_context context,
return err;
}
int test_mem_host_write_only_subbuffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_write_only_subbuffer)
{
cl_mem_flags parent_buffer_mem_flags[1] = { CL_MEM_READ_WRITE
| CL_MEM_USE_HOST_PTR
@@ -336,17 +329,17 @@ int test_mem_host_write_only_subbuffer(cl_device_id deviceID,
for (int i = 0; i < 2; i++)
{
err = test_mem_host_write_only_buffer_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[m],
device, context, queue, blocking[i], buffer_mem_flags[m],
parent_buffer_mem_flags[p], _Sub_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_write_only_buffer_RW_Rect(
deviceID, context, queue, blocking[i], buffer_mem_flags[m],
device, context, queue, blocking[i], buffer_mem_flags[m],
parent_buffer_mem_flags[p], _Sub_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_write_only_buffer_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[m],
device, context, queue, blocking[i], buffer_mem_flags[m],
parent_buffer_mem_flags[p], _Sub_BUFFER);
test_error(err, __FUNCTION__);
}
@@ -359,14 +352,14 @@ int test_mem_host_write_only_subbuffer(cl_device_id deviceID,
//===================== NO ACCESS
static cl_int test_mem_host_no_access_buffer_RW(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_no_access<TEST_ELEMENT_TYPE> checker(deviceID,
context, queue);
cBuffer_check_mem_host_no_access<TEST_ELEMENT_TYPE> checker(device, context,
queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -389,14 +382,14 @@ static cl_int test_mem_host_no_access_buffer_RW(
}
static cl_int test_mem_host_no_access_buffer_RW_Rect(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_no_access<TEST_ELEMENT_TYPE> checker(deviceID,
context, queue);
cBuffer_check_mem_host_no_access<TEST_ELEMENT_TYPE> checker(device, context,
queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
cl_int err;
@@ -418,14 +411,14 @@ static cl_int test_mem_host_no_access_buffer_RW_Rect(
}
static cl_int test_mem_host_no_access_buffer_RW_Mapping(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_flags parent_buffer_flag, enum BUFFER_TYPE buffer_type)
{
log_info("%s\n", __FUNCTION__);
cBuffer_check_mem_host_no_access<TEST_ELEMENT_TYPE> checker(deviceID,
context, queue);
cBuffer_check_mem_host_no_access<TEST_ELEMENT_TYPE> checker(device, context,
queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -447,8 +440,7 @@ static cl_int test_mem_host_no_access_buffer_RW_Mapping(
return err;
}
int test_mem_host_no_access_buffer(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_no_access_buffer)
{
cl_mem_flags buffer_mem_flag[2] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_NO_ACCESS,
@@ -462,17 +454,17 @@ int test_mem_host_no_access_buffer(cl_device_id deviceID, cl_context context,
for (int i = 0; i < 2; i++)
{
err = test_mem_host_no_access_buffer_RW(
deviceID, context, queue, blocking[i], buffer_mem_flag[k], 0,
device, context, queue, blocking[i], buffer_mem_flag[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_no_access_buffer_RW_Rect(
deviceID, context, queue, blocking[i], buffer_mem_flag[k], 0,
device, context, queue, blocking[i], buffer_mem_flag[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
err = test_mem_host_no_access_buffer_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flag[k], 0,
device, context, queue, blocking[i], buffer_mem_flag[k], 0,
_BUFFER);
test_error(err, __FUNCTION__);
}
@@ -480,8 +472,7 @@ int test_mem_host_no_access_buffer(cl_device_id deviceID, cl_context context,
return err;
}
int test_mem_host_no_access_subbuffer(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_no_access_subbuffer)
{
cl_mem_flags parent_buffer_mem_flags[3] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_NO_ACCESS,
@@ -505,15 +496,15 @@ int test_mem_host_no_access_subbuffer(cl_device_id deviceID, cl_context context,
for (int i = 0; i < 2; i++)
{
err += test_mem_host_no_access_buffer_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
parent_buffer_mem_flags[p], _Sub_BUFFER);
err += test_mem_host_no_access_buffer_RW_Rect(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
parent_buffer_mem_flags[p], _Sub_BUFFER);
err += test_mem_host_no_access_buffer_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
parent_buffer_mem_flags[p], _Sub_BUFFER);
}
}

65
test_conformance/mem_host_flags/mem_host_image.cpp
View File

@@ -20,22 +20,20 @@
#include <sys/types.h>
#include <sys/stat.h>
#include "procs.h"
#include "checker_image_mem_host_read_only.hpp"
#include "checker_image_mem_host_no_access.hpp"
#include "checker_image_mem_host_write_only.hpp"
//======================================
static cl_int test_mem_host_read_only_RW_Image(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_object_type image_type_in, size_t array_size, size_t *img_dim)
{
log_info("%s ... \n ", __FUNCTION__);
cl_int err = CL_SUCCESS;
cImage_check_mem_host_read_only<int> checker(deviceID, context, queue);
cImage_check_mem_host_read_only<int> checker(device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -58,14 +56,14 @@ static cl_int test_mem_host_read_only_RW_Image(
}
static cl_int test_mem_host_read_only_RW_Image_Mapping(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_object_type image_type_in, size_t array_size, size_t *img_dim)
{
log_info("%s ... \n ", __FUNCTION__);
cl_int err = CL_SUCCESS;
cImage_check_mem_host_read_only<int> checker(deviceID, context, queue);
cImage_check_mem_host_read_only<int> checker(device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -87,8 +85,7 @@ static cl_int test_mem_host_read_only_RW_Image_Mapping(
return err;
}
int test_mem_host_read_only_image(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_read_only_image)
{
cl_mem_flags buffer_mem_flags[2] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_READ_ONLY,
@@ -98,8 +95,8 @@ int test_mem_host_read_only_image(cl_device_id deviceID, cl_context context,
cl_int err = CL_SUCCESS;
cl_bool image_support;
err = clGetDeviceInfo(deviceID, CL_DEVICE_IMAGE_SUPPORT,
sizeof image_support, &image_support, NULL);
err = clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof image_support,
&image_support, NULL);
if (err)
{
test_error(err, __FUNCTION__);
@@ -132,16 +129,14 @@ int test_mem_host_read_only_image(cl_device_id deviceID, cl_context context,
for (int p = 0; p < 3; p++)
{
err = test_mem_host_read_only_RW_Image(
deviceID, context, queue, blocking[i],
buffer_mem_flags[flag], img_type[p], array_size[p],
img_dims[p]);
device, context, queue, blocking[i], buffer_mem_flags[flag],
img_type[p], array_size[p], img_dims[p]);
test_error(err, __FUNCTION__);
err = test_mem_host_read_only_RW_Image_Mapping(
deviceID, context, queue, blocking[i],
buffer_mem_flags[flag], img_type[p], array_size[p],
img_dims[p]);
device, context, queue, blocking[i], buffer_mem_flags[flag],
img_type[p], array_size[p], img_dims[p]);
test_error(err, __FUNCTION__);
}
@@ -152,14 +147,14 @@ int test_mem_host_read_only_image(cl_device_id deviceID, cl_context context,
//----------------------------
static cl_int test_MEM_HOST_WRITE_ONLY_Image_RW(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_object_type image_type_in, size_t array_size, size_t *img_dim)
{
log_info(" %s ... \n ", __FUNCTION__);
cl_int err = CL_SUCCESS;
cImage_check_mem_host_write_only<int> checker(deviceID, context, queue);
cImage_check_mem_host_write_only<int> checker(device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -185,14 +180,14 @@ static cl_int test_MEM_HOST_WRITE_ONLY_Image_RW(
}
static cl_int test_MEM_HOST_WRITE_ONLY_Image_RW_Mapping(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_object_type image_type_in, size_t array_size, size_t *img_dim)
{
log_info("%s ... \n ", __FUNCTION__);
cl_int err = CL_SUCCESS;
cImage_check_mem_host_write_only<int> checker(deviceID, context, queue);
cImage_check_mem_host_write_only<int> checker(device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -217,8 +212,7 @@ static cl_int test_MEM_HOST_WRITE_ONLY_Image_RW_Mapping(
return err;
}
int test_mem_host_write_only_image(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_write_only_image)
{
cl_mem_flags buffer_mem_flags[2] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_WRITE_ONLY,
@@ -228,8 +222,8 @@ int test_mem_host_write_only_image(cl_device_id deviceID, cl_context context,
cl_int err = CL_SUCCESS;
cl_bool image_support;
err = clGetDeviceInfo(deviceID, CL_DEVICE_IMAGE_SUPPORT,
sizeof image_support, &image_support, NULL);
err = clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof image_support,
&image_support, NULL);
if (err)
{
test_error(err, __FUNCTION__);
@@ -261,12 +255,12 @@ int test_mem_host_write_only_image(cl_device_id deviceID, cl_context context,
for (int p = 0; p < 3; p++)
{
err = test_MEM_HOST_WRITE_ONLY_Image_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
img_type[p], array_size[p], img_dims[p]);
test_error(err, __FUNCTION__);
err = test_MEM_HOST_WRITE_ONLY_Image_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
img_type[p], array_size[p], img_dims[p]);
test_error(err, __FUNCTION__);
}
@@ -278,14 +272,14 @@ int test_mem_host_write_only_image(cl_device_id deviceID, cl_context context,
//--------
static cl_int test_mem_host_no_access_Image_RW(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_object_type image_type_in, size_t array_size, size_t *img_dim)
{
log_info("%s ... \n", __FUNCTION__);
cl_int err = CL_SUCCESS;
cImage_check_mem_host_no_access<int> checker(deviceID, context, queue);
cImage_check_mem_host_no_access<int> checker(device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -310,14 +304,14 @@ static cl_int test_mem_host_no_access_Image_RW(
}
static cl_int test_mem_host_no_access_Image_RW_Mapping(
cl_device_id deviceID, cl_context context, cl_command_queue queue,
cl_device_id device, cl_context context, cl_command_queue queue,
cl_bool blocking, cl_mem_flags buffer_mem_flag,
cl_mem_object_type image_type_in, size_t array_size, size_t *img_dim)
{
log_info("%s ... \n ", __FUNCTION__);
cl_int err = CL_SUCCESS;
cImage_check_mem_host_no_access<int> checker(deviceID, context, queue);
cImage_check_mem_host_no_access<int> checker(device, context, queue);
checker.m_blocking = blocking;
checker.buffer_mem_flag = buffer_mem_flag;
@@ -341,8 +335,7 @@ static cl_int test_mem_host_no_access_Image_RW_Mapping(
return err;
}
int test_mem_host_no_access_image(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(mem_host_no_access_image)
{
cl_mem_flags buffer_mem_flags[2] = {
CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR | CL_MEM_HOST_NO_ACCESS,
@@ -352,8 +345,8 @@ int test_mem_host_no_access_image(cl_device_id deviceID, cl_context context,
cl_int err = CL_SUCCESS;
cl_bool image_support;
err = clGetDeviceInfo(deviceID, CL_DEVICE_IMAGE_SUPPORT,
sizeof image_support, &image_support, NULL);
err = clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof image_support,
&image_support, NULL);
if (err)
{
test_error(err, __FUNCTION__);
@@ -385,11 +378,11 @@ int test_mem_host_no_access_image(cl_device_id deviceID, cl_context context,
for (int p = 0; p < 3; p++)
{
err += test_mem_host_no_access_Image_RW(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
img_type[p], array_size[p], img_dims[p]);
err += test_mem_host_no_access_Image_RW_Mapping(
deviceID, context, queue, blocking[i], buffer_mem_flags[k],
device, context, queue, blocking[i], buffer_mem_flags[k],
img_type[p], array_size[p], img_dims[p]);
}
}

63
test_conformance/mem_host_flags/procs.h
View File

@@ -1,63 +0,0 @@
//
// Copyright (c) 2017 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#ifndef __PROCS_H__
#define __PROCS_H__
#include "testBase.h"
#define NUM_FLAGS 4
extern int test_mem_host_read_only_buffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_read_only_subbuffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_write_only_buffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_write_only_subbuffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_no_access_buffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_no_access_subbuffer(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_read_only_image(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_write_only_image(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_mem_host_no_access_image(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
#endif // #ifndef __PROCS_H__

1
test_conformance/non_uniform_work_group/TestNonUniformWorkGroup.h
View File

@@ -16,7 +16,6 @@
#ifndef TESTNONUNIFORMWORKGROUP_H
#define TESTNONUNIFORMWORKGROUP_H
#include "procs.h"
#include <vector>
#include "tools.h"
#include <algorithm>

28
test_conformance/non_uniform_work_group/main.cpp
View File

@@ -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,31 +13,10 @@
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "procs.h"
#include "tools.h"
#include "harness/testHarness.h"
#include "TestNonUniformWorkGroup.h"
test_definition test_list[] = {
ADD_TEST( non_uniform_1d_basic ),
ADD_TEST( non_uniform_1d_atomics ),
ADD_TEST( non_uniform_1d_barriers ),
ADD_TEST( non_uniform_2d_basic ),
ADD_TEST( non_uniform_2d_atomics ),
ADD_TEST( non_uniform_2d_barriers ),
ADD_TEST( non_uniform_3d_basic ),
ADD_TEST( non_uniform_3d_atomics ),
ADD_TEST( non_uniform_3d_barriers ),
ADD_TEST( non_uniform_other_basic ),
ADD_TEST( non_uniform_other_atomics ),
ADD_TEST( non_uniform_other_barriers ),
};
const int test_num = ARRAY_SIZE( test_list );
test_status InitCL(cl_device_id device) {
auto version = get_device_cl_version(device);
auto expected_min_version = Version(2, 0);
@@ -69,6 +48,9 @@ int main(int argc, const char *argv[])
PrimeNumbers::generatePrimeNumbers(100000);
return runTestHarnessWithCheck(static_cast<int>(programArgs.size()), &programArgs.front(), test_num, test_list, false, false, InitCL);
return runTestHarnessWithCheck(
static_cast<int>(programArgs.size()), &programArgs.front(),
test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), false, false, InitCL);
}

32
test_conformance/non_uniform_work_group/procs.h
View File

@@ -1,32 +0,0 @@
//
// Copyright (c) 2017 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "harness/typeWrappers.h"
extern int test_non_uniform_1d_basic(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_1d_atomics(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_1d_barriers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_2d_basic(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_2d_atomics(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_2d_barriers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_3d_basic(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_3d_atomics(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_3d_barriers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_other_basic(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_other_atomics(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);
extern int test_non_uniform_other_barriers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements);

12
test_conformance/non_uniform_work_group/test_advanced_2d.cpp
View File

@@ -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,13 +13,11 @@
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "procs.h"
#include "tools.h"
#include "TestNonUniformWorkGroup.h"
int
test_non_uniform_2d_basic(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_2d_basic)
{
SubTestExecutor exec(device, context, queue);
@@ -140,8 +138,7 @@ int
return exec.status();
}
int
test_non_uniform_2d_atomics(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_2d_atomics)
{
SubTestExecutor exec(device, context, queue);
@@ -262,8 +259,7 @@ int
return exec.status();
}
int
test_non_uniform_2d_barriers(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_2d_barriers)
{
SubTestExecutor exec(device, context, queue);

12
test_conformance/non_uniform_work_group/test_advanced_3d.cpp
View File

@@ -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,13 +13,11 @@
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "procs.h"
#include "tools.h"
#include "TestNonUniformWorkGroup.h"
int
test_non_uniform_3d_basic(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_3d_basic)
{
SubTestExecutor exec(device, context, queue);
@@ -148,8 +146,7 @@ int
return exec.status();
}
int
test_non_uniform_3d_atomics(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_3d_atomics)
{
SubTestExecutor exec(device, context, queue);
@@ -278,8 +275,7 @@ int
return exec.status();
}
int
test_non_uniform_3d_barriers(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_3d_barriers)
{
SubTestExecutor exec(device, context, queue);

12
test_conformance/non_uniform_work_group/test_advanced_other.cpp
View File

@@ -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,13 +13,11 @@
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "procs.h"
#include "tools.h"
#include "TestNonUniformWorkGroup.h"
int
test_non_uniform_other_basic(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_other_basic)
{
SubTestExecutor exec(device, context, queue);
@@ -105,8 +103,7 @@ int
return exec.status();
}
int
test_non_uniform_other_atomics(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_other_atomics)
{
SubTestExecutor exec(device, context, queue);
@@ -190,8 +187,7 @@ int
return exec.status();
}
int
test_non_uniform_other_barriers(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_other_barriers)
{
SubTestExecutor exec(device, context, queue);

12
test_conformance/non_uniform_work_group/test_basic.cpp
View File

@@ -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,13 +13,11 @@
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "procs.h"
#include "tools.h"
#include "TestNonUniformWorkGroup.h"
int
test_non_uniform_1d_basic(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_1d_basic)
{
SubTestExecutor exec(device, context, queue);
@@ -144,8 +142,7 @@ int
return exec.status();
}
int
test_non_uniform_1d_atomics(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_1d_atomics)
{
SubTestExecutor exec(device, context, queue);
@@ -270,8 +267,7 @@ int
return exec.status();
}
int
test_non_uniform_1d_barriers(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements)
REGISTER_TEST(non_uniform_1d_barriers)
{
SubTestExecutor exec(device, context, queue);

4
test_conformance/non_uniform_work_group/tools.h
View File

@@ -16,7 +16,9 @@
#ifndef TOOLS_H
#define TOOLS_H
#include "procs.h"
#include "harness/testHarness.h"
#include "harness/typeWrappers.h"
#include <vector>
#include <map>
#include <string>

16
test_conformance/thread_dimensions/main.cpp
View File

@@ -14,27 +14,16 @@
// limitations under the License.
//
#include "harness/compat.h"
#include "harness/testHarness.h"
#include <stdio.h>
#include <string.h>
#include "procs.h"
// Additional parameters to limit test scope (-n,-b,-x)
cl_uint maxThreadDimension = 0;
cl_uint bufferSize = 0;
cl_uint bufferStep = 0;
test_definition test_list[] = {
ADD_TEST(quick_1d_explicit_local), ADD_TEST(quick_2d_explicit_local),
ADD_TEST(quick_3d_explicit_local), ADD_TEST(quick_1d_implicit_local),
ADD_TEST(quick_2d_implicit_local), ADD_TEST(quick_3d_implicit_local),
ADD_TEST(full_1d_explicit_local), ADD_TEST(full_2d_explicit_local),
ADD_TEST(full_3d_explicit_local), ADD_TEST(full_1d_implicit_local),
ADD_TEST(full_2d_implicit_local), ADD_TEST(full_3d_implicit_local),
};
const int test_num = ARRAY_SIZE(test_list);
int main(int argc, const char *argv[])
{
int delArg = 0;
@@ -88,5 +77,6 @@ int main(int argc, const char *argv[])
i -= delArg;
}
return runTestHarness(argc, argv, test_num, test_list, false, 0);
return runTestHarness(argc, argv, test_registry::getInstance().num_tests(),
test_registry::getInstance().definitions(), false, 0);
}

72
test_conformance/thread_dimensions/procs.h
View File

@@ -1,72 +0,0 @@
//
// Copyright (c) 2017 The Khronos Group Inc.
//
// Licensed under the Apache License, Version 2.0 (the "License");
// you may not use this file except in compliance with the License.
// You may obtain a copy of the License at
//
// http://www.apache.org/licenses/LICENSE-2.0
//
// Unless required by applicable law or agreed to in writing, software
// distributed under the License is distributed on an "AS IS" BASIS,
// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
// See the License for the specific language governing permissions and
// limitations under the License.
//
#include "harness/kernelHelpers.h"
#include "harness/testHarness.h"
#include "harness/errorHelpers.h"
#include "harness/conversions.h"
#include "harness/mt19937.h"
extern const int kVectorSizeCount;
extern int test_quick_1d_explicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_quick_2d_explicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_quick_3d_explicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_quick_1d_implicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_quick_2d_implicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_quick_3d_implicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_full_1d_explicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_full_2d_explicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_full_3d_explicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_full_1d_implicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_full_2d_implicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_full_3d_implicit_local(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);

67
test_conformance/thread_dimensions/test_thread_dimensions.cpp
View File

@@ -14,6 +14,11 @@
// limitations under the License.
//
#include "harness/compat.h"
#include "harness/kernelHelpers.h"
#include "harness/testHarness.h"
#include "harness/errorHelpers.h"
#include "harness/conversions.h"
#include "harness/mt19937.h"
#include <stdio.h>
#include <string.h>
@@ -22,8 +27,6 @@
#include <cinttypes>
#include "procs.h"
#define ITERATIONS 4
#define DEBUG 0
@@ -1118,101 +1121,89 @@ int test_thread_dimensions(cl_device_id device, cl_context context,
#define QUICK 1
#define FULL 0
int test_quick_1d_explicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(quick_1d_explicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 1, 1,
device, context, queue, 1, 1,
maxThreadDimension ? maxThreadDimension : 65536 * 512, QUICK, 4, 1);
}
int test_quick_2d_explicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(quick_2d_explicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 2, 1,
device, context, queue, 2, 1,
maxThreadDimension ? maxThreadDimension : 65536 / 4, QUICK, 16, 1);
}
int test_quick_3d_explicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(quick_3d_explicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 3, 1,
device, context, queue, 3, 1,
maxThreadDimension ? maxThreadDimension : 1024, QUICK, 32, 1);
}
int test_quick_1d_implicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(quick_1d_implicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 1, 1,
device, context, queue, 1, 1,
maxThreadDimension ? maxThreadDimension : 65536 * 256, QUICK, 4, 0);
}
int test_quick_2d_implicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(quick_2d_implicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 2, 1,
device, context, queue, 2, 1,
maxThreadDimension ? maxThreadDimension : 65536 / 4, QUICK, 16, 0);
}
int test_quick_3d_implicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(quick_3d_implicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 3, 1,
device, context, queue, 3, 1,
maxThreadDimension ? maxThreadDimension : 1024, QUICK, 32, 0);
}
int test_full_1d_explicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(full_1d_explicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 1, 1,
device, context, queue, 1, 1,
maxThreadDimension ? maxThreadDimension : 65536 * 512, FULL, 4, 1);
}
int test_full_2d_explicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(full_2d_explicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 2, 1,
device, context, queue, 2, 1,
maxThreadDimension ? maxThreadDimension : 65536 / 4, FULL, 16, 1);
}
int test_full_3d_explicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(full_3d_explicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 3, 1,
device, context, queue, 3, 1,
maxThreadDimension ? maxThreadDimension : 1024, FULL, 32, 1);
}
int test_full_1d_implicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(full_1d_implicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 1, 1,
device, context, queue, 1, 1,
maxThreadDimension ? maxThreadDimension : 65536 * 256, FULL, 4, 0);
}
int test_full_2d_implicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(full_2d_implicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 2, 1,
device, context, queue, 2, 1,
maxThreadDimension ? maxThreadDimension : 65536 / 4, FULL, 16, 0);
}
int test_full_3d_implicit_local(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements)
REGISTER_TEST(full_3d_implicit_local)
{
return test_thread_dimensions(
deviceID, context, queue, 3, 1,
device, context, queue, 3, 1,
maxThreadDimension ? maxThreadDimension : 1024, FULL, 32, 0);
}

24
test_conformance/workgroups/main.cpp
View File

@@ -16,32 +16,12 @@
#include "harness/compat.h"
#include "harness/testHarness.h"
#include "procs.h"
#include <stdio.h>
#include <string.h>
#if !defined(_WIN32)
#include <unistd.h>
#endif
test_definition test_list[] = {
ADD_TEST_VERSION(work_group_all, Version(2, 0)),
ADD_TEST_VERSION(work_group_any, Version(2, 0)),
ADD_TEST_VERSION(work_group_reduce_add, Version(2, 0)),
ADD_TEST_VERSION(work_group_reduce_min, Version(2, 0)),
ADD_TEST_VERSION(work_group_reduce_max, Version(2, 0)),
ADD_TEST_VERSION(work_group_scan_inclusive_add, Version(2, 0)),
ADD_TEST_VERSION(work_group_scan_inclusive_min, Version(2, 0)),
ADD_TEST_VERSION(work_group_scan_inclusive_max, Version(2, 0)),
ADD_TEST_VERSION(work_group_scan_exclusive_add, Version(2, 0)),
ADD_TEST_VERSION(work_group_scan_exclusive_min, Version(2, 0)),
ADD_TEST_VERSION(work_group_scan_exclusive_max, Version(2, 0)),
ADD_TEST_VERSION(work_group_broadcast_1D, Version(2, 0)),
ADD_TEST_VERSION(work_group_broadcast_2D, Version(2, 0)),
ADD_TEST_VERSION(work_group_broadcast_3D, Version(2, 0)),
};
const int test_num = ARRAY_SIZE(test_list);
test_status InitCL(cl_device_id device) {
auto version = get_device_cl_version(device);
auto expected_min_version = Version(1, 2);
@@ -77,6 +57,8 @@ 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);
}

74
test_conformance/workgroups/procs.h
View File

@@ -1,74 +0,0 @@
//
// Copyright (c) 2017, 2021 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/testHarness.h"
#include "harness/kernelHelpers.h"
#include "harness/errorHelpers.h"
#include "harness/typeWrappers.h"
#include "harness/conversions.h"
#include "harness/mt19937.h"
extern int create_program_and_kernel(const char *source,
const char *kernel_name,
cl_program *program_ret,
cl_kernel *kernel_ret);
extern int test_work_group_all(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_work_group_any(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_work_group_broadcast_1D(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_broadcast_2D(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_broadcast_3D(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_reduce_add(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_work_group_reduce_min(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_work_group_reduce_max(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_work_group_scan_exclusive_add(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_scan_exclusive_min(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_scan_exclusive_max(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_scan_inclusive_add(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_scan_inclusive_min(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);
extern int test_work_group_scan_inclusive_max(cl_device_id deviceID,
cl_context context,
cl_command_queue queue,
int num_elements);

11
test_conformance/workgroups/testBase.h
View File

@@ -16,6 +16,12 @@
#ifndef _testBase_h
#define _testBase_h
#include "harness/testHarness.h"
#include "harness/kernelHelpers.h"
#include "harness/errorHelpers.h"
#include "harness/typeWrappers.h"
#include "harness/conversions.h"
#include "harness/mt19937.h"
#include "harness/compat.h"
#include <stdio.h>
@@ -23,9 +29,4 @@
#include <sys/types.h>
#include <sys/stat.h>
#include "procs.h"
#endif // _testBase_h

15
test_conformance/workgroups/test_wg_all.cpp
View File

@@ -20,8 +20,7 @@
#include <sys/types.h>
#include <sys/stat.h>
#include "procs.h"
#include "testBase.h"
const char *wg_all_kernel_code =
"__kernel void test_wg_all(global float *input, global int *output)\n"
@@ -65,8 +64,7 @@ verify_wg_all(float *inptr, int *outptr, size_t n, size_t wg_size)
return 0;
}
int
test_work_group_all(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_all, Version(2, 0))
{
cl_mem streams[2];
cl_float *input_ptr[1], *p;
@@ -74,8 +72,7 @@ test_work_group_all(cl_device_id device, cl_context context, cl_command_queue qu
cl_program program;
cl_kernel kernel;
size_t threads[1];
size_t wg_size[1];
size_t num_elements;
size_t wg_size[1];
int err;
MTdata d;
@@ -88,8 +85,6 @@ test_work_group_all(cl_device_id device, cl_context context, cl_command_queue qu
err = get_max_allowed_1d_work_group_size_on_device(device, kernel, wg_size);
test_error(err, "get_max_allowed_1d_work_group_size_on_device failed");
num_elements = n_elems;
input_ptr[0] = (cl_float*)malloc(sizeof(cl_float) * (num_elements+1));
output_ptr = (cl_int*)malloc(sizeof(cl_int) * (num_elements+1));
streams[0] =
@@ -111,7 +106,7 @@ test_work_group_all(cl_device_id device, cl_context context, cl_command_queue qu
p = input_ptr[0];
d = init_genrand( gRandomSeed );
for (size_t i = 0; i < (num_elements + 1); i++)
for (int i = 0; i < (num_elements + 1); i++)
{
p[i] = get_random_float((float)(-100000.f * M_PI), (float)(100000.f * M_PI) ,d);
}
@@ -133,7 +128,7 @@ test_work_group_all(cl_device_id device, cl_context context, cl_command_queue qu
}
// Line below is troublesome...
threads[0] = (size_t)n_elems;
threads[0] = (size_t)num_elements;
err = clEnqueueNDRangeKernel( queue, kernel, 1, NULL, threads, wg_size, 0, NULL, NULL );
if (err != CL_SUCCESS)
{

15
test_conformance/workgroups/test_wg_any.cpp
View File

@@ -20,8 +20,7 @@
#include <sys/types.h>
#include <sys/stat.h>
#include "procs.h"
#include "testBase.h"
const char *wg_any_kernel_code =
"__kernel void test_wg_any(global float *input, global int *output)\n"
@@ -65,8 +64,7 @@ verify_wg_any(float *inptr, int *outptr, size_t n, size_t wg_size)
return 0;
}
int
test_work_group_any(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_any, Version(2, 0))
{
cl_mem streams[2];
cl_float *input_ptr[1], *p;
@@ -74,8 +72,7 @@ test_work_group_any(cl_device_id device, cl_context context, cl_command_queue qu
cl_program program;
cl_kernel kernel;
size_t threads[1];
size_t wg_size[1];
size_t num_elements;
size_t wg_size[1];
int err;
MTdata d;
@@ -88,8 +85,6 @@ test_work_group_any(cl_device_id device, cl_context context, cl_command_queue qu
err = get_max_allowed_1d_work_group_size_on_device(device, kernel, wg_size);
test_error(err, "get_max_allowed_1d_work_group_size_on_device failed");
num_elements = n_elems;
input_ptr[0] = (cl_float*)malloc(sizeof(cl_float) * (num_elements+1));
output_ptr = (cl_int*)malloc(sizeof(cl_int) * (num_elements+1));
streams[0] =
@@ -111,7 +106,7 @@ test_work_group_any(cl_device_id device, cl_context context, cl_command_queue qu
p = input_ptr[0];
d = init_genrand( gRandomSeed );
for (size_t i = 0; i < (num_elements + 1); i++)
for (int i = 0; i < (num_elements + 1); i++)
{
p[i] = get_random_float((float)(-100000.f * M_PI), (float)(100000.f * M_PI) ,d);
}
@@ -133,7 +128,7 @@ test_work_group_any(cl_device_id device, cl_context context, cl_command_queue qu
}
// Line below is troublesome...
threads[0] = (size_t)n_elems;
threads[0] = (size_t)num_elements;
err = clEnqueueNDRangeKernel( queue, kernel, 1, NULL, threads, wg_size, 0, NULL, NULL );
if (err != CL_SUCCESS)
{

50
test_conformance/workgroups/test_wg_broadcast.cpp
View File

@@ -22,8 +22,7 @@
#include <algorithm>
#include "procs.h"
#include "testBase.h"
const char *wg_broadcast_1D_kernel_code =
"__kernel void test_wg_broadcast_1D(global float *input, global float *output)\n"
@@ -168,8 +167,7 @@ verify_wg_broadcast_3D(float *inptr, float *outptr, size_t nx, size_t ny, size_t
}
int
test_work_group_broadcast_1D(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_broadcast_1D, Version(2, 0))
{
cl_mem streams[2];
cl_float *input_ptr[1], *p;
@@ -177,8 +175,7 @@ test_work_group_broadcast_1D(cl_device_id device, cl_context context, cl_command
cl_program program;
cl_kernel kernel;
size_t globalsize[1];
size_t wg_size[1];
size_t num_elements;
size_t wg_size[1];
int err;
MTdata d;
@@ -192,8 +189,6 @@ test_work_group_broadcast_1D(cl_device_id device, cl_context context, cl_command
err = get_max_allowed_1d_work_group_size_on_device(device, kernel, wg_size);
test_error(err, "get_max_allowed_1d_work_group_size_on_device failed");
num_elements = n_elems;
input_ptr[0] = (cl_float*)malloc(sizeof(cl_float) * num_elements);
output_ptr = (cl_float*)malloc(sizeof(cl_float) * num_elements);
streams[0] = clCreateBuffer(context, CL_MEM_READ_WRITE,
@@ -214,7 +209,7 @@ test_work_group_broadcast_1D(cl_device_id device, cl_context context, cl_command
p = input_ptr[0];
d = init_genrand( gRandomSeed );
for (size_t i = 0; i < num_elements; i++)
for (int i = 0; i < num_elements; i++)
{
p[i] = get_random_float((float)(-100000.f * M_PI), (float)(100000.f * M_PI) ,d);
}
@@ -236,7 +231,7 @@ test_work_group_broadcast_1D(cl_device_id device, cl_context context, cl_command
}
// Line below is troublesome...
globalsize[0] = (size_t)n_elems;
globalsize[0] = (size_t)num_elements;
err = clEnqueueNDRangeKernel( queue, kernel, 1, NULL, globalsize, wg_size, 0, NULL, NULL );
if (err != CL_SUCCESS)
{
@@ -271,8 +266,7 @@ test_work_group_broadcast_1D(cl_device_id device, cl_context context, cl_command
}
int
test_work_group_broadcast_2D(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_broadcast_2D, Version(2, 0))
{
cl_mem streams[2];
cl_float *input_ptr[1], *p;
@@ -282,8 +276,7 @@ test_work_group_broadcast_2D(cl_device_id device, cl_context context, cl_command
size_t globalsize[2];
size_t localsize[2];
size_t wg_size[1];
size_t num_workgroups;
size_t num_elements;
size_t num_workgroups;
int err;
MTdata d;
@@ -314,7 +307,7 @@ test_work_group_broadcast_2D(cl_device_id device, cl_context context, cl_command
localsize[0] = localsize[1] = 1;
}
num_workgroups = std::max(n_elems / wg_size[0], (size_t)16);
num_workgroups = std::max(num_elements / wg_size[0], (size_t)16);
globalsize[0] = num_workgroups * localsize[0];
globalsize[1] = num_workgroups * localsize[1];
num_elements = globalsize[0] * globalsize[1];
@@ -339,7 +332,7 @@ test_work_group_broadcast_2D(cl_device_id device, cl_context context, cl_command
p = input_ptr[0];
d = init_genrand( gRandomSeed );
for (size_t i = 0; i < num_elements; i++)
for (int i = 0; i < num_elements; i++)
{
p[i] = get_random_float((float)(-100000.f * M_PI), (float)(100000.f * M_PI) ,d);
}
@@ -394,8 +387,7 @@ test_work_group_broadcast_2D(cl_device_id device, cl_context context, cl_command
}
int
test_work_group_broadcast_3D(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_broadcast_3D, Version(2, 0))
{
cl_mem streams[2];
cl_float *input_ptr[1], *p;
@@ -405,8 +397,7 @@ test_work_group_broadcast_3D(cl_device_id device, cl_context context, cl_command
size_t globalsize[3];
size_t localsize[3];
size_t wg_size[1];
size_t num_workgroups;
size_t num_elements;
size_t num_workgroups;
int err;
MTdata d;
@@ -437,7 +428,7 @@ test_work_group_broadcast_3D(cl_device_id device, cl_context context, cl_command
localsize[0] = localsize[1] = localsize[2] = 1;
}
num_workgroups = std::max(n_elems / wg_size[0], (size_t)8);
num_workgroups = std::max(num_elements / wg_size[0], (size_t)8);
globalsize[0] = num_workgroups * localsize[0];
globalsize[1] = num_workgroups * localsize[1];
globalsize[2] = num_workgroups * localsize[2];
@@ -463,7 +454,7 @@ test_work_group_broadcast_3D(cl_device_id device, cl_context context, cl_command
p = input_ptr[0];
d = init_genrand( gRandomSeed );
for (size_t i = 0; i < num_elements; i++)
for (int i = 0; i < num_elements; i++)
{
p[i] = get_random_float((float)(-100000.f * M_PI), (float)(100000.f * M_PI) ,d);
}
@@ -516,18 +507,3 @@ test_work_group_broadcast_3D(cl_device_id device, cl_context context, cl_command
return err;
}
int
test_work_group_broadcast(cl_device_id device, cl_context context, cl_command_queue queue, int n_elems)
{
int err;
err = test_work_group_broadcast_1D(device, context, queue, n_elems);
if (err) return err;
err = test_work_group_broadcast_2D(device, context, queue, n_elems);
if (err) return err;
return err;
}

205
test_conformance/workgroups/test_wg_scan_reduce.cpp
View File

@@ -19,7 +19,7 @@
#include <limits>
#include <vector>
#include "procs.h"
#include "testBase.h"
static std::string make_kernel_string(const std::string &type,
const std::string &kernelName,
@@ -272,184 +272,181 @@ static int run_test(cl_device_id device, cl_context context,
return TEST_PASS;
}
int test_work_group_reduce_add(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
{
int result = TEST_PASS;
result |= run_test<Reduce<Add<cl_int>>>(device, context, queue, n_elems);
result |= run_test<Reduce<Add<cl_uint>>>(device, context, queue, n_elems);
if (gHasLong)
{
result |=
run_test<Reduce<Add<cl_long>>>(device, context, queue, n_elems);
result |=
run_test<Reduce<Add<cl_ulong>>>(device, context, queue, n_elems);
}
return result;
}
int test_work_group_reduce_max(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
{
int result = TEST_PASS;
result |= run_test<Reduce<Max<cl_int>>>(device, context, queue, n_elems);
result |= run_test<Reduce<Max<cl_uint>>>(device, context, queue, n_elems);
if (gHasLong)
{
result |=
run_test<Reduce<Max<cl_long>>>(device, context, queue, n_elems);
result |=
run_test<Reduce<Max<cl_ulong>>>(device, context, queue, n_elems);
}
return result;
}
int test_work_group_reduce_min(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
{
int result = TEST_PASS;
result |= run_test<Reduce<Min<cl_int>>>(device, context, queue, n_elems);
result |= run_test<Reduce<Min<cl_uint>>>(device, context, queue, n_elems);
if (gHasLong)
{
result |=
run_test<Reduce<Min<cl_long>>>(device, context, queue, n_elems);
result |=
run_test<Reduce<Min<cl_ulong>>>(device, context, queue, n_elems);
}
return result;
}
int test_work_group_scan_inclusive_add(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_reduce_add, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<ScanInclusive<Add<cl_int>>>(device, context, queue, n_elems);
run_test<Reduce<Add<cl_int>>>(device, context, queue, num_elements);
result |=
run_test<ScanInclusive<Add<cl_uint>>>(device, context, queue, n_elems);
run_test<Reduce<Add<cl_uint>>>(device, context, queue, num_elements);
if (gHasLong)
{
result |= run_test<Reduce<Add<cl_long>>>(device, context, queue,
num_elements);
result |= run_test<Reduce<Add<cl_ulong>>>(device, context, queue,
num_elements);
}
return result;
}
REGISTER_TEST_VERSION(work_group_reduce_max, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<Reduce<Max<cl_int>>>(device, context, queue, num_elements);
result |=
run_test<Reduce<Max<cl_uint>>>(device, context, queue, num_elements);
if (gHasLong)
{
result |= run_test<Reduce<Max<cl_long>>>(device, context, queue,
num_elements);
result |= run_test<Reduce<Max<cl_ulong>>>(device, context, queue,
num_elements);
}
return result;
}
REGISTER_TEST_VERSION(work_group_reduce_min, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<Reduce<Min<cl_int>>>(device, context, queue, num_elements);
result |=
run_test<Reduce<Min<cl_uint>>>(device, context, queue, num_elements);
if (gHasLong)
{
result |= run_test<Reduce<Min<cl_long>>>(device, context, queue,
num_elements);
result |= run_test<Reduce<Min<cl_ulong>>>(device, context, queue,
num_elements);
}
return result;
}
REGISTER_TEST_VERSION(work_group_scan_inclusive_add, Version(2, 0))
{
int result = TEST_PASS;
result |= run_test<ScanInclusive<Add<cl_int>>>(device, context, queue,
num_elements);
result |= run_test<ScanInclusive<Add<cl_uint>>>(device, context, queue,
num_elements);
if (gHasLong)
{
result |= run_test<ScanInclusive<Add<cl_long>>>(device, context, queue,
n_elems);
num_elements);
result |= run_test<ScanInclusive<Add<cl_ulong>>>(device, context, queue,
n_elems);
num_elements);
}
return result;
}
int test_work_group_scan_inclusive_max(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_scan_inclusive_max, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<ScanInclusive<Max<cl_int>>>(device, context, queue, n_elems);
result |=
run_test<ScanInclusive<Max<cl_uint>>>(device, context, queue, n_elems);
result |= run_test<ScanInclusive<Max<cl_int>>>(device, context, queue,
num_elements);
result |= run_test<ScanInclusive<Max<cl_uint>>>(device, context, queue,
num_elements);
if (gHasLong)
{
result |= run_test<ScanInclusive<Max<cl_long>>>(device, context, queue,
n_elems);
num_elements);
result |= run_test<ScanInclusive<Max<cl_ulong>>>(device, context, queue,
n_elems);
num_elements);
}
return result;
}
int test_work_group_scan_inclusive_min(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_scan_inclusive_min, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<ScanInclusive<Min<cl_int>>>(device, context, queue, n_elems);
result |=
run_test<ScanInclusive<Min<cl_uint>>>(device, context, queue, n_elems);
result |= run_test<ScanInclusive<Min<cl_int>>>(device, context, queue,
num_elements);
result |= run_test<ScanInclusive<Min<cl_uint>>>(device, context, queue,
num_elements);
if (gHasLong)
{
result |= run_test<ScanInclusive<Min<cl_long>>>(device, context, queue,
n_elems);
num_elements);
result |= run_test<ScanInclusive<Min<cl_ulong>>>(device, context, queue,
n_elems);
num_elements);
}
return result;
}
int test_work_group_scan_exclusive_add(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_scan_exclusive_add, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<ScanExclusive<Add<cl_int>>>(device, context, queue, n_elems);
result |=
run_test<ScanExclusive<Add<cl_uint>>>(device, context, queue, n_elems);
result |= run_test<ScanExclusive<Add<cl_int>>>(device, context, queue,
num_elements);
result |= run_test<ScanExclusive<Add<cl_uint>>>(device, context, queue,
num_elements);
if (gHasLong)
{
result |= run_test<ScanExclusive<Add<cl_long>>>(device, context, queue,
n_elems);
num_elements);
result |= run_test<ScanExclusive<Add<cl_ulong>>>(device, context, queue,
n_elems);
num_elements);
}
return result;
}
int test_work_group_scan_exclusive_max(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_scan_exclusive_max, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<ScanExclusive<Max<cl_int>>>(device, context, queue, n_elems);
result |=
run_test<ScanExclusive<Max<cl_uint>>>(device, context, queue, n_elems);
result |= run_test<ScanExclusive<Max<cl_int>>>(device, context, queue,
num_elements);
result |= run_test<ScanExclusive<Max<cl_uint>>>(device, context, queue,
num_elements);
if (gHasLong)
{
result |= run_test<ScanExclusive<Max<cl_long>>>(device, context, queue,
n_elems);
num_elements);
result |= run_test<ScanExclusive<Max<cl_ulong>>>(device, context, queue,
n_elems);
num_elements);
}
return result;
}
int test_work_group_scan_exclusive_min(cl_device_id device, cl_context context,
cl_command_queue queue, int n_elems)
REGISTER_TEST_VERSION(work_group_scan_exclusive_min, Version(2, 0))
{
int result = TEST_PASS;
result |=
run_test<ScanExclusive<Min<cl_int>>>(device, context, queue, n_elems);
result |=
run_test<ScanExclusive<Min<cl_uint>>>(device, context, queue, n_elems);
result |= run_test<ScanExclusive<Min<cl_int>>>(device, context, queue,
num_elements);
result |= run_test<ScanExclusive<Min<cl_uint>>>(device, context, queue,
num_elements);
if (gHasLong)
{
result |= run_test<ScanExclusive<Min<cl_long>>>(device, context, queue,
n_elems);
num_elements);
result |= run_test<ScanExclusive<Min<cl_ulong>>>(device, context, queue,
n_elems);
num_elements);
}
return result;