1 //
2 // Copyright (c) 2017 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 "harness/testHarness.h"
18 #include "harness/parseParameters.h"
19
20 const char *sample_kernel_code_single_line[] = {
21 "__kernel void sample_test(__global float *src, __global int *dst)\n"
22 "{\n"
23 " int tid = get_global_id(0);\n"
24 "\n"
25 " dst[tid] = (int)src[tid];\n"
26 "\n"
27 "}\n" };
28
29 const char *sample_kernel_code_multi_line[] = {
30 "__kernel void sample_test(__global float *src, __global int *dst)",
31 "{",
32 " int tid = get_global_id(0);",
33 "",
34 " dst[tid] = (int)src[tid];",
35 "",
36 "}" };
37
38 const char *sample_kernel_code_two_line[] = {
39 "__kernel void sample_test(__global float *src, __global int *dst)\n"
40 "{\n"
41 " int tid = get_global_id(0);\n"
42 "\n"
43 " dst[tid] = (int)src[tid];\n"
44 "\n"
45 "}\n",
46 "__kernel void sample_test2(__global int *src, __global float *dst)\n"
47 "{\n"
48 " int tid = get_global_id(0);\n"
49 "\n"
50 " dst[tid] = (float)src[tid];\n"
51 "\n"
52 "}\n" };
53
54
55 const char *sample_kernel_code_bad_multi_line[] = {
56 "__kernel void sample_test(__global float *src, __global int *dst)",
57 "{",
58 " int tid = get_global_id(0);thisisanerror",
59 "",
60 " dst[tid] = (int)src[tid];",
61 "",
62 "}" };
63
64
test_load_program_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)65 int test_load_program_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
66 {
67 int error;
68 clProgramWrapper program;
69 size_t length;
70 char *buffer;
71
72 /* Preprocess: calc the length of the source file line */
73 size_t line_length = strlen(sample_kernel_code_single_line[0]);
74
75 /* New OpenCL API only has one entry point, so go ahead and just try it */
76 program = clCreateProgramWithSource(
77 context, 1, sample_kernel_code_single_line, &line_length, &error);
78 test_error( error, "Unable to create reference program" );
79
80 /* Now get the source and compare against our original */
81 error = clGetProgramInfo( program, CL_PROGRAM_SOURCE, 0, NULL, &length );
82 test_error( error, "Unable to get length of first program source" );
83
84 // Note: according to spec section 5.4.5, the length returned should include the null terminator
85 if (length != line_length + 1)
86 {
87 log_error("ERROR: Length of program (%ld) does not match reference "
88 "length (%ld)!\n",
89 length, line_length + 1);
90 return -1;
91 }
92
93 buffer = (char *)malloc( length );
94 error = clGetProgramInfo( program, CL_PROGRAM_SOURCE, length, buffer, NULL );
95 test_error( error, "Unable to get buffer of first program source" );
96
97 if( strcmp( (char *)buffer, sample_kernel_code_single_line[ 0 ] ) != 0 )
98 {
99 log_error( "ERROR: Program sources do not match!\n" );
100 return -1;
101 }
102
103 /* All done */
104 free( buffer );
105
106 return 0;
107 }
108
test_load_multistring_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)109 int test_load_multistring_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
110 {
111 int error;
112 clProgramWrapper program;
113
114 int i;
115
116 constexpr int num_lines = ARRAY_SIZE(sample_kernel_code_multi_line);
117
118 /* Preprocess: calc the length of each source file line */
119 size_t line_lengths[num_lines];
120 for (i = 0; i < num_lines; i++)
121 {
122 line_lengths[i] = strlen(sample_kernel_code_multi_line[i]);
123 }
124
125 /* Create another program using the macro function */
126 program = clCreateProgramWithSource(context, num_lines,
127 sample_kernel_code_multi_line,
128 line_lengths, &error);
129 if( program == NULL )
130 {
131 log_error( "ERROR: Unable to create reference program!\n" );
132 return -1;
133 }
134
135 /* Try compiling */
136 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
137 test_error( error, "Unable to build multi-line program source" );
138
139 /* Should probably check binary here to verify the same results... */
140
141 /* All done! */
142
143 return 0;
144 }
145
test_load_two_kernel_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)146 int test_load_two_kernel_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
147 {
148 int error;
149 cl_program program;
150
151 int i;
152
153 constexpr int num_lines = ARRAY_SIZE(sample_kernel_code_two_line);
154
155 /* Preprocess: calc the length of each source file line */
156 size_t line_lengths[num_lines];
157 for (i = 0; i < num_lines; i++)
158 {
159 line_lengths[i] = strlen(sample_kernel_code_two_line[i]);
160 }
161
162 /* Now create a program using the macro function */
163 program = clCreateProgramWithSource(
164 context, num_lines, sample_kernel_code_two_line, line_lengths, &error);
165 if( program == NULL )
166 {
167 log_error( "ERROR: Unable to create two-kernel program!\n" );
168 return -1;
169 }
170
171 /* Try compiling */
172 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
173 test_error( error, "Unable to build two-kernel program source" );
174
175 /* Should probably check binary here to verify the same results... */
176
177 /* All done! */
178 error = clReleaseProgram( program );
179 test_error( error, "Unable to release program object" );
180
181 return 0;
182 }
183
test_load_null_terminated_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)184 int test_load_null_terminated_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
185 {
186 int error;
187 cl_program program;
188
189
190 /* Now create a program using the macro function */
191 program = clCreateProgramWithSource( context, 1, sample_kernel_code_single_line, NULL, &error );
192 if( program == NULL )
193 {
194 log_error( "ERROR: Unable to create null-terminated program!" );
195 return -1;
196 }
197
198 /* Try compiling */
199 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
200 test_error( error, "Unable to build null-terminated program source" );
201
202 /* Should probably check binary here to verify the same results... */
203
204 /* All done! */
205 error = clReleaseProgram( program );
206 test_error( error, "Unable to release program object" );
207
208 return 0;
209 }
210
test_load_null_terminated_multi_line_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)211 int test_load_null_terminated_multi_line_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
212 {
213 int error;
214 cl_program program;
215
216
217 int num_lines = ARRAY_SIZE(sample_kernel_code_multi_line);
218
219 /* Now create a program using the macro function */
220 program = clCreateProgramWithSource(
221 context, num_lines, sample_kernel_code_multi_line, NULL, &error);
222 if( program == NULL )
223 {
224 log_error( "ERROR: Unable to create null-terminated program!" );
225 return -1;
226 }
227
228 /* Try compiling */
229 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
230 test_error( error, "Unable to build null-terminated program source" );
231
232 /* Should probably check binary here to verify the same results... */
233
234 /* All done! */
235 error = clReleaseProgram( program );
236 test_error( error, "Unable to release program object" );
237
238 return 0;
239 }
240
241
test_load_discreet_length_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)242 int test_load_discreet_length_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
243 {
244 int error;
245 cl_program program;
246
247 int i;
248
249 constexpr int num_lines = ARRAY_SIZE(sample_kernel_code_bad_multi_line);
250
251 /* Preprocess: calc the length of each source file line */
252 size_t line_lengths[num_lines];
253 for (i = 0; i < num_lines; i++)
254 {
255 line_lengths[i] = strlen(sample_kernel_code_bad_multi_line[i]);
256 }
257
258 /* Now force the length of the third line to skip the actual error */
259 static_assert(num_lines >= 3, "expected at least 3 lines in source");
260 line_lengths[2] -= strlen("thisisanerror");
261
262 /* Now create a program using the macro function */
263 program = clCreateProgramWithSource(context, num_lines,
264 sample_kernel_code_bad_multi_line,
265 line_lengths, &error);
266 if( program == NULL )
267 {
268 log_error( "ERROR: Unable to create null-terminated program!" );
269 return -1;
270 }
271
272 /* Try compiling */
273 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
274 test_error( error, "Unable to build null-terminated program source" );
275
276 /* Should probably check binary here to verify the same results... */
277
278 /* All done! */
279 error = clReleaseProgram( program );
280 test_error( error, "Unable to release program object" );
281
282 return 0;
283 }
284
test_load_null_terminated_partial_multi_line_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)285 int test_load_null_terminated_partial_multi_line_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
286 {
287 int error;
288 cl_program program;
289 int i;
290
291 constexpr int num_lines = ARRAY_SIZE(sample_kernel_code_multi_line);
292
293 /* Preprocess: calc the length of each source file line */
294 size_t line_lengths[num_lines];
295 for (i = 0; i < num_lines; i++)
296 {
297 if( i & 0x01 )
298 line_lengths[i] =
299 0; /* Should force for null-termination on this line only */
300 else
301 line_lengths[i] = strlen(sample_kernel_code_multi_line[i]);
302 }
303
304 /* Now create a program using the macro function */
305 program = clCreateProgramWithSource(context, num_lines,
306 sample_kernel_code_multi_line,
307 line_lengths, &error);
308 if( program == NULL )
309 {
310 log_error( "ERROR: Unable to create null-terminated program!" );
311 return -1;
312 }
313
314 /* Try compiling */
315 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
316 test_error( error, "Unable to build null-terminated program source" );
317
318 /* Should probably check binary here to verify the same results... */
319
320 /* All done! */
321 error = clReleaseProgram( program );
322 test_error( error, "Unable to release program object" );
323
324 return 0;
325 }
326
test_get_program_info(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)327 int test_get_program_info(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
328 {
329 int error;
330 cl_program program;
331 cl_device_id device1;
332 cl_context context1;
333 size_t paramSize;
334 cl_uint numInstances;
335
336
337 error = create_single_kernel_helper_create_program(context, &program, 1, sample_kernel_code_single_line);
338 if( program == NULL )
339 {
340 log_error( "ERROR: Unable to create reference program!\n" );
341 return -1;
342 }
343
344 /* Test that getting the device works. */
345 device1 = (cl_device_id)0xbaadfeed;
346 error = clGetProgramInfo( program, CL_PROGRAM_DEVICES, sizeof( device1 ), &device1, NULL );
347 test_error( error, "Unable to get device of program" );
348
349 /* Since the device IDs are opaque types we check the CL_DEVICE_VENDOR_ID which is unique for identical hardware. */
350 cl_uint device1_vid, deviceID_vid;
351 error = clGetDeviceInfo(device1, CL_DEVICE_VENDOR_ID, sizeof(device1_vid), &device1_vid, NULL );
352 test_error( error, "Unable to get device CL_DEVICE_VENDOR_ID" );
353 error = clGetDeviceInfo(deviceID, CL_DEVICE_VENDOR_ID, sizeof(deviceID_vid), &deviceID_vid, NULL );
354 test_error( error, "Unable to get device CL_DEVICE_VENDOR_ID" );
355
356 if( device1_vid != deviceID_vid )
357 {
358 log_error( "ERROR: Incorrect device returned for program! (Expected vendor ID 0x%x, got 0x%x)\n", deviceID_vid, device1_vid );
359 return -1;
360 }
361
362 cl_uint devCount;
363 error = clGetProgramInfo( program, CL_PROGRAM_NUM_DEVICES, sizeof( devCount ), &devCount, NULL );
364 test_error( error, "Unable to get device count of program" );
365
366 if( devCount != 1 )
367 {
368 log_error( "ERROR: Invalid device count returned for program! (Expected 1, got %d)\n", (int)devCount );
369 return -1;
370 }
371
372 context1 = (cl_context)0xbaadfeed;
373 error = clGetProgramInfo( program, CL_PROGRAM_CONTEXT, sizeof( context1 ), &context1, NULL );
374 test_error( error, "Unable to get device of program" );
375
376 if( context1 != context )
377 {
378 log_error( "ERROR: Invalid context returned for program! (Expected %p, got %p)\n", context, context1 );
379 return -1;
380 }
381
382 error = clGetProgramInfo( program, CL_PROGRAM_REFERENCE_COUNT, sizeof( numInstances ), &numInstances, NULL );
383 test_error( error, "Unable to get instance count" );
384
385 /* While we're at it, test the sizes of programInfo too */
386 error = clGetProgramInfo( program, CL_PROGRAM_DEVICES, 0, NULL, ¶mSize );
387 test_error( error, "Unable to get device param size" );
388 if( paramSize != sizeof( cl_device_id ) )
389 {
390 log_error( "ERROR: Size returned for device is wrong!\n" );
391 return -1;
392 }
393
394 error = clGetProgramInfo( program, CL_PROGRAM_CONTEXT, 0, NULL, ¶mSize );
395 test_error( error, "Unable to get context param size" );
396 if( paramSize != sizeof( cl_context ) )
397 {
398 log_error( "ERROR: Size returned for context is wrong!\n" );
399 return -1;
400 }
401
402 error = clGetProgramInfo( program, CL_PROGRAM_REFERENCE_COUNT, 0, NULL, ¶mSize );
403 test_error( error, "Unable to get instance param size" );
404 if( paramSize != sizeof( cl_uint ) )
405 {
406 log_error( "ERROR: Size returned for num instances is wrong!\n" );
407 return -1;
408 }
409
410 error = clGetProgramInfo( program, CL_PROGRAM_NUM_DEVICES, 0, NULL, ¶mSize );
411 test_error( error, "Unable to get device count param size" );
412 if( paramSize != sizeof( cl_uint ) )
413 {
414 log_error( "ERROR: Size returned for device count is wrong!\n" );
415 return -1;
416 }
417
418 /* All done! */
419 error = clReleaseProgram( program );
420 test_error( error, "Unable to release program object" );
421
422 return 0;
423 }
424
test_get_program_source(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)425 int test_get_program_source(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
426 {
427 cl_program program;
428 int error;
429 char buffer[10240];
430 size_t length;
431 size_t line_length = strlen(sample_kernel_code_single_line[0]);
432 bool online_compilation = (gCompilationMode == kOnline);
433
434 error = create_single_kernel_helper_create_program(context, &program, 1, sample_kernel_code_single_line);
435 if( program == NULL )
436 {
437 log_error( "ERROR: Unable to create test program!\n" );
438 return -1;
439 }
440
441 /* Try getting the length */
442 error = clGetProgramInfo( program, CL_PROGRAM_SOURCE, 0, NULL, &length );
443 test_error( error, "Unable to get program source length" );
444 if (length != line_length + 1 && online_compilation)
445 {
446 log_error( "ERROR: Length returned for program source is incorrect!\n" );
447 return -1;
448 }
449
450 /* Try normal source */
451 error = clGetProgramInfo( program, CL_PROGRAM_SOURCE, sizeof( buffer ), buffer, NULL );
452 test_error( error, "Unable to get program source" );
453 if (strlen(buffer) != line_length && online_compilation)
454 {
455 log_error( "ERROR: Length of program source is incorrect!\n" );
456 return -1;
457 }
458
459 /* Try both at once */
460 error = clGetProgramInfo( program, CL_PROGRAM_SOURCE, sizeof( buffer ), buffer, &length );
461 test_error( error, "Unable to get program source" );
462 if (strlen(buffer) != line_length && online_compilation)
463 {
464 log_error( "ERROR: Length of program source is incorrect!\n" );
465 return -1;
466 }
467 if (length != line_length + 1 && online_compilation)
468 {
469 log_error( "ERROR: Returned length of program source is incorrect!\n" );
470 return -1;
471 }
472
473 /* All done! */
474 error = clReleaseProgram( program );
475 test_error( error, "Unable to release program object" );
476
477 return 0;
478 }
479
test_get_program_build_info(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)480 int test_get_program_build_info(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
481 {
482 cl_program program;
483 int error;
484 char *buffer;
485 size_t length, newLength;
486 cl_build_status status;
487
488
489 error = create_single_kernel_helper_create_program(context, &program, 1, sample_kernel_code_single_line);
490 if( program == NULL )
491 {
492 log_error( "ERROR: Unable to create test program!\n" );
493 return -1;
494 }
495
496 /* Make sure getting the length works */
497 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, 0, NULL, &length );
498 test_error( error, "Unable to get program build status length" );
499 if( length != sizeof( status ) )
500 {
501 log_error( "ERROR: Returned length of program build status is invalid! (Expected %d, got %d)\n", (int)sizeof( status ), (int)length );
502 return -1;
503 }
504
505 /* Now actually build it and verify the status */
506 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
507 test_error( error, "Unable to build program source" );
508
509 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
510 test_error( error, "Unable to get program build status" );
511 if( status != CL_BUILD_SUCCESS )
512 {
513 log_error( "ERROR: Getting built program build status did not return CL_BUILD_SUCCESS! (%d)\n", (int)status );
514 return -1;
515 }
516
517 /***** Build log *****/
518
519 /* Try getting the length */
520 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &length );
521 test_error( error, "Unable to get program build log length" );
522
523 log_info("Build log is %ld long.\n", length);
524
525 buffer = (char*)malloc(length);
526
527 /* Try normal source */
528 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, length, buffer, NULL );
529 test_error( error, "Unable to get program build log" );
530
531 if( buffer[length-1] != '\0' )
532 {
533 log_error( "clGetProgramBuildInfo overwrote allocated space for build log! '%c'\n", buffer[length-1] );
534 return -1;
535 }
536
537 /* Try both at once */
538 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, length, buffer, &newLength );
539 test_error( error, "Unable to get program build log" );
540
541 free(buffer);
542
543 /***** Build options *****/
544 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &length );
545 test_error( error, "Unable to get program build options length" );
546
547 buffer = (char*)malloc(length);
548
549 /* Try normal source */
550 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, length, buffer, NULL );
551 test_error( error, "Unable to get program build options" );
552
553 /* Try both at once */
554 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, length, buffer, &newLength );
555 test_error( error, "Unable to get program build options" );
556
557 free(buffer);
558
559 /* Try with a valid option */
560 error = clReleaseProgram( program );
561 test_error( error, "Unable to release program object" );
562
563 program = clCreateProgramWithSource( context, 1, sample_kernel_code_single_line, NULL, &error );
564 if( program == NULL )
565 {
566 log_error( "ERROR: Unable to create test program!\n" );
567 return -1;
568 }
569
570 error = clBuildProgram( program, 1, &deviceID, "-cl-opt-disable", NULL, NULL );
571 if( error != CL_SUCCESS )
572 {
573 print_error( error, "Building with valid options failed!" );
574 return -1;
575 }
576
577 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &length );
578 test_error( error, "Unable to get program build options" );
579
580 buffer = (char*)malloc(length);
581
582 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, length, buffer, NULL );
583 test_error( error, "Unable to get program build options" );
584 if( strcmp( (char *)buffer, "-cl-opt-disable" ) != 0 )
585 {
586 log_error( "ERROR: Getting program build options for program with -cl-opt-disable build options did not return expected value (got %s)\n", buffer );
587 return -1;
588 }
589
590 /* All done */
591 free( buffer );
592
593 error = clReleaseProgram( program );
594 test_error( error, "Unable to release program object" );
595
596 return 0;
597 }
598