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

Report unsupported extended subgroup tests as skipped rather than passed (#1301)

Browse Source 
* Report unsupported extended subgroup tests as skipped rather than passed

Also don't check the presence of extensions for each sub-test.

Signed-off-by: Kévin Petit <kpet@free.fr>

* address review comments
This commit is contained in:
Kévin Petit
2021-09-30 13:33:18 +01:00
committed by GitHub
parent 2b770c4f34
commit 903f1bf65d
8 changed files with 52 additions and 49 deletions
Download Patch File Download Diff File Expand all files Collapse all files

17
test_conformance/subgroups/subhelpers.h
View File

@@ -33,10 +33,9 @@ extern cl_half_rounding_mode g_rounding_mode;
struct WorkGroupParams
{
WorkGroupParams(size_t gws, size_t lws,
const std::vector<std::string> &req_ext = {},
const std::vector<uint32_t> &all_wim = {})
: global_workgroup_size(gws), local_workgroup_size(lws),
required_extensions(req_ext), all_work_item_masks(all_wim)
all_work_item_masks(all_wim)
{
subgroup_size = 0;
work_items_mask = 0;
@@ -49,7 +48,6 @@ struct WorkGroupParams
uint32_t work_items_mask;
int dynsc;
bool use_core_subgroups;
std::vector<std::string> required_extensions;
std::vector<uint32_t> all_work_item_masks;
};
@@ -1297,19 +1295,6 @@ template <typename Ty, typename Fns, size_t TSIZE = 0> struct test
}
}
for (std::string extension : test_params.required_extensions)
{
if (!is_extension_available(device, extension.c_str()))
{
log_info("The extension %s not supported on this device. SKIP "
"testing - kernel %s data type %s\n",
extension.c_str(), kname, TypeManager<Ty>::name());
return TEST_PASS;
}
kernel_sstr << "#pragma OPENCL EXTENSION " + extension
+ ": enable\n";
}
error = clGetDeviceInfo(device, CL_DEVICE_PLATFORM, sizeof(platform),
(void *)&platform, NULL);
test_error(error, "clGetDeviceInfo failed for CL_DEVICE_PLATFORM");

10
test_conformance/subgroups/test_subgroup_ballot.cpp
View File

