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 #include "testBase.h"
17 #include <vector>
18 #include <algorithm>
19 #include "errorHelpers.h"
20
21 const char* macro_supported_source = R"(kernel void enabled(global int * buf) {
22 int n = get_global_id(0);
23 buf[n] = 0;
24 #ifndef %s
25 #error Feature macro was not defined
26 #endif
27 })";
28
29 const char* macro_not_supported_source =
30 R"(kernel void not_enabled(global int * buf) {
31 int n = get_global_id(0);
32 buf[n] = 0;
33 #ifdef %s
34 #error Feature macro was defined
35 #endif
36 })";
37
38 template <typename T>
check_api_feature_info_capabilities(cl_device_id deviceID,cl_context context,cl_bool & status,cl_device_info check_property,cl_bitfield check_cap)39 cl_int check_api_feature_info_capabilities(cl_device_id deviceID,
40 cl_context context, cl_bool& status,
41 cl_device_info check_property,
42 cl_bitfield check_cap)
43 {
44 cl_int error = CL_SUCCESS;
45 T response;
46 error = clGetDeviceInfo(deviceID, check_property, sizeof(response),
47 &response, NULL);
48 test_error(error, "clGetDeviceInfo failed.\n");
49
50 if ((response & check_cap) == check_cap)
51 {
52 status = CL_TRUE;
53 }
54 else
55 {
56 status = CL_FALSE;
57 }
58 return error;
59 }
60
check_api_feature_info_support(cl_device_id deviceID,cl_context context,cl_bool & status,cl_device_info check_property)61 cl_int check_api_feature_info_support(cl_device_id deviceID, cl_context context,
62 cl_bool& status,
63 cl_device_info check_property)
64 {
65 cl_int error = CL_SUCCESS;
66 cl_bool response;
67 error = clGetDeviceInfo(deviceID, check_property, sizeof(response),
68 &response, NULL);
69 test_error(error, "clGetDeviceInfo failed.\n");
70 status = response;
71 return error;
72 }
73
74 template <typename T>
check_api_feature_info_number(cl_device_id deviceID,cl_context context,cl_bool & status,cl_device_info check_property)75 cl_int check_api_feature_info_number(cl_device_id deviceID, cl_context context,
76 cl_bool& status,
77 cl_device_info check_property)
78 {
79 cl_int error = CL_SUCCESS;
80 T response;
81 error = clGetDeviceInfo(deviceID, check_property, sizeof(response),
82 &response, NULL);
83 test_error(error, "clGetDeviceInfo failed.\n");
84 if (response > 0)
85 {
86 status = CL_TRUE;
87 }
88 else
89 {
90 status = CL_FALSE;
91 }
92 return error;
93 }
94
check_api_feature_info_supported_image_formats(cl_device_id deviceID,cl_context context,cl_bool & status)95 cl_int check_api_feature_info_supported_image_formats(cl_device_id deviceID,
96 cl_context context,
97 cl_bool& status)
98 {
99 cl_int error = CL_SUCCESS;
100 cl_uint response = 0;
101 cl_uint image_format_count;
102 error = clGetSupportedImageFormats(context, CL_MEM_WRITE_ONLY,
103 CL_MEM_OBJECT_IMAGE3D, 0, NULL,
104 &image_format_count);
105 test_error(error, "clGetSupportedImageFormats failed");
106 response += image_format_count;
107 error = clGetSupportedImageFormats(context, CL_MEM_READ_WRITE,
108 CL_MEM_OBJECT_IMAGE3D, 0, NULL,
109 &image_format_count);
110 test_error(error, "clGetSupportedImageFormats failed");
111 response += image_format_count;
112 error = clGetSupportedImageFormats(context, CL_MEM_KERNEL_READ_AND_WRITE,
113 CL_MEM_OBJECT_IMAGE3D, 0, NULL,
114 &image_format_count);
115 test_error(error, "clGetSupportedImageFormats failed");
116 response += image_format_count;
117 if (response > 0)
118 {
119 status = CL_TRUE;
120 }
121 else
122 {
123 status = CL_FALSE;
124 }
125 return error;
126 }
127
check_compiler_feature_info(cl_device_id deviceID,cl_context context,std::string feature_macro,cl_bool & status)128 cl_int check_compiler_feature_info(cl_device_id deviceID, cl_context context,
129 std::string feature_macro, cl_bool& status)
130 {
131 cl_int error = CL_SUCCESS;
132 clProgramWrapper program_supported;
133 clProgramWrapper program_not_supported;
134 char kernel_supported_src[1024];
135 char kernel_not_supported_src[1024];
136 sprintf(kernel_supported_src, macro_supported_source,
137 feature_macro.c_str());
138 const char* ptr_supported = kernel_supported_src;
139 const char* build_options = "-cl-std=CL3.0";
140
141 error = create_single_kernel_helper_create_program(
142 context, &program_supported, 1, &ptr_supported, build_options);
143 test_error(error, "create_single_kernel_helper_create_program failed.\n");
144
145 sprintf(kernel_not_supported_src, macro_not_supported_source,
146 feature_macro.c_str());
147 const char* ptr_not_supported = kernel_not_supported_src;
148 error = create_single_kernel_helper_create_program(
149 context, &program_not_supported, 1, &ptr_not_supported,
150 "-cl-std=CL3.0");
151 test_error(error, "create_single_kernel_helper_create_program failed.\n");
152
153 cl_int status_supported = CL_SUCCESS;
154 cl_int status_not_supported = CL_SUCCESS;
155 status_supported = clBuildProgram(program_supported, 1, &deviceID,
156 build_options, NULL, NULL);
157 status_not_supported = clBuildProgram(program_not_supported, 1, &deviceID,
158 build_options, NULL, NULL);
159 if (status_supported != status_not_supported)
160 {
161 if (status_not_supported == CL_SUCCESS)
162 {
163 // kernel which verifies not supporting return passed
164 status = CL_FALSE;
165 }
166 else
167 {
168 // kernel which verifies supporting return passed
169 status = CL_TRUE;
170 }
171 }
172 else
173 {
174 log_error("Error: The macro feature is defined and undefined "
175 "in the same time\n");
176 error = OutputBuildLogs(program_supported, 1, &deviceID);
177 test_error(error, "OutputBuildLogs failed.\n");
178 error = OutputBuildLogs(program_not_supported, 1, &deviceID);
179 test_error(error, "OutputBuildLogs failed.\n");
180 return TEST_FAIL;
181 }
182 return error;
183 }
184
feature_macro_verify_results(std::string test_macro_name,cl_bool api_status,cl_bool compiler_status,cl_bool & supported)185 int feature_macro_verify_results(std::string test_macro_name,
186 cl_bool api_status, cl_bool compiler_status,
187 cl_bool& supported)
188 {
189 cl_int error = TEST_PASS;
190 log_info("Feature status: API - %s, compiler - %s\n",
191 api_status == CL_TRUE ? "supported" : "not supported",
192 compiler_status == CL_TRUE ? "supported" : "not supported");
193 if (api_status != compiler_status)
194 {
195 log_info("%s - failed\n", test_macro_name.c_str());
196 supported = CL_FALSE;
197 return TEST_FAIL;
198 }
199 else
200 {
201 log_info("%s - passed\n", test_macro_name.c_str());
202 }
203 supported = api_status;
204 return error;
205 }
206
test_feature_macro_atomic_order_acq_rel(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)207 int test_feature_macro_atomic_order_acq_rel(cl_device_id deviceID,
208 cl_context context,
209 std::string test_macro_name,
210 cl_bool& supported)
211 {
212 cl_int error = TEST_FAIL;
213 cl_bool api_status;
214 cl_bool compiler_status;
215 log_info("\n%s ...\n", test_macro_name.c_str());
216 error = check_api_feature_info_capabilities<cl_device_atomic_capabilities>(
217 deviceID, context, api_status, CL_DEVICE_ATOMIC_MEMORY_CAPABILITIES,
218 CL_DEVICE_ATOMIC_ORDER_ACQ_REL);
219 if (error != CL_SUCCESS)
220 {
221 return error;
222 }
223
224 error = check_compiler_feature_info(deviceID, context, test_macro_name,
225 compiler_status);
226 if (error != CL_SUCCESS)
227 {
228 return error;
229 }
230
231 return feature_macro_verify_results(test_macro_name, api_status,
232 compiler_status, supported);
233 }
234
test_feature_macro_atomic_order_seq_cst(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)235 int test_feature_macro_atomic_order_seq_cst(cl_device_id deviceID,
236 cl_context context,
237 std::string test_macro_name,
238 cl_bool& supported)
239 {
240 cl_int error = TEST_FAIL;
241 cl_bool api_status;
242 cl_bool compiler_status;
243 log_info("\n%s ...\n", test_macro_name.c_str());
244
245 error = check_api_feature_info_capabilities<cl_device_atomic_capabilities>(
246 deviceID, context, api_status, CL_DEVICE_ATOMIC_MEMORY_CAPABILITIES,
247 CL_DEVICE_ATOMIC_ORDER_SEQ_CST);
248 if (error != CL_SUCCESS)
249 {
250 return error;
251 }
252
253 error = check_compiler_feature_info(deviceID, context, test_macro_name,
254 compiler_status);
255 if (error != CL_SUCCESS)
256 {
257 return error;
258 }
259
260 return feature_macro_verify_results(test_macro_name, api_status,
261 compiler_status, supported);
262 }
263
test_feature_macro_atomic_scope_device(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)264 int test_feature_macro_atomic_scope_device(cl_device_id deviceID,
265 cl_context context,
266 std::string test_macro_name,
267 cl_bool& supported)
268 {
269 cl_int error = TEST_FAIL;
270 cl_bool api_status;
271 cl_bool compiler_status;
272 log_info("\n%s ...\n", test_macro_name.c_str());
273 error = check_api_feature_info_capabilities<cl_device_atomic_capabilities>(
274 deviceID, context, api_status, CL_DEVICE_ATOMIC_MEMORY_CAPABILITIES,
275 CL_DEVICE_ATOMIC_SCOPE_DEVICE);
276 if (error != CL_SUCCESS)
277 {
278 return error;
279 }
280 error = check_compiler_feature_info(deviceID, context, test_macro_name,
281 compiler_status);
282 if (error != CL_SUCCESS)
283 {
284 return error;
285 }
286
287 return feature_macro_verify_results(test_macro_name, api_status,
288 compiler_status, supported);
289 }
290
test_feature_macro_atomic_scope_all_devices(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)291 int test_feature_macro_atomic_scope_all_devices(cl_device_id deviceID,
292 cl_context context,
293 std::string test_macro_name,
294 cl_bool& supported)
295 {
296 cl_int error = TEST_FAIL;
297 cl_bool api_status;
298 cl_bool compiler_status;
299 log_info("\n%s ...\n", test_macro_name.c_str());
300 error = check_api_feature_info_capabilities<cl_device_atomic_capabilities>(
301 deviceID, context, api_status, CL_DEVICE_ATOMIC_MEMORY_CAPABILITIES,
302 CL_DEVICE_ATOMIC_SCOPE_ALL_DEVICES);
303 if (error != CL_SUCCESS)
304 {
305 return error;
306 }
307 error = check_compiler_feature_info(deviceID, context, test_macro_name,
308 compiler_status);
309 if (error != CL_SUCCESS)
310 {
311 return error;
312 }
313
314 return feature_macro_verify_results(test_macro_name, api_status,
315 compiler_status, supported);
316 }
317
test_feature_macro_3d_image_writes(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)318 int test_feature_macro_3d_image_writes(cl_device_id deviceID,
319 cl_context context,
320 std::string test_macro_name,
321 cl_bool& supported)
322 {
323 cl_int error = TEST_FAIL;
324 cl_bool api_status;
325 cl_bool compiler_status;
326 log_info("\n%s ...\n", test_macro_name.c_str());
327 error = check_api_feature_info_supported_image_formats(deviceID, context,
328 api_status);
329 if (error != CL_SUCCESS)
330 {
331 return error;
332 }
333
334 error = check_compiler_feature_info(deviceID, context, test_macro_name,
335 compiler_status);
336 if (error != CL_SUCCESS)
337 {
338 return error;
339 }
340
341 return feature_macro_verify_results(test_macro_name, api_status,
342 compiler_status, supported);
343 }
344
test_feature_macro_device_enqueue(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)345 int test_feature_macro_device_enqueue(cl_device_id deviceID, cl_context context,
346 std::string test_macro_name,
347 cl_bool& supported)
348 {
349 cl_int error = TEST_FAIL;
350 cl_bool api_status;
351 cl_bool compiler_status;
352 log_info("\n%s ...\n", test_macro_name.c_str());
353 error = check_api_feature_info_capabilities<
354 cl_device_device_enqueue_capabilities>(
355 deviceID, context, api_status, CL_DEVICE_DEVICE_ENQUEUE_CAPABILITIES,
356 CL_DEVICE_QUEUE_SUPPORTED);
357 if (error != CL_SUCCESS)
358 {
359 return error;
360 }
361
362 error = check_compiler_feature_info(deviceID, context, test_macro_name,
363 compiler_status);
364 if (error != CL_SUCCESS)
365 {
366 return error;
367 }
368
369 return feature_macro_verify_results(test_macro_name, api_status,
370 compiler_status, supported);
371 }
372
test_feature_macro_generic_address_space(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)373 int test_feature_macro_generic_address_space(cl_device_id deviceID,
374 cl_context context,
375 std::string test_macro_name,
376 cl_bool& supported)
377 {
378 cl_int error = TEST_FAIL;
379 cl_bool api_status;
380 cl_bool compiler_status;
381 log_info("\n%s ...\n", test_macro_name.c_str());
382 error = check_api_feature_info_support(
383 deviceID, context, api_status, CL_DEVICE_GENERIC_ADDRESS_SPACE_SUPPORT);
384 if (error != CL_SUCCESS)
385 {
386 return error;
387 }
388
389 error = check_compiler_feature_info(deviceID, context, test_macro_name,
390 compiler_status);
391 if (error != CL_SUCCESS)
392 {
393 return error;
394 }
395
396 return feature_macro_verify_results(test_macro_name, api_status,
397 compiler_status, supported);
398 }
399
test_feature_macro_pipes(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)400 int test_feature_macro_pipes(cl_device_id deviceID, cl_context context,
401 std::string test_macro_name, cl_bool& supported)
402 {
403 cl_int error = TEST_FAIL;
404 cl_bool api_status;
405 cl_bool compiler_status;
406 log_info("\n%s ...\n", test_macro_name.c_str());
407 error = check_api_feature_info_support(deviceID, context, api_status,
408 CL_DEVICE_PIPE_SUPPORT);
409 if (error != CL_SUCCESS)
410 {
411 return error;
412 }
413
414 error = check_compiler_feature_info(deviceID, context, test_macro_name,
415 compiler_status);
416 if (error != CL_SUCCESS)
417 {
418 return error;
419 }
420
421 return feature_macro_verify_results(test_macro_name, api_status,
422 compiler_status, supported);
423 }
424
test_feature_macro_program_scope_global_variables(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)425 int test_feature_macro_program_scope_global_variables(
426 cl_device_id deviceID, cl_context context, std::string test_macro_name,
427 cl_bool& supported)
428 {
429 cl_int error = TEST_FAIL;
430 cl_bool api_status;
431 cl_bool compiler_status;
432 log_info("\n%s ...\n", test_macro_name.c_str());
433 error = check_api_feature_info_number<size_t>(
434 deviceID, context, api_status, CL_DEVICE_MAX_GLOBAL_VARIABLE_SIZE);
435 if (error != CL_SUCCESS)
436 {
437 return error;
438 }
439
440 error = check_compiler_feature_info(deviceID, context, test_macro_name,
441 compiler_status);
442 if (error != CL_SUCCESS)
443 {
444 return error;
445 }
446
447 return feature_macro_verify_results(test_macro_name, api_status,
448 compiler_status, supported);
449 }
450
test_feature_macro_read_write_images(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)451 int test_feature_macro_read_write_images(cl_device_id deviceID,
452 cl_context context,
453 std::string test_macro_name,
454 cl_bool& supported)
455 {
456 cl_int error = TEST_FAIL;
457 cl_bool api_status;
458 cl_bool compiler_status;
459 log_info("\n%s ...\n", test_macro_name.c_str());
460 error = check_api_feature_info_number<cl_uint>(
461 deviceID, context, api_status, CL_DEVICE_MAX_READ_WRITE_IMAGE_ARGS);
462 if (error != CL_SUCCESS)
463 {
464 return error;
465 }
466
467 error = check_compiler_feature_info(deviceID, context, test_macro_name,
468 compiler_status);
469 if (error != CL_SUCCESS)
470 {
471 return error;
472 }
473
474 return feature_macro_verify_results(test_macro_name, api_status,
475 compiler_status, supported);
476 }
477
test_feature_macro_subgroups(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)478 int test_feature_macro_subgroups(cl_device_id deviceID, cl_context context,
479 std::string test_macro_name,
480 cl_bool& supported)
481 {
482 cl_int error = TEST_FAIL;
483 cl_bool api_status;
484 cl_bool compiler_status;
485 log_info("\n%s ...\n", test_macro_name.c_str());
486 error = check_api_feature_info_number<cl_uint>(
487 deviceID, context, api_status, CL_DEVICE_MAX_NUM_SUB_GROUPS);
488 if (error != CL_SUCCESS)
489 {
490 return error;
491 }
492
493 error = check_compiler_feature_info(deviceID, context, test_macro_name,
494 compiler_status);
495 if (error != CL_SUCCESS)
496 {
497 return error;
498 }
499
500 return feature_macro_verify_results(test_macro_name, api_status,
501 compiler_status, supported);
502 }
503
test_feature_macro_work_group_collective_functions(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)504 int test_feature_macro_work_group_collective_functions(
505 cl_device_id deviceID, cl_context context, std::string test_macro_name,
506 cl_bool& supported)
507 {
508 cl_int error = TEST_FAIL;
509 cl_bool api_status;
510 cl_bool compiler_status;
511 log_info("\n%s ...\n", test_macro_name.c_str());
512 error = check_api_feature_info_support(
513 deviceID, context, api_status,
514 CL_DEVICE_WORK_GROUP_COLLECTIVE_FUNCTIONS_SUPPORT);
515 if (error != CL_SUCCESS)
516 {
517 return error;
518 }
519
520 error = check_compiler_feature_info(deviceID, context, test_macro_name,
521 compiler_status);
522 if (error != CL_SUCCESS)
523 {
524 return error;
525 }
526
527 return feature_macro_verify_results(test_macro_name, api_status,
528 compiler_status, supported);
529 }
530
test_feature_macro_images(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)531 int test_feature_macro_images(cl_device_id deviceID, cl_context context,
532 std::string test_macro_name, cl_bool& supported)
533 {
534 cl_int error = TEST_FAIL;
535 cl_bool api_status;
536 cl_bool compiler_status;
537 log_info("\n%s ...\n", test_macro_name.c_str());
538 error = check_api_feature_info_support(deviceID, context, api_status,
539 CL_DEVICE_IMAGE_SUPPORT);
540 if (error != CL_SUCCESS)
541 {
542 return error;
543 }
544
545 error = check_compiler_feature_info(deviceID, context, test_macro_name,
546 compiler_status);
547 if (error != CL_SUCCESS)
548 {
549 return error;
550 }
551
552 return feature_macro_verify_results(test_macro_name, api_status,
553 compiler_status, supported);
554 }
555
test_feature_macro_fp64(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)556 int test_feature_macro_fp64(cl_device_id deviceID, cl_context context,
557 std::string test_macro_name, cl_bool& supported)
558 {
559 cl_int error = TEST_FAIL;
560 cl_bool api_status;
561 cl_bool compiler_status;
562 log_info("\n%s ...\n", test_macro_name.c_str());
563 error = check_api_feature_info_capabilities<cl_device_fp_config>(
564 deviceID, context, api_status, CL_DEVICE_DOUBLE_FP_CONFIG,
565 CL_FP_FMA | CL_FP_ROUND_TO_NEAREST | CL_FP_INF_NAN | CL_FP_DENORM);
566 if (error != CL_SUCCESS)
567 {
568 return error;
569 }
570
571 error = check_compiler_feature_info(deviceID, context, test_macro_name,
572 compiler_status);
573 if (error != CL_SUCCESS)
574 {
575 return error;
576 }
577
578 return feature_macro_verify_results(test_macro_name, api_status,
579 compiler_status, supported);
580 }
581
test_feature_macro_int64(cl_device_id deviceID,cl_context context,std::string test_macro_name,cl_bool & supported)582 int test_feature_macro_int64(cl_device_id deviceID, cl_context context,
583 std::string test_macro_name, cl_bool& supported)
584 {
585 cl_int error = TEST_FAIL;
586 cl_bool api_status;
587 cl_bool compiler_status;
588 cl_int full_profile = 0;
589 log_info("\n%s ...\n", test_macro_name.c_str());
590 size_t ret_len;
591 char profile[32] = { 0 };
592 error = clGetDeviceInfo(deviceID, CL_DEVICE_PROFILE, sizeof(profile),
593 profile, &ret_len);
594 test_error(error, "clGetDeviceInfo(CL_DEVICE_PROFILE) failed");
595 if (ret_len < sizeof(profile) && strcmp(profile, "FULL_PROFILE") == 0)
596 {
597 full_profile = 1;
598 }
599 else if (ret_len < sizeof(profile)
600 && strcmp(profile, "EMBEDDED_PROFILE") == 0)
601 {
602 full_profile = 0;
603 }
604 else
605 {
606 log_error("Unknown device profile: %s\n", profile);
607 return TEST_FAIL;
608 }
609
610 if (full_profile)
611 {
612 api_status = CL_TRUE;
613 }
614 else
615 {
616 if (is_extension_available(deviceID, "cles_khr_int64"))
617 {
618 api_status = CL_TRUE;
619 }
620 else
621 {
622 cl_bool double_supported = CL_FALSE;
623 error = check_api_feature_info_capabilities<cl_device_fp_config>(
624 deviceID, context, double_supported, CL_DEVICE_DOUBLE_FP_CONFIG,
625 CL_FP_FMA | CL_FP_ROUND_TO_NEAREST | CL_FP_INF_NAN
626 | CL_FP_DENORM);
627 test_error(error, "checking CL_DEVICE_DOUBLE_FP_CONFIG failed");
628 if (double_supported == CL_FALSE)
629 {
630 api_status = CL_FALSE;
631 }
632 else
633 {
634 log_error("FP double type is supported and cles_khr_int64 "
635 "extension not supported\n");
636 return TEST_FAIL;
637 }
638 }
639 }
640
641 error = check_compiler_feature_info(deviceID, context, test_macro_name,
642 compiler_status);
643 if (error != CL_SUCCESS)
644 {
645 return error;
646 }
647
648 return feature_macro_verify_results(test_macro_name, api_status,
649 compiler_status, supported);
650 }
651
test_consistency_c_features_list(cl_device_id deviceID,std::vector<std::string> vec_to_cmp)652 int test_consistency_c_features_list(cl_device_id deviceID,
653 std::vector<std::string> vec_to_cmp)
654 {
655 log_info("\nComparison list of features: CL_DEVICE_OPENCL_C_FEATURES vs "
656 "API/compiler queries.\n");
657 cl_int error;
658 size_t config_size;
659 std::vector<cl_name_version> vec_device_feature;
660 std::vector<std::string> vec_device_feature_names;
661 error = clGetDeviceInfo(deviceID, CL_DEVICE_OPENCL_C_FEATURES, 0, NULL,
662 &config_size);
663
664 test_error(
665 error,
666 "clGetDeviceInfo asking for CL_DEVICE_OPENCL_C_FEATURES failed.\n");
667 if (config_size == 0)
668 {
669 log_info("Empty list of CL_DEVICE_OPENCL_C_FEATURES returned by "
670 "clGetDeviceInfo on this device.\n");
671 }
672 else
673 {
674 int vec_elements = config_size / sizeof(cl_name_version);
675 vec_device_feature.resize(vec_elements);
676 error = clGetDeviceInfo(deviceID, CL_DEVICE_OPENCL_C_FEATURES,
677 config_size, vec_device_feature.data(), 0);
678 test_error(
679 error,
680 "clGetDeviceInfo asking for CL_DEVICE_OPENCL_C_FEATURES failed.\n");
681 }
682 for (auto each_f : vec_device_feature)
683 {
684 vec_device_feature_names.push_back(each_f.name);
685 }
686 sort(vec_to_cmp.begin(), vec_to_cmp.end());
687 sort(vec_device_feature_names.begin(), vec_device_feature_names.end());
688
689 if (vec_device_feature_names == vec_to_cmp)
690 {
691 log_info("Comparison list of features - passed\n");
692 }
693 else
694 {
695 log_info("Comparison list of features - failed\n");
696 error = TEST_FAIL;
697 }
698 log_info(
699 "Supported features based on CL_DEVICE_OPENCL_C_FEATURES API query:\n");
700 for (auto each_f : vec_device_feature_names)
701 {
702 log_info("%s\n", each_f.c_str());
703 }
704
705 log_info("\nSupported features based on queries to API/compiler :\n");
706 for (auto each_f : vec_to_cmp)
707 {
708 log_info("%s\n", each_f.c_str());
709 }
710
711 return error;
712 }
713
714 #define NEW_FEATURE_MACRO_TEST(feat) \
715 test_macro_name = "__opencl_c_" #feat; \
716 error |= test_feature_macro_##feat(deviceID, context, test_macro_name, \
717 supported); \
718 if (supported) supported_features_vec.push_back(test_macro_name);
719
720
test_features_macro(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)721 int test_features_macro(cl_device_id deviceID, cl_context context,
722 cl_command_queue queue, int num_elements)
723 {
724
725 // Note: Not checking that the feature array is empty for the compiler not
726 // available case because the specification says "For devices that do not
727 // support compilation from OpenCL C source, this query may return an empty
728 // array." It "may" return an empty array implies that an implementation
729 // also "may not".
730 check_compiler_available(deviceID);
731
732 int error = TEST_PASS;
733 cl_bool supported = CL_FALSE;
734 std::string test_macro_name = "";
735 std::vector<std::string> supported_features_vec;
736 NEW_FEATURE_MACRO_TEST(program_scope_global_variables);
737 NEW_FEATURE_MACRO_TEST(3d_image_writes);
738 NEW_FEATURE_MACRO_TEST(atomic_order_acq_rel);
739 NEW_FEATURE_MACRO_TEST(atomic_order_seq_cst);
740 NEW_FEATURE_MACRO_TEST(atomic_scope_device);
741 NEW_FEATURE_MACRO_TEST(atomic_scope_all_devices);
742 NEW_FEATURE_MACRO_TEST(device_enqueue);
743 NEW_FEATURE_MACRO_TEST(generic_address_space);
744 NEW_FEATURE_MACRO_TEST(pipes);
745 NEW_FEATURE_MACRO_TEST(read_write_images);
746 NEW_FEATURE_MACRO_TEST(subgroups);
747 NEW_FEATURE_MACRO_TEST(work_group_collective_functions);
748 NEW_FEATURE_MACRO_TEST(images);
749 NEW_FEATURE_MACRO_TEST(fp64);
750 NEW_FEATURE_MACRO_TEST(int64);
751
752 error |= test_consistency_c_features_list(deviceID, supported_features_vec);
753
754 return error;
755 }
756