• Home
  • Line#
  • Scopes#
  • Navigate#
  • Raw
  • Download
1 //
2 // Copyright (c) 2020 The Khronos Group Inc.
3 //
4 // Licensed under the Apache License, Version 2.0 (the "License");
5 // you may not use this file except in compliance with the License.
6 // You may obtain a copy of the License at
7 //
8 //    http://www.apache.org/licenses/LICENSE-2.0
9 //
10 // Unless required by applicable law or agreed to in writing, software
11 // distributed under the License is distributed on an "AS IS" BASIS,
12 // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13 // See the License for the specific language governing permissions and
14 // limitations under the License.
15 //
16 
17 #include "testBase.h"
18 #include "harness/featureHelpers.h"
19 
20 #include <vector>
21 
22 static const char* test_kernel = R"CLC(
23 __kernel void test(__global int* dst) {
24     dst[0] = 0;
25 }
26 )CLC";
27 
28 // This sub-test checks that CL_DEVICE_OPENCL_C_VERSION meets any API
29 // requirements and that programs can be built for the reported OpenCL C version
30 // and all previous versions.
test_CL_DEVICE_OPENCL_C_VERSION(cl_device_id device,cl_context context)31 static int test_CL_DEVICE_OPENCL_C_VERSION(cl_device_id device,
32                                            cl_context context)
33 {
34     const Version latest_version = Version(3, 0);
35 
36     const Version api_version = get_device_cl_version(device);
37     const Version clc_version = get_device_cl_c_version(device);
38 
39     if (api_version > latest_version)
40     {
41         log_info("CL_DEVICE_VERSION is %s, which is bigger than %s.\n"
42                  "Need to update the opencl_c_versions test!\n",
43                  api_version.to_string().c_str(),
44                  latest_version.to_string().c_str());
45     }
46 
47     if (clc_version > latest_version)
48     {
49         log_info("CL_DEVICE_OPENCL_C_VERSION is %s, which is bigger than %s.\n"
50                  "Need to update the opencl_c_versions test!\n",
51                  clc_version.to_string().c_str(),
52                  latest_version.to_string().c_str());
53     }
54 
55     // For OpenCL 3.0, the minimum required OpenCL C version is OpenCL C 1.2.
56     // For OpenCL 2.x, the minimum required OpenCL C version is OpenCL C 2.0.
57     // For other OpenCL versions, the minimum required OpenCL C version is
58     // the same as the API version.
59     const Version min_clc_version = api_version == Version(3, 0)
60         ? Version(1, 2)
61         : api_version >= Version(2, 0) ? Version(2, 0) : api_version;
62     if (clc_version < min_clc_version)
63     {
64         log_error("The minimum required OpenCL C version for API version %s is "
65                   "%s (got %s)!\n",
66                   api_version.to_string().c_str(),
67                   min_clc_version.to_string().c_str(),
68                   clc_version.to_string().c_str());
69         return TEST_FAIL;
70     }
71 
72     log_info("  testing compilation based on CL_DEVICE_OPENCL_C_VERSION\n");
73 
74     struct TestCase
75     {
76         Version version;
77         const char* buildOptions;
78     };
79 
80     std::vector<TestCase> tests;
81     tests.push_back({ Version(1, 1), "-cl-std=CL1.1" });
82     tests.push_back({ Version(1, 2), "-cl-std=CL1.2" });
83     tests.push_back({ Version(2, 0), "-cl-std=CL2.0" });
84     tests.push_back({ Version(3, 0), "-cl-std=CL3.0" });
85 
86     for (const auto& testcase : tests)
87     {
88         if (clc_version >= testcase.version)
89         {
90             clProgramWrapper program;
91             cl_int error =
92                 create_single_kernel_helper_create_program_for_device(
93                     context, device, &program, 1, &test_kernel,
94                     testcase.buildOptions);
95             test_error(error, "Unable to build program!");
96 
97             log_info("    successfully built program with build options '%s'\n",
98                      testcase.buildOptions);
99         }
100     }
101 
102     return TEST_PASS;
103 }
104 
105 // This sub-test checks that CL_DEVICE_OPENCL_C_ALL_VERSIONS includes any
106 // requirements for the API version, and that programs can be built for all
107 // reported versions.
test_CL_DEVICE_OPENCL_C_ALL_VERSIONS(cl_device_id device,cl_context context)108 static int test_CL_DEVICE_OPENCL_C_ALL_VERSIONS(cl_device_id device,
109                                                 cl_context context)
110 {
111     // For now, the required OpenCL C version is the same as the API version.
112     const Version api_version = get_device_cl_version(device);
113     bool found_api_version = false;
114 
115     log_info(
116         "  testing compilation based on CL_DEVICE_OPENCL_C_ALL_VERSIONS\n");
117 
118     cl_int error = CL_SUCCESS;
119 
120     size_t sz = 0;
121     error =
122         clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, 0, NULL, &sz);
123     test_error(error, "Unable to query CL_DEVICE_OPENCL_C_ALL_VERSIONS size");
124 
125     std::vector<cl_name_version> clc_versions(sz / sizeof(cl_name_version));
126     error = clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, sz,
127                             clc_versions.data(), NULL);
128     test_error(error, "Unable to query CL_DEVICE_OPENCL_C_FEATURES");
129 
130     for (const auto& clc_version : clc_versions)
131     {
132         const unsigned major = CL_VERSION_MAJOR(clc_version.version);
133         const unsigned minor = CL_VERSION_MINOR(clc_version.version);
134 
135         if (strcmp(clc_version.name, "OpenCL C") == 0)
136         {
137             if (api_version == Version(major, minor))
138             {
139                 found_api_version = true;
140             }
141 
142             if (major == 1 && minor == 0)
143             {
144                 log_info(
145                     "    skipping OpenCL C 1.0, there is no -cl-std=CL1.0.\n");
146                 continue;
147             }
148 
149             std::string buildOptions = "-cl-std=CL";
150             buildOptions += std::to_string(major);
151             buildOptions += ".";
152             buildOptions += std::to_string(minor);
153 
154             clProgramWrapper program;
155             error = create_single_kernel_helper_create_program_for_device(
156                 context, device, &program, 1, &test_kernel,
157                 buildOptions.c_str());
158             test_error(error, "Unable to build program!");
159 
160             log_info("    successfully built program with build options '%s'\n",
161                      buildOptions.c_str());
162         }
163         else
164         {
165             log_error("    unknown OpenCL C name '%s'.\n", clc_version.name);
166             return TEST_FAIL;
167         }
168     }
169 
170     if (!found_api_version)
171     {
172         log_error("    didn't find required OpenCL C version '%s'!\n",
173                   api_version.to_string().c_str());
174         return TEST_FAIL;
175     }
176 
177     return TEST_PASS;
178 }
179 
180 // This sub-test checks that any required features are present for a specific
181 // CL_DEVICE_OPENCL_C_VERSION.
test_CL_DEVICE_OPENCL_C_VERSION_features(cl_device_id device,cl_context context)182 static int test_CL_DEVICE_OPENCL_C_VERSION_features(cl_device_id device,
183                                                     cl_context context)
184 {
185     log_info("  testing for OPENCL_C_VERSION required features\n");
186 
187     OpenCLCFeatures features;
188     int error = get_device_cl_c_features(device, features);
189     if (error)
190     {
191         log_error("Couldn't query OpenCL C features for the device!\n");
192         return TEST_FAIL;
193     }
194 
195     const Version clc_version = get_device_cl_c_version(device);
196     if (clc_version >= Version(2, 0))
197     {
198         bool has_all_OpenCL_C_20_features =
199             features.supports__opencl_c_atomic_order_acq_rel
200             && features.supports__opencl_c_atomic_order_seq_cst
201             && features.supports__opencl_c_atomic_scope_device
202             && features.supports__opencl_c_atomic_scope_all_devices
203             && features.supports__opencl_c_device_enqueue
204             && features.supports__opencl_c_generic_address_space
205             && features.supports__opencl_c_pipes
206             && features.supports__opencl_c_program_scope_global_variables
207             && features.supports__opencl_c_work_group_collective_functions;
208 
209         if (features.supports__opencl_c_images)
210         {
211             has_all_OpenCL_C_20_features = has_all_OpenCL_C_20_features
212                 && features.supports__opencl_c_3d_image_writes
213                 && features.supports__opencl_c_read_write_images;
214         }
215 
216         test_assert_error(
217             has_all_OpenCL_C_20_features,
218             "At least one required OpenCL C 2.0 feature is missing!");
219     }
220 
221     return TEST_PASS;
222 }
223 
224 // This sub-test checks that all required OpenCL C versions are present for a
225 // specific CL_DEVICE_OPENCL_C_VERSION.
test_CL_DEVICE_OPENCL_C_VERSION_versions(cl_device_id device,cl_context context)226 static int test_CL_DEVICE_OPENCL_C_VERSION_versions(cl_device_id device,
227                                                     cl_context context)
228 {
229     log_info("  testing for OPENCL_C_VERSION required versions\n");
230 
231     const Version device_clc_version = get_device_cl_c_version(device);
232 
233     std::vector<Version> test_clc_versions;
234     test_clc_versions.push_back(Version(1, 0));
235     test_clc_versions.push_back(Version(1, 1));
236     test_clc_versions.push_back(Version(1, 2));
237     test_clc_versions.push_back(Version(2, 0));
238     test_clc_versions.push_back(Version(3, 0));
239 
240     cl_int error = CL_SUCCESS;
241 
242     size_t sz = 0;
243     error =
244         clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, 0, NULL, &sz);
245     test_error(error, "Unable to query CL_DEVICE_OPENCL_C_ALL_VERSIONS size");
246 
247     std::vector<cl_name_version> device_clc_versions(sz
248                                                      / sizeof(cl_name_version));
249     error = clGetDeviceInfo(device, CL_DEVICE_OPENCL_C_ALL_VERSIONS, sz,
250                             device_clc_versions.data(), NULL);
251     test_error(error, "Unable to query CL_DEVICE_OPENCL_C_FEATURES");
252 
253     for (const auto& test_clc_version : test_clc_versions)
254     {
255         if (device_clc_version >= test_clc_version)
256         {
257             bool found = false;
258             for (const auto& check : device_clc_versions)
259             {
260                 const unsigned major = CL_VERSION_MAJOR(check.version);
261                 const unsigned minor = CL_VERSION_MINOR(check.version);
262 
263                 if (strcmp(check.name, "OpenCL C") == 0
264                     && test_clc_version == Version(major, minor))
265                 {
266                     found = true;
267                     break;
268                 }
269             }
270 
271             if (found)
272             {
273                 log_info("    found OpenCL C version '%s'\n",
274                          test_clc_version.to_string().c_str());
275             }
276             else
277             {
278                 log_error("Didn't find OpenCL C version '%s'!\n",
279                           test_clc_version.to_string().c_str());
280                 return TEST_FAIL;
281             }
282         }
283     }
284 
285 
286     return TEST_PASS;
287 }
288 
test_opencl_c_versions(cl_device_id device,cl_context context,cl_command_queue queue,int num_elements)289 int test_opencl_c_versions(cl_device_id device, cl_context context,
290                            cl_command_queue queue, int num_elements)
291 {
292     check_compiler_available(device);
293 
294     const Version version = get_device_cl_version(device);
295 
296     int result = TEST_PASS;
297 
298     result |= test_CL_DEVICE_OPENCL_C_VERSION(device, context);
299 
300     if (version >= Version(3, 0))
301     {
302         result |= test_CL_DEVICE_OPENCL_C_ALL_VERSIONS(device, context);
303         result |= test_CL_DEVICE_OPENCL_C_VERSION_features(device, context);
304         result |= test_CL_DEVICE_OPENCL_C_VERSION_versions(device, context);
305     }
306 
307     return result;
308 }
309