@@ -926,11 +926,15 @@ template <typename T> int run_non_uniform_broadcast_for_type(RunTestForType rft)
int test_subgroup_functions_ballot(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements)
{
std::vector<std::string> required_extensions = { "cl_khr_subgroup_ballot" };
if (!is_extension_available(device, "cl_khr_subgroup_ballot"))
{
log_info("cl_khr_subgroup_ballot is not supported on this device, "
"skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
constexpr size_t global_work_size = 170;
constexpr size_t local_work_size = 64;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions);
WorkGroupParams test_params(global_work_size, local_work_size);
RunTestForType rft(device, context, queue, num_elements, test_params);
// non uniform broadcast functions

12
test_conformance/subgroups/test_subgroup_clustered_reduce.cpp
View File

@@ -305,13 +305,15 @@ int test_subgroup_functions_clustered_reduce(cl_device_id device,
cl_command_queue queue,
int num_elements)
{
std::vector<std::string> required_extensions = {
"cl_khr_subgroup_clustered_reduce"
};
if (!is_extension_available(device, "cl_khr_subgroup_clustered_reduce"))
{
log_info("cl_khr_subgroup_clustered_reduce is not supported on this "
"device, skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
constexpr size_t global_work_size = 2000;
constexpr size_t local_work_size = 200;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions);
WorkGroupParams test_params(global_work_size, local_work_size);
RunTestForType rft(device, context, queue, num_elements, test_params);
int error = run_cluster_red_add_max_min_mul_for_type<cl_int>(rft);

12
test_conformance/subgroups/test_subgroup_extended_types.cpp
View File

@@ -59,13 +59,15 @@ int test_subgroup_functions_extended_types(cl_device_id device,
cl_command_queue queue,
int num_elements)
{
std::vector<std::string> required_extensions = {
"cl_khr_subgroup_extended_types"
};
if (!is_extension_available(device, "cl_khr_subgroup_extended_types"))
{
log_info("cl_khr_subgroup_extended_types is not supported on this "
"device, skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
constexpr size_t global_work_size = 2000;
constexpr size_t local_work_size = 200;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions);
WorkGroupParams test_params(global_work_size, local_work_size);
RunTestForType rft(device, context, queue, num_elements, test_params);
int error = run_broadcast_for_extended_type<cl_uint2>(rft);

15
test_conformance/subgroups/test_subgroup_non_uniform_arithmetic.cpp
View File

@@ -434,17 +434,20 @@ int test_subgroup_functions_non_uniform_arithmetic(cl_device_id device,
cl_command_queue queue,
int num_elements)
{
std::vector<std::string> required_extensions = {
"cl_khr_subgroup_non_uniform_arithmetic"
};
if (!is_extension_available(device,
"cl_khr_subgroup_non_uniform_arithmetic"))
{
log_info("cl_khr_subgroup_non_uniform_arithmetic is not supported on "
"this device, skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
std::vector<uint32_t> masks{ 0xffffffff, 0x55aaaa55, 0x5555aaaa, 0xaaaa5555,
0x0f0ff0f0, 0x0f0f0f0f, 0xff0000ff, 0xff00ff00,
0x00ffff00, 0x80000000, 0xaaaaaaaa };
constexpr size_t global_work_size = 2000;
constexpr size_t local_work_size = 200;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions, masks);
WorkGroupParams test_params(global_work_size, local_work_size, masks);
RunTestForType rft(device, context, queue, num_elements, test_params);
int error = run_functions_add_mul_max_min_for_type<cl_int>(rft);
@@ -470,4 +473,4 @@ int test_subgroup_functions_non_uniform_arithmetic(cl_device_id device,
error |= run_functions_logical_and_or_xor_for_type<cl_int>(rft);
return error;
}
}

13
test_conformance/subgroups/test_subgroup_non_uniform_vote.cpp
View File

@@ -272,17 +272,18 @@ int test_subgroup_functions_non_uniform_vote(cl_device_id device,
cl_command_queue queue,
int num_elements)
{
std::vector<std::string> required_extensions = {
"cl_khr_subgroup_non_uniform_vote"
};
if (!is_extension_available(device, "cl_khr_subgroup_non_uniform_vote"))
{
log_info("cl_khr_subgroup_non_uniform_vote is not supported on this "
"device, skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
std::vector<uint32_t> masks{ 0xffffffff, 0x55aaaa55, 0x5555aaaa, 0xaaaa5555,
0x0f0ff0f0, 0x0f0f0f0f, 0xff0000ff, 0xff00ff00,
0x00ffff00, 0x80000000 };
constexpr size_t global_work_size = 170;
constexpr size_t local_work_size = 64;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions, masks);
WorkGroupParams test_params(global_work_size, local_work_size, masks);
RunTestForType rft(device, context, queue, num_elements, test_params);
int error = run_vote_all_equal_for_type<cl_int>(rft);

10
test_conformance/subgroups/test_subgroup_shuffle.cpp
View File

@@ -55,11 +55,15 @@ template <typename T> int run_shuffle_for_type(RunTestForType rft)
int test_subgroup_functions_shuffle(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements)
{
std::vector<std::string> required_extensions{ "cl_khr_subgroup_shuffle" };
if (!is_extension_available(device, "cl_khr_subgroup_shuffle"))
{
log_info("cl_khr_subgroup_shuffle is not supported on this device, "
"skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
constexpr size_t global_work_size = 2000;
constexpr size_t local_work_size = 200;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions);
WorkGroupParams test_params(global_work_size, local_work_size);
RunTestForType rft(device, context, queue, num_elements, test_params);
int error = run_shuffle_for_type<cl_int>(rft);

12
test_conformance/subgroups/test_subgroup_shuffle_relative.cpp
View File

@@ -56,13 +56,15 @@ int test_subgroup_functions_shuffle_relative(cl_device_id device,
cl_command_queue queue,
int num_elements)
{
std::vector<std::string> required_extensions = {
"cl_khr_subgroup_shuffle_relative"
};
if (!is_extension_available(device, "cl_khr_subgroup_shuffle_relative"))
{
log_info("cl_khr_subgroup_shuffle_relative is not supported on this "
"device, skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
constexpr size_t global_work_size = 2000;
constexpr size_t local_work_size = 200;
WorkGroupParams test_params(global_work_size, local_work_size,
required_extensions);
WorkGroupParams test_params(global_work_size, local_work_size);
RunTestForType rft(device, context, queue, num_elements, test_params);
int error = run_shuffle_relative_for_type<cl_int>(rft);