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

add checks for OpenCL C versions (#968)

Browse Source 
* add test_compiler test for OpenCL C versions

* add checks for required OpenCL C versions

* bugfix: only OpenCL C 2.0 is required for OpenCL 2.1 and 2.2

* pass reference to feature struct vs. pointer

* address review comments regarding C++ identifiers and 3d image writes
This commit is contained in:
Ben Ashbaugh
2020-09-27 16:26:57 -07:00
committed by GitHub
parent 951d010eaf
commit beaecb71a3
7 changed files with 439 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files

1
test_common/CMakeLists.txt
View File

@@ -8,6 +8,7 @@ set(HARNESS_SOURCES
harness/msvc9.c
harness/crc32.cpp
harness/errorHelpers.cpp
harness/featureHelpers.cpp
harness/genericThread.cpp
harness/imageHelpers.cpp
harness/kernelHelpers.cpp

75
test_common/harness/featureHelpers.cpp Normal file
View File

@@ -0,0 +1,75 @@
//
// Copyright (c) 2020 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 "featureHelpers.h"
#include "errorHelpers.h"
#include <assert.h>
#include <string.h>
#include <vector>
int get_device_cl_c_features(cl_device_id device, OpenCLCFeatures& features)
{
// Initially, all features are unsupported.
features = { 0 };
// The CL_DEVICE_OPENCL_C_FEATURES query does not exist pre-3.0.
const Version version = get_device_cl_version(device);
if (version < Version(3, 0))
{
return TEST_PASS;
}
cl_int error = CL_SUCCESS;
size_t sz = 0;
error = clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_FEATURES, 0, NULL, &sz);
test_error(error, "Unable to query CL_DEVICE_OPENCL_C_FEATURES size");
std::vector<cl_name_version> clc_features(sz / sizeof(cl_name_version));
error = clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_FEATURES, sz,
clc_features.data(), NULL);
test_error(error, "Unable to query CL_DEVICE_OPENCL_C_FEATURES");
#define CHECK_OPENCL_C_FEATURE(_feature) \
if (strcmp(clc_feature.name, #_feature) == 0) \
{ \
features.supports##_feature = true; \
}
for (const auto& clc_feature : clc_features)
{
CHECK_OPENCL_C_FEATURE(__opencl_c_3d_image_writes);
CHECK_OPENCL_C_FEATURE(__opencl_c_atomic_order_acq_rel);
CHECK_OPENCL_C_FEATURE(__opencl_c_atomic_order_seq_cst);
CHECK_OPENCL_C_FEATURE(__opencl_c_atomic_scope_device);
CHECK_OPENCL_C_FEATURE(__opencl_c_atomic_scope_all_devices);
CHECK_OPENCL_C_FEATURE(__opencl_c_device_enqueue);
CHECK_OPENCL_C_FEATURE(__opencl_c_generic_address_space);
CHECK_OPENCL_C_FEATURE(__opencl_c_fp64);
CHECK_OPENCL_C_FEATURE(__opencl_c_images);
CHECK_OPENCL_C_FEATURE(__opencl_c_int64);
CHECK_OPENCL_C_FEATURE(__opencl_c_pipes);
CHECK_OPENCL_C_FEATURE(__opencl_c_program_scope_global_variables);
CHECK_OPENCL_C_FEATURE(__opencl_c_read_write_images);
CHECK_OPENCL_C_FEATURE(__opencl_c_subgroups);
CHECK_OPENCL_C_FEATURE(__opencl_c_work_group_collective_functions);
}
#undef CHECK_OPENCL_C_FEATURE
return TEST_PASS;
}

43
test_common/harness/featureHelpers.h Normal file
View File

@@ -0,0 +1,43 @@
//
// Copyright (c) 2020 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 _featureHelpers_h
#define _featureHelpers_h
#include "compat.h"
#include "testHarness.h"
struct OpenCLCFeatures
{
bool supports__opencl_c_3d_image_writes;
bool supports__opencl_c_atomic_order_acq_rel;
bool supports__opencl_c_atomic_order_seq_cst;
bool supports__opencl_c_atomic_scope_device;
bool supports__opencl_c_atomic_scope_all_devices;
bool supports__opencl_c_device_enqueue;
bool supports__opencl_c_generic_address_space;
bool supports__opencl_c_fp64;
bool supports__opencl_c_images;
bool supports__opencl_c_int64;
bool supports__opencl_c_pipes;
bool supports__opencl_c_program_scope_global_variables;
bool supports__opencl_c_read_write_images;
bool supports__opencl_c_subgroups;
bool supports__opencl_c_work_group_collective_functions;
};
int get_device_cl_c_features(cl_device_id device, OpenCLCFeatures& features);
#endif // _featureHelpers_h

1
test_conformance/compiler/CMakeLists.txt
View File

@@ -7,6 +7,7 @@ set(${MODULE_NAME}_SOURCES
test_async_build.cpp
test_build_options.cpp
test_preprocessor.cpp
test_opencl_c_versions.cpp
test_image_macro.cpp
test_compiler_defines_for_extensions.cpp
test_pragma_unroll.cpp

1
test_conformance/compiler/main.cpp
View File

@@ -50,6 +50,7 @@ test_definition test_list[] = {
ADD_TEST(preprocessor_line_error),
ADD_TEST(preprocessor_pragma),
ADD_TEST(opencl_c_versions),
ADD_TEST(compiler_defines_for_extensions),
ADD_TEST(image_macro),

2
test_conformance/compiler/procs.h
View File

@@ -88,6 +88,8 @@ extern int test_preprocessor_line_error(cl_device_id deviceID,
extern int test_preprocessor_pragma(cl_device_id deviceID, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_opencl_c_versions(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements);
extern int test_compiler_defines_for_extensions(cl_device_id device,
cl_context context,
cl_command_queue queue,

316
test_conformance/compiler/test_opencl_c_versions.cpp Normal file
View File

@@ -0,0 +1,316 @@
//
// Copyright (c) 2020 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 "testBase.h"
#include "harness/featureHelpers.h"
#include <vector>
static const char* test_kernel = R"CLC(
__kernel void test(__global int* dst) {
dst[0] = 0;
}
)CLC";
// This sub-test checks that CL_DEVICE_OPENCL_C_VERSION meets any API
// requirements and that programs can be built for the reported OpenCL C version
// and all previous versions.
static int test_CL_DEVICE_OPENCL_C_VERSION(cl_device_id device,
cl_context context)
{
const Version latest_version = Version(3, 0);
const Version api_version = get_device_cl_version(device);
const Version clc_version = get_device_cl_c_version(device);
if (api_version > latest_version)
{
log_info("CL_DEVICE_VERSION is %s, which is bigger than %s.\n"
"Need to update the opencl_c_versions test!\n",
api_version.to_string().c_str(),
latest_version.to_string().c_str());
}
if (clc_version > latest_version)
{
log_info("CL_DEVICE_OPENCL_C_VERSION is %s, which is bigger than %s.\n"
"Need to update the opencl_c_versions test!\n",
clc_version.to_string().c_str(),
latest_version.to_string().c_str());
}
// For OpenCL 3.0, the minimum required OpenCL C version is OpenCL C 1.2.
// For OpenCL 2.x, the minimum required OpenCL C version is OpenCL C 2.0.
// For other OpenCL versions, the minimum required OpenCL C version is
// the same as the API version.
const Version min_clc_version = api_version == Version(3, 0)
? Version(1, 2)
: api_version >= Version(2, 0) ? Version(2, 0) : api_version;
if (clc_version < min_clc_version)
{
log_error("The minimum required OpenCL C version for API version %s is "
"%s (got %s)!\n",
api_version.to_string().c_str(),
min_clc_version.to_string().c_str(),
clc_version.to_string().c_str());
return TEST_FAIL;
}
log_info(" testing compilation based on CL_DEVICE_OPENCL_C_VERSION\n");
struct TestCase
{
Version version;
const char* buildOptions;
};
std::vector<TestCase> tests;
tests.push_back({ Version(1, 1), "-cl-std=CL1.1" });
tests.push_back({ Version(1, 2), "-cl-std=CL1.2" });
tests.push_back({ Version(2, 0), "-cl-std=CL2.0" });
tests.push_back({ Version(3, 0), "-cl-std=CL3.0" });
for (const auto& testcase : tests)
{
if (clc_version >= testcase.version)
{
clProgramWrapper program;
cl_int error =
create_single_kernel_helper_create_program_for_device(
context, device, &program, 1, &test_kernel,
testcase.buildOptions);
test_error(error, "Unable to build program!");
log_info(" successfully built program with build options '%s'\n",
testcase.buildOptions);
}
}
return TEST_PASS;
}
// This sub-test checks that CL_DEVICE_OPENCL_C_ALL_VERSIONS includes any
// requirements for the API version, and that programs can be built for all
// reported versions.
static int test_CL_DEVICE_OPENCL_C_ALL_VERSIONS(cl_device_id device,
cl_context context)
{
// For now, the required OpenCL C version is the same as the API version.
const Version api_version = get_device_cl_version(device);
bool found_api_version = false;
log_info(
" testing compilation based on CL_DEVICE_OPENCL_C_ALL_VERSIONS\n");
cl_int error = CL_SUCCESS;
size_t sz = 0;
error =
clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, 0, NULL, &sz);
test_error(error, "Unable to query CL_DEVICE_OPENCL_C_ALL_VERSIONS size");
std::vector<cl_name_version> clc_versions(sz / sizeof(cl_name_version));
error = clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, sz,
clc_versions.data(), NULL);
test_error(error, "Unable to query CL_DEVICE_OPENCL_C_FEATURES");
for (const auto& clc_version : clc_versions)
{
const unsigned major = CL_VERSION_MAJOR(clc_version.version);
const unsigned minor = CL_VERSION_MINOR(clc_version.version);
if (strcmp(clc_version.name, "OpenCL C") == 0)
{
if (api_version == Version(major, minor))
{
found_api_version = true;
}
if (major == 1 && minor == 0)
{
log_info(
" skipping OpenCL C 1.0, there is no -cl-std=CL1.0.\n");
continue;
}
std::string buildOptions = "-cl-std=CL";
buildOptions += std::to_string(major);
buildOptions += ".";
buildOptions += std::to_string(minor);
clProgramWrapper program;
error = create_single_kernel_helper_create_program_for_device(
context, device, &program, 1, &test_kernel,
buildOptions.c_str());
test_error(error, "Unable to build program!");
log_info(" successfully built program with build options '%s'\n",
buildOptions.c_str());
}
else
{
log_error(" unknown OpenCL C name '%s'.\n", clc_version.name);
return TEST_FAIL;
}
}
if (!found_api_version)
{
log_error(" didn't find required OpenCL C version '%s'!\n",
api_version.to_string().c_str());
return TEST_FAIL;
}
return TEST_PASS;
}
// This sub-test checks that any required features are present for a specific
// CL_DEVICE_OPENCL_C_VERSION.
static int test_CL_DEVICE_OPENCL_C_VERSION_features(cl_device_id device,
cl_context context)
{
log_info(" testing for OPENCL_C_VERSION required features\n");
OpenCLCFeatures features;
int error = get_device_cl_c_features(device, features);
if (error)
{
log_error("Couldn't query OpenCL C features for the device!\n");
return TEST_FAIL;
}
const Version clc_version = get_device_cl_c_version(device);
if (clc_version >= Version(2, 0))
{
bool has_all_OpenCL_C_20_features =
features.supports__opencl_c_atomic_order_acq_rel
&& features.supports__opencl_c_atomic_order_seq_cst
&& features.supports__opencl_c_atomic_scope_device
&& features.supports__opencl_c_atomic_scope_all_devices
&& features.supports__opencl_c_device_enqueue
&& features.supports__opencl_c_generic_address_space
&& features.supports__opencl_c_pipes
&& features.supports__opencl_c_program_scope_global_variables
&& features.supports__opencl_c_work_group_collective_functions;
if (features.supports__opencl_c_images)
{
has_all_OpenCL_C_20_features = has_all_OpenCL_C_20_features
&& features.supports__opencl_c_3d_image_writes
&& features.supports__opencl_c_read_write_images;
}
test_assert_error(
has_all_OpenCL_C_20_features,
"At least one required OpenCL C 2.0 feature is missing!");
}
return TEST_PASS;
}
// This sub-test checks that all required OpenCL C versions are present for a
// specific CL_DEVICE_OPENCL_C_VERSION.
static int test_CL_DEVICE_OPENCL_C_VERSION_versions(cl_device_id device,
cl_context context)
{
log_info(" testing for OPENCL_C_VERSION required versions\n");
const Version device_clc_version = get_device_cl_c_version(device);
std::vector<Version> test_clc_versions;
test_clc_versions.push_back(Version(1, 0));
test_clc_versions.push_back(Version(1, 1));
test_clc_versions.push_back(Version(1, 2));
test_clc_versions.push_back(Version(2, 0));
test_clc_versions.push_back(Version(3, 0));
cl_int error = CL_SUCCESS;
size_t sz = 0;
error =
clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, 0, NULL, &sz);
test_error(error, "Unable to query CL_DEVICE_OPENCL_C_ALL_VERSIONS size");
std::vector<cl_name_version> device_clc_versions(sz
/ sizeof(cl_name_version));
error = clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, sz,
device_clc_versions.data(), NULL);
test_error(error, "Unable to query CL_DEVICE_OPENCL_C_FEATURES");
for (const auto& test_clc_version : test_clc_versions)
{
if (device_clc_version >= test_clc_version)
{
bool found = false;
for (const auto& check : device_clc_versions)
{
const unsigned major = CL_VERSION_MAJOR(check.version);
const unsigned minor = CL_VERSION_MINOR(check.version);
if (strcmp(check.name, "OpenCL C") == 0
&& test_clc_version == Version(major, minor))
{
found = true;
break;
}
}
if (found)
{
log_info(" found OpenCL C version '%s'\n",
test_clc_version.to_string().c_str());
}
else
{
log_error("Didn't find OpenCL C version '%s'!\n",
test_clc_version.to_string().c_str());
return TEST_FAIL;
}
}
}
return TEST_PASS;
}
int test_opencl_c_versions(cl_device_id device, cl_context context,
cl_command_queue queue, int num_elements)
{
cl_bool compilerAvailable = CL_FALSE;
cl_int error =
clGetDeviceInfo(device, CL_DEVICE_COMPILER_AVAILABLE,
sizeof(compilerAvailable), &compilerAvailable, NULL);
if (compilerAvailable == CL_FALSE)
{
log_info("Skipping test - no compiler is available.\n");
return TEST_SKIPPED_ITSELF;
}
const Version version = get_device_cl_version(device);
int result = TEST_PASS;
result |= test_CL_DEVICE_OPENCL_C_VERSION(device, context);
if (version >= Version(3, 0))
{
result |= test_CL_DEVICE_OPENCL_C_ALL_VERSIONS(device, context);
result |= test_CL_DEVICE_OPENCL_C_VERSION_features(device, context);
result |= test_CL_DEVICE_OPENCL_C_VERSION_versions(device, context);
}
return result;
}