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 #if defined(_WIN32)
18 #include <time.h>
19 #elif defined(__linux__) || defined(__APPLE__)
20 #include <sys/time.h>
21 #include <unistd.h>
22 #endif
23 #include "harness/conversions.h"
24
25 #define MAX_LINE_SIZE_IN_PROGRAM 1024
26 #define MAX_LOG_SIZE_IN_PROGRAM 2048
27
28 const char *sample_kernel_start =
29 "__kernel void sample_test(__global float *src, __global int *dst)\n"
30 "{\n"
31 " float temp;\n"
32 " int tid = get_global_id(0);\n";
33
34 const char *sample_kernel_end = "}\n";
35
36 const char *sample_kernel_lines[] = {
37 "dst[tid] = src[tid];\n",
38 "dst[tid] = src[tid] * 3.f;\n",
39 "temp = src[tid] / 4.f;\n",
40 "dst[tid] = dot(temp,src[tid]);\n",
41 "dst[tid] = dst[tid] + temp;\n" };
42
43 /* I compile and link therefore I am. Robert Ioffe */
44 /* The following kernels are used in testing Improved Compilation and Linking feature */
45
46 const char *simple_kernel =
47 "__kernel void\n"
48 "CopyBuffer(\n"
49 " __global float* src,\n"
50 " __global float* dst )\n"
51 "{\n"
52 " int id = (int)get_global_id(0);\n"
53 " dst[id] = src[id];\n"
54 "}\n";
55
56 const char *simple_kernel_with_defines =
57 "__kernel void\n"
58 "CopyBuffer(\n"
59 " __global float* src,\n"
60 " __global float* dst )\n"
61 "{\n"
62 " int id = (int)get_global_id(0);\n"
63 " float temp = src[id] - 42;\n"
64 " dst[id] = FIRST + temp + SECOND;\n"
65 "}\n";
66
67 const char *simple_kernel_template =
68 "__kernel void\n"
69 "CopyBuffer%d(\n"
70 " __global float* src,\n"
71 " __global float* dst )\n"
72 "{\n"
73 " int id = (int)get_global_id(0);\n"
74 " dst[id] = src[id];\n"
75 "}\n";
76
77 const char *composite_kernel_start =
78 "__kernel void\n"
79 "CompositeKernel(\n"
80 " __global float* src,\n"
81 " __global float* dst )\n"
82 "{\n";
83
84 const char *composite_kernel_end = "}\n";
85
86 const char *composite_kernel_template =
87 " CopyBuffer%d(src, dst);\n";
88
89 const char *composite_kernel_extern_template =
90 "extern __kernel void\n"
91 "CopyBuffer%d(\n"
92 " __global float* src,\n"
93 " __global float* dst );\n";
94
95 const char *another_simple_kernel =
96 "extern __kernel void\n"
97 "CopyBuffer(\n"
98 " __global float* src,\n"
99 " __global float* dst );\n"
100 "__kernel void\n"
101 "AnotherCopyBuffer(\n"
102 " __global float* src,\n"
103 " __global float* dst )\n"
104 "{\n"
105 " CopyBuffer(src, dst);\n"
106 "}\n";
107
108 const char* simple_header =
109 "extern __kernel void\n"
110 "CopyBuffer(\n"
111 " __global float* src,\n"
112 " __global float* dst );\n";
113
114 const char* simple_header_name = "simple_header.h";
115
116 const char* another_simple_kernel_with_header =
117 "#include \"simple_header.h\"\n"
118 "__kernel void\n"
119 "AnotherCopyBuffer(\n"
120 " __global float* src,\n"
121 " __global float* dst )\n"
122 "{\n"
123 " CopyBuffer(src, dst);\n"
124 "}\n";
125
126 const char* header_name_templates[4] = { "simple_header%d.h",
127 "foo/simple_header%d.h",
128 "foo/bar/simple_header%d.h",
129 "foo/bar/baz/simple_header%d.h"};
130
131 const char* include_header_name_templates[4] = { "#include \"simple_header%d.h\"\n",
132 "#include \"foo/simple_header%d.h\"\n",
133 "#include \"foo/bar/simple_header%d.h\"\n",
134 "#include \"foo/bar/baz/simple_header%d.h\"\n"};
135
136 const char* compile_extern_var = "extern constant float foo;\n";
137 const char* compile_extern_struct = "extern constant struct bar bart;\n";
138 const char* compile_extern_function = "extern int baz(int, int);\n";
139
140 const char* compile_static_var = "static constant float foo = 2.78;\n";
141 const char* compile_static_struct = "static constant struct bar {float x, y, z, r; int color; } foo = {3.14159};\n";
142 const char* compile_static_function = "static int foo(int x, int y) { return x*x + y*y; }\n";
143
144 const char* compile_regular_var = "constant float foo = 4.0f;\n";
145 const char* compile_regular_struct = "constant struct bar {float x, y, z, r; int color; } foo = {0.f, 0.f, 0.f, 0.f, 0};\n";
146 const char* compile_regular_function = "int foo(int x, int y) { return x*x + y*y; }\n";
147
148 const char* link_static_var_access = // use with compile_static_var
149 "extern constant float foo;\n"
150 "float access_foo() { return foo; }\n";
151
152 const char* link_static_struct_access = // use with compile_static_struct
153 "extern constant struct bar{float x, y, z, r; int color; } foo;\n"
154 "struct bar access_foo() {return foo; }\n";
155
156 const char* link_static_function_access = // use with compile_static_function
157 "extern int foo(int, int);\n"
158 "int access_foo() { int blah = foo(3, 4); return blah + 5; }\n";
159
test_large_single_compile(cl_context context,cl_device_id deviceID,unsigned int numLines)160 int test_large_single_compile(cl_context context, cl_device_id deviceID, unsigned int numLines)
161 {
162 int error;
163 cl_program program;
164 const char **lines;
165 unsigned int numChoices, i;
166 MTdata d;
167
168 /* First, allocate the array for our line pointers */
169 lines = (const char **)malloc( numLines * sizeof( const char * ) );
170 if (lines == NULL) {
171 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", numLines, __FILE__, __LINE__);
172 return -1;
173 }
174
175 /* First and last lines are easy */
176 lines[ 0 ] = sample_kernel_start;
177 lines[ numLines - 1 ] = sample_kernel_end;
178
179 numChoices = sizeof( sample_kernel_lines ) / sizeof( sample_kernel_lines[ 0 ] );
180
181 /* Fill the rest with random lines to hopefully prevent much optimization */
182 d = init_genrand( gRandomSeed );
183 for( i = 1; i < numLines - 1; i++ )
184 {
185 lines[ i ] = sample_kernel_lines[ genrand_int32(d) % numChoices ];
186 }
187 free_mtdata(d); d = NULL;
188
189 /* Try to create a program with these lines */
190 error = create_single_kernel_helper_create_program(context, &program, numLines, lines);
191 if( program == NULL || error != CL_SUCCESS )
192 {
193 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
194 free( lines );
195 if (program != NULL)
196 {
197 error = clReleaseProgram( program );
198 test_error( error, "Unable to release a program object" );
199 }
200 return -1;
201 }
202
203 /* Build it */
204 error = clBuildProgram( program, 1, &deviceID, NULL, NULL, NULL );
205 test_error( error, "Unable to build a long program" );
206
207 /* All done! */
208 error = clReleaseProgram( program );
209 test_error( error, "Unable to release a program object" );
210
211 free( lines );
212
213 return 0;
214 }
215
test_large_compile(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)216 int test_large_compile(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
217 {
218 unsigned int toTest[] = { 64, 128, 256, 512, 1024, 2048, 4096, 0 }; //8192, 16384, 32768, 0 };
219 unsigned int i;
220
221 log_info( "Testing large compiles...this might take awhile...\n" );
222
223 for( i = 0; toTest[ i ] != 0; i++ )
224 {
225 log_info( " %d...\n", toTest[ i ] );
226
227 #if defined(_WIN32)
228 clock_t start = clock();
229 #elif defined(__linux__) || defined(__APPLE__)
230 timeval time1, time2;
231 gettimeofday(&time1, NULL);
232 #endif
233
234 if( test_large_single_compile( context, deviceID, toTest[ i ] ) != 0 )
235 {
236 log_error( "ERROR: long program test failed for %d lines! (in %s:%d)\n", toTest[ i ], __FILE__, __LINE__);
237 return -1;
238 }
239
240 #if defined(_WIN32)
241 clock_t end = clock();
242 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
243 #elif defined(__linux__) || defined(__APPLE__)
244 gettimeofday(&time2, NULL);
245 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
246 #endif
247 }
248
249 return 0;
250 }
251
252 static int verifyCopyBuffer(cl_context context, cl_command_queue queue, cl_kernel kernel);
253
254 #if defined(__APPLE__) || defined(__linux)
255 #define _strdup strdup
256 #endif
257
test_large_multi_file_library(cl_context context,cl_device_id deviceID,cl_command_queue queue,unsigned int numLines)258 int test_large_multi_file_library(cl_context context, cl_device_id deviceID, cl_command_queue queue, unsigned int numLines)
259 {
260 int error;
261 cl_program program;
262 cl_program *simple_kernels;
263 const char **lines;
264 unsigned int i;
265 char buffer[MAX_LINE_SIZE_IN_PROGRAM];
266
267 simple_kernels = (cl_program*)malloc(numLines*sizeof(cl_program));
268 if (simple_kernels == NULL) {
269 log_error( "ERROR: Unable to allocate kernels array with %d kernels! (in %s:%d)\n", numLines, __FILE__, __LINE__);
270 return -1;
271 }
272 /* First, allocate the array for our line pointers */
273 lines = (const char **)malloc( (2*numLines + 2) * sizeof( const char * ) );
274 if (lines == NULL) {
275 free(simple_kernels);
276 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", (2*numLines + 2), __FILE__, __LINE__ );
277 return -1;
278 }
279
280 for( i = 0; i < numLines; i++)
281 {
282 sprintf(buffer, composite_kernel_extern_template, i);
283 lines[i] = _strdup(buffer);
284 }
285 /* First and last lines are easy */
286 lines[ numLines ] = composite_kernel_start;
287 lines[ 2* numLines + 1] = composite_kernel_end;
288
289 /* Fill the rest with templated kernels */
290 for( i = numLines + 1; i < 2* numLines + 1; i++ )
291 {
292 sprintf(buffer, composite_kernel_template, i - numLines - 1);
293 lines[ i ] = _strdup(buffer);
294 }
295
296 /* Try to create a program with these lines */
297 error = create_single_kernel_helper_create_program(context, &program, 2 * numLines + 2, lines);
298 if( program == NULL || error != CL_SUCCESS )
299 {
300 log_error( "ERROR: Unable to create long test program with %d lines! (%s) (in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
301 free( simple_kernels );
302 for( i = 0; i < numLines; i++)
303 {
304 free( (void*)lines[i] );
305 free( (void*)lines[i+numLines+1] );
306 }
307 free( lines );
308 if (program != NULL)
309 {
310 error = clReleaseProgram( program );
311 test_error( error, "Unable to release program object" );
312 }
313
314 return -1;
315 }
316
317 /* Compile it */
318 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
319 test_error( error, "Unable to compile a simple program" );
320
321 /* Create and compile templated kernels */
322 for( i = 0; i < numLines; i++)
323 {
324 sprintf(buffer, simple_kernel_template, i);
325 const char* kernel_source = _strdup(buffer);
326 simple_kernels[i] = clCreateProgramWithSource( context, 1, &kernel_source, NULL, &error );
327 if( simple_kernels[i] == NULL || error != CL_SUCCESS )
328 {
329 log_error( "ERROR: Unable to create long test program with %d lines! (%s) (in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
330 return -1;
331 }
332
333 /* Compile it */
334 error = clCompileProgram(simple_kernels[i], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
335 test_error( error, "Unable to compile a simple program" );
336
337 free((void*)kernel_source);
338 }
339
340 /* Create library out of compiled templated kernels */
341 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", numLines, simple_kernels, NULL, NULL, &error);
342 test_error( error, "Unable to create a multi-line library" );
343
344 /* Link the program that calls the kernels and the library that contains them */
345 cl_program programs[2] = { program, my_newly_minted_library };
346 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, programs, NULL, NULL, &error);
347 test_error( error, "Unable to link a program with a library" );
348
349 // Create the composite kernel
350 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CompositeKernel", &error);
351 test_error( error, "Unable to create a composite kernel" );
352
353 // Run the composite kernel and verify the results
354 error = verifyCopyBuffer(context, queue, kernel);
355 if (error != CL_SUCCESS)
356 return error;
357
358 /* All done! */
359 error = clReleaseProgram( program );
360 test_error( error, "Unable to release program object" );
361
362 for( i = 0; i < numLines; i++)
363 {
364 free( (void*)lines[i] );
365 free( (void*)lines[i+numLines+1] );
366 }
367 free( lines );
368
369 for(i = 0; i < numLines; i++)
370 {
371 error = clReleaseProgram( simple_kernels[i] );
372 test_error( error, "Unable to release program object" );
373 }
374 free( simple_kernels );
375
376 error = clReleaseKernel( kernel );
377 test_error( error, "Unable to release kernel object" );
378
379 error = clReleaseProgram( my_newly_minted_library );
380 test_error( error, "Unable to release program object" );
381
382 error = clReleaseProgram( my_newly_linked_program );
383 test_error( error, "Unable to release program object" );
384
385 return 0;
386 }
387
test_multi_file_libraries(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)388 int test_multi_file_libraries(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
389 {
390 unsigned int toTest[] = { 2, 4, 8, 16, 32, 64, 128, 256, 0 }; // 512, 1024, 2048, 4096, 8192, 16384, 32768, 0 };
391 unsigned int i;
392
393 log_info( "Testing multi-file libraries ...this might take awhile...\n" );
394
395 for( i = 0; toTest[ i ] != 0; i++ )
396 {
397 log_info( " %d...\n", toTest[ i ] );
398
399 #if defined(_WIN32)
400 clock_t start = clock();
401 #elif defined(__linux__) || defined(__APPLE__)
402 timeval time1, time2;
403 gettimeofday(&time1, NULL);
404 #endif
405
406 if( test_large_multi_file_library( context, deviceID, queue, toTest[ i ] ) != 0 )
407 {
408 log_error( "ERROR: multi-file library program test failed for %d lines! (in %s:%d)\n\n", toTest[ i ], __FILE__, __LINE__ );
409 return -1;
410 }
411
412 #if defined(_WIN32)
413 clock_t end = clock();
414 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
415 #elif defined(__linux__) || defined(__APPLE__)
416 gettimeofday(&time2, NULL);
417 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
418 #endif
419 }
420
421 return 0;
422 }
423
test_large_multiple_embedded_headers(cl_context context,cl_device_id deviceID,cl_command_queue queue,unsigned int numLines)424 int test_large_multiple_embedded_headers(cl_context context, cl_device_id deviceID, cl_command_queue queue, unsigned int numLines)
425 {
426 int error;
427 cl_program program;
428 cl_program *simple_kernels;
429 cl_program *headers;
430 const char **header_names;
431 const char **lines;
432 unsigned int i;
433 char buffer[MAX_LINE_SIZE_IN_PROGRAM];
434
435 simple_kernels = (cl_program*)malloc(numLines*sizeof(cl_program));
436 if (simple_kernels == NULL) {
437 log_error( "ERROR: Unable to allocate simple_kernels array with %d lines! (in %s:%d)\n", numLines, __FILE__, __LINE__ );
438 return -1;
439 }
440 headers = (cl_program*)malloc(numLines*sizeof(cl_program));
441 if (headers == NULL) {
442 log_error( "ERROR: Unable to allocate headers array with %d lines! (in %s:%d)\n", numLines, __FILE__, __LINE__ );
443 return -1;
444 }
445 /* First, allocate the array for our line pointers */
446 header_names = (const char**)malloc( numLines*sizeof( const char * ) );
447 if (header_names == NULL) {
448 log_error( "ERROR: Unable to allocate header_names array with %d lines! (in %s:%d)\n", numLines, __FILE__, __LINE__ );
449 return -1;
450 }
451 lines = (const char **)malloc( (2*numLines + 2)*sizeof( const char * ) );
452 if (lines == NULL) {
453 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", (2*numLines + 2), __FILE__, __LINE__ );
454 return -1;
455 }
456
457 for( i = 0; i < numLines; i++)
458 {
459 sprintf(buffer, include_header_name_templates[i % 4], i);
460 lines[i] = _strdup(buffer);
461 sprintf(buffer, header_name_templates[i % 4], i);
462 header_names[i] = _strdup(buffer);
463
464 sprintf(buffer, composite_kernel_extern_template, i);
465 const char* line = _strdup(buffer);
466 error = create_single_kernel_helper_create_program(context, &headers[i], 1, &line);
467 if( headers[i] == NULL || error != CL_SUCCESS )
468 {
469 log_error( "ERROR: Unable to create a simple header program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__);
470 return -1;
471 }
472 }
473 /* First and last lines are easy */
474 lines[ numLines ] = composite_kernel_start;
475 lines[ 2* numLines + 1 ] = composite_kernel_end;
476
477 /* Fill the rest with templated kernels */
478 for( i = numLines + 1; i < 2* numLines + 1; i++ )
479 {
480 sprintf(buffer, composite_kernel_template, i - numLines - 1);
481 lines[ i ] = _strdup(buffer);
482 }
483
484 /* Try to create a program with these lines */
485 error = create_single_kernel_helper_create_program(context, &program, 2 * numLines + 2, lines);
486 if( program == NULL || error != CL_SUCCESS )
487 {
488 log_error( "ERROR: Unable to create long test program with %d lines! (%s) (in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
489 return -1;
490 }
491
492 /* Compile it */
493 error = clCompileProgram(program, 1, &deviceID, NULL, numLines, headers, header_names, NULL, NULL);
494 test_error( error, "Unable to compile a simple program" );
495
496 /* Create and compile templated kernels */
497 for( i = 0; i < numLines; i++)
498 {
499 sprintf(buffer, simple_kernel_template, i);
500 const char* kernel_source = _strdup(buffer);
501 error = create_single_kernel_helper_create_program(context, &simple_kernels[i], 1, &kernel_source);
502 if( simple_kernels[i] == NULL || error != CL_SUCCESS )
503 {
504 log_error( "ERROR: Unable to create long test program with %d lines! (%s) (in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
505 return -1;
506 }
507
508 /* Compile it */
509 error = clCompileProgram(simple_kernels[i], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
510 test_error( error, "Unable to compile a simple program" );
511
512 free((void*)kernel_source);
513 }
514
515 /* Create library out of compiled templated kernels */
516 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", numLines, simple_kernels, NULL, NULL, &error);
517 test_error( error, "Unable to create a multi-line library" );
518
519 /* Link the program that calls the kernels and the library that contains them */
520 cl_program programs[2] = { program, my_newly_minted_library };
521 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, programs, NULL, NULL, &error);
522 test_error( error, "Unable to link a program with a library" );
523
524 // Create the composite kernel
525 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CompositeKernel", &error);
526 test_error( error, "Unable to create a composite kernel" );
527
528 // Run the composite kernel and verify the results
529 error = verifyCopyBuffer(context, queue, kernel);
530 if (error != CL_SUCCESS)
531 return error;
532
533 /* All done! */
534 error = clReleaseProgram( program );
535 test_error( error, "Unable to release program object" );
536
537 for( i = 0; i < numLines; i++)
538 {
539 free( (void*)lines[i] );
540 free( (void*)header_names[i] );
541 }
542 for( i = numLines + 1; i < 2* numLines + 1; i++ )
543 {
544 free( (void*)lines[i] );
545 }
546 free( lines );
547 free( header_names );
548
549 for(i = 0; i < numLines; i++)
550 {
551 error = clReleaseProgram( simple_kernels[i] );
552 test_error( error, "Unable to release program object" );
553 error = clReleaseProgram( headers[i] );
554 test_error( error, "Unable to release header program object" );
555 }
556 free( simple_kernels );
557 free( headers );
558
559 error = clReleaseKernel( kernel );
560 test_error( error, "Unable to release kernel object" );
561
562 error = clReleaseProgram( my_newly_minted_library );
563 test_error( error, "Unable to release program object" );
564
565 error = clReleaseProgram( my_newly_linked_program );
566 test_error( error, "Unable to release program object" );
567
568 return 0;
569 }
570
test_multiple_embedded_headers(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)571 int test_multiple_embedded_headers(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
572 {
573 unsigned int toTest[] = { 2, 4, 8, 16, 32, 64, 128, 256, 0 }; // 512, 1024, 2048, 4096, 8192, 16384, 32768, 0 };
574 unsigned int i;
575
576 log_info( "Testing multiple embedded headers ...this might take awhile...\n" );
577
578 for( i = 0; toTest[ i ] != 0; i++ )
579 {
580 log_info( " %d...\n", toTest[ i ] );
581
582 #if defined(_WIN32)
583 clock_t start = clock();
584 #elif defined(__linux__) || defined(__APPLE__)
585 timeval time1, time2;
586 gettimeofday(&time1, NULL);
587 #endif
588
589 if( test_large_multiple_embedded_headers( context, deviceID, queue, toTest[ i ] ) != 0 )
590 {
591 log_error( "ERROR: multiple embedded headers program test failed for %d lines! (in %s:%d)\n", toTest[ i ], __FILE__, __LINE__ );
592 return -1;
593 }
594
595 #if defined(_WIN32)
596 clock_t end = clock();
597 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
598 #elif defined(__linux__) || defined(__APPLE__)
599 gettimeofday(&time2, NULL);
600 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
601 #endif
602 }
603
604 return 0;
605 }
606
logbase(double a,double base)607 double logbase(double a, double base)
608 {
609 return log(a) / log(base);
610 }
611
test_large_multiple_libraries(cl_context context,cl_device_id deviceID,cl_command_queue queue,unsigned int numLines)612 int test_large_multiple_libraries(cl_context context, cl_device_id deviceID, cl_command_queue queue, unsigned int numLines)
613 {
614 int error;
615 cl_program *simple_kernels;
616 const char **lines;
617 unsigned int i;
618 char buffer[MAX_LINE_SIZE_IN_PROGRAM];
619 /* I want to create (log2(N)+1)/2 libraries */
620 unsigned int level = (unsigned int)(logbase(numLines, 2.0) + 1.000001)/2;
621 unsigned int numLibraries = (unsigned int)pow(2.0, level - 1.0);
622 unsigned int numFilesInLib = numLines/numLibraries;
623 cl_program *my_program_and_libraries = (cl_program*)malloc((1+numLibraries)*sizeof(cl_program));
624 if (my_program_and_libraries == NULL) {
625 log_error( "ERROR: Unable to allocate program array with %d programs! (in %s:%d)\n", (1+numLibraries), __FILE__, __LINE__);
626 return -1;
627 }
628
629 log_info("level - %d, numLibraries - %d, numFilesInLib - %d\n", level, numLibraries, numFilesInLib);
630
631 simple_kernels = (cl_program*)malloc(numLines*sizeof(cl_program));
632 if (simple_kernels == NULL) {
633 log_error( "ERROR: Unable to allocate kernels array with %d kernels! (in %s:%d)\n", numLines, __FILE__, __LINE__);
634 return -1;
635 }
636 /* First, allocate the array for our line pointers */
637 lines = (const char **)malloc( (2*numLines + 2) * sizeof( const char * ) );
638 if (lines == NULL) {
639 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", (2*numLines + 2), __FILE__, __LINE__);
640 return -1;
641 }
642
643 for(i = 0; i < numLines; i++)
644 {
645 sprintf(buffer, composite_kernel_extern_template, i);
646 lines[i] = _strdup(buffer);
647 }
648 /* First and last lines are easy */
649 lines[ numLines ] = composite_kernel_start;
650 lines[ 2*numLines + 1] = composite_kernel_end;
651
652 /* Fill the rest with templated kernels */
653 for(i = numLines + 1; i < 2*numLines + 1; i++ )
654 {
655 sprintf(buffer, composite_kernel_template, i - numLines - 1);
656 lines[ i ] = _strdup(buffer);
657 }
658
659 /* Try to create a program with these lines */
660 error = create_single_kernel_helper_create_program(context, &my_program_and_libraries[0], 2 * numLines + 2, lines);
661 if( my_program_and_libraries[0] == NULL || error != CL_SUCCESS )
662 {
663 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
664 return -1;
665 }
666
667 /* Compile it */
668 error = clCompileProgram(my_program_and_libraries[0], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
669 test_error( error, "Unable to compile a simple program" );
670
671 /* Create and compile templated kernels */
672 for(i = 0; i < numLines; i++)
673 {
674 sprintf(buffer, simple_kernel_template, i);
675 const char* kernel_source = _strdup(buffer);
676 error = create_single_kernel_helper_create_program(context, &simple_kernels[i], 1, &kernel_source);
677 if( simple_kernels[i] == NULL || error != CL_SUCCESS )
678 {
679 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
680 return -1;
681 }
682
683 /* Compile it */
684 error = clCompileProgram(simple_kernels[i], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
685 test_error( error, "Unable to compile a simple program" );
686
687 free((void*)kernel_source);
688 }
689
690 /* Create library out of compiled templated kernels */
691 for(i = 0; i < numLibraries; i++) {
692 my_program_and_libraries[i+1] = clLinkProgram(context, 1, &deviceID, "-create-library", numFilesInLib, simple_kernels+i*numFilesInLib, NULL, NULL, &error);
693 test_error( error, "Unable to create a multi-line library" );
694 }
695
696 /* Link the program that calls the kernels and the library that contains them */
697 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, numLibraries+1, my_program_and_libraries, NULL, NULL, &error);
698 test_error( error, "Unable to link a program with a library" );
699
700 // Create the composite kernel
701 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CompositeKernel", &error);
702 test_error( error, "Unable to create a composite kernel" );
703
704 // Run the composite kernel and verify the results
705 error = verifyCopyBuffer(context, queue, kernel);
706 if (error != CL_SUCCESS)
707 return error;
708
709 /* All done! */
710 for(i = 0; i <= numLibraries; i++) {
711 error = clReleaseProgram( my_program_and_libraries[i] );
712 test_error( error, "Unable to release program object" );
713 }
714 free( my_program_and_libraries );
715 for(i = 0; i < numLines; i++)
716 {
717 free( (void*)lines[i] );
718 }
719 for(i = numLines + 1; i < 2*numLines + 1; i++ )
720 {
721 free( (void*)lines[i] );
722 }
723 free( lines );
724
725 for(i = 0; i < numLines; i++)
726 {
727 error = clReleaseProgram( simple_kernels[i] );
728 test_error( error, "Unable to release program object" );
729 }
730 free( simple_kernels );
731
732 error = clReleaseKernel( kernel );
733 test_error( error, "Unable to release kernel object" );
734
735 error = clReleaseProgram( my_newly_linked_program );
736 test_error( error, "Unable to release program object" );
737
738 return 0;
739 }
740
test_multiple_libraries(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)741 int test_multiple_libraries(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
742 {
743 unsigned int toTest[] = { 2, 8, 32, 128, 256, 0 }; // 512, 2048, 8192, 32768, 0 };
744 unsigned int i;
745
746 log_info( "Testing multiple libraries ...this might take awhile...\n" );
747
748 for( i = 0; toTest[ i ] != 0; i++ )
749 {
750 log_info( " %d...\n", toTest[ i ] );
751
752 #if defined(_WIN32)
753 clock_t start = clock();
754 #elif defined(__linux__) || defined(__APPLE__)
755 timeval time1, time2;
756 gettimeofday(&time1, NULL);
757 #endif
758
759 if( test_large_multiple_libraries( context, deviceID, queue, toTest[ i ] ) != 0 )
760 {
761 log_error( "ERROR: multiple library program test failed for %d lines! (in %s:%d)\n\n", toTest[ i ], __FILE__, __LINE__ );
762 return -1;
763 }
764
765 #if defined(_WIN32)
766 clock_t end = clock();
767 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
768 #elif defined(__linux__) || defined(__APPLE__)
769 gettimeofday(&time2, NULL);
770 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
771 #endif
772 }
773
774 return 0;
775 }
776
test_large_multiple_files_multiple_libraries(cl_context context,cl_device_id deviceID,cl_command_queue queue,unsigned int numLines)777 int test_large_multiple_files_multiple_libraries(cl_context context, cl_device_id deviceID, cl_command_queue queue, unsigned int numLines)
778 {
779 int error;
780 cl_program *simple_kernels;
781 const char **lines;
782 unsigned int i;
783 char buffer[MAX_LINE_SIZE_IN_PROGRAM];
784 /* I want to create (log2(N)+1)/4 libraries */
785 unsigned int level = (unsigned int)(logbase(numLines, 2.0) + 1.000001)/2;
786 unsigned int numLibraries = (unsigned int)pow(2.0, level - 2.0);
787 unsigned int numFilesInLib = numLines/(2*numLibraries);
788 cl_program *my_programs_and_libraries = (cl_program*)malloc((1+numLibraries+numLibraries*numFilesInLib)*sizeof(cl_program));
789 if (my_programs_and_libraries == NULL) {
790 log_error( "ERROR: Unable to allocate program array with %d programs! (in %s:%d)\n", (1+numLibraries+numLibraries*numFilesInLib), __FILE__, __LINE__ );
791 return -1;
792 }
793 log_info("level - %d, numLibraries - %d, numFilesInLib - %d\n", level, numLibraries, numFilesInLib);
794
795 simple_kernels = (cl_program*)malloc(numLines*sizeof(cl_program));
796 if (simple_kernels == NULL) {
797 log_error( "ERROR: Unable to allocate kernels array with %d kernels! (in %s:%d)\n", numLines, __FILE__, __LINE__ );
798 return -1;
799 }
800 /* First, allocate the array for our line pointers */
801 lines = (const char **)malloc( (2*numLines + 2) * sizeof( const char * ) );
802 if (lines == NULL) {
803 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", (2*numLines + 2), __FILE__, __LINE__ );
804 return -1;
805 }
806
807 for(i = 0; i < numLines; i++)
808 {
809 sprintf(buffer, composite_kernel_extern_template, i);
810 lines[i] = _strdup(buffer);
811 }
812 /* First and last lines are easy */
813 lines[ numLines ] = composite_kernel_start;
814 lines[ 2*numLines + 1] = composite_kernel_end;
815
816 /* Fill the rest with templated kernels */
817 for(i = numLines + 1; i < 2*numLines + 1; i++ )
818 {
819 sprintf(buffer, composite_kernel_template, i - numLines - 1);
820 lines[ i ] = _strdup(buffer);
821 }
822
823 /* Try to create a program with these lines */
824 error = create_single_kernel_helper_create_program(context, &my_programs_and_libraries[0], 2 * numLines + 2, lines);
825 if( my_programs_and_libraries[0] == NULL || error != CL_SUCCESS )
826 {
827 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
828 return -1;
829 }
830
831 /* Compile it */
832 error = clCompileProgram(my_programs_and_libraries[0], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
833 test_error( error, "Unable to compile a simple program" );
834
835 /* Create and compile templated kernels */
836 for(i = 0; i < numLines; i++)
837 {
838 sprintf(buffer, simple_kernel_template, i);
839 const char* kernel_source = _strdup(buffer);
840 error = create_single_kernel_helper_create_program(context, &simple_kernels[i], 1, &kernel_source);
841 if( simple_kernels[i] == NULL || error != CL_SUCCESS )
842 {
843 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
844 return -1;
845 }
846
847 /* Compile it */
848 error = clCompileProgram(simple_kernels[i], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
849 test_error( error, "Unable to compile a simple program" );
850
851 free((void*)kernel_source);
852 }
853
854 /* Copy already compiled kernels */
855 for( i = 0; i < numLibraries*numFilesInLib; i++) {
856 my_programs_and_libraries[i+1] = simple_kernels[i];
857 }
858
859 /* Create library out of compiled templated kernels */
860 for( i = 0; i < numLibraries; i++) {
861 my_programs_and_libraries[i+1+numLibraries*numFilesInLib] = clLinkProgram(context, 1, &deviceID, "-create-library", numFilesInLib, simple_kernels+(i*numFilesInLib+numLibraries*numFilesInLib), NULL, NULL, &error);
862 test_error( error, "Unable to create a multi-line library" );
863 }
864
865 /* Link the program that calls the kernels and the library that contains them */
866 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, numLibraries+1+numLibraries*numFilesInLib, my_programs_and_libraries, NULL, NULL, &error);
867 test_error( error, "Unable to link a program with a library" );
868
869 // Create the composite kernel
870 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CompositeKernel", &error);
871 test_error( error, "Unable to create a composite kernel" );
872
873 // Run the composite kernel and verify the results
874 error = verifyCopyBuffer(context, queue, kernel);
875 if (error != CL_SUCCESS)
876 return error;
877
878 /* All done! */
879 for(i = 0; i < numLibraries+1+numLibraries*numFilesInLib; i++) {
880 error = clReleaseProgram( my_programs_and_libraries[i] );
881 test_error( error, "Unable to release program object" );
882 }
883 free( my_programs_and_libraries );
884
885 for(i = 0; i < numLines; i++)
886 {
887 free( (void*)lines[i] );
888 }
889 for(i = numLines + 1; i < 2*numLines + 1; i++ )
890 {
891 free( (void*)lines[i] );
892 }
893 free( lines );
894
895 for(i = numLibraries*numFilesInLib; i < numLines; i++)
896 {
897 error = clReleaseProgram( simple_kernels[i] );
898 test_error( error, "Unable to release program object" );
899 }
900 free( simple_kernels );
901
902 error = clReleaseKernel( kernel );
903 test_error( error, "Unable to release kernel object" );
904
905 error = clReleaseProgram( my_newly_linked_program );
906 test_error( error, "Unable to release program object" );
907
908 return 0;
909 }
910
test_multiple_files_multiple_libraries(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)911 int test_multiple_files_multiple_libraries(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
912 {
913 unsigned int toTest[] = { 8, 32, 128, 256, 0 }; // 512, 2048, 8192, 32768, 0 };
914 unsigned int i;
915
916 log_info( "Testing multiple files and multiple libraries ...this might take awhile...\n" );
917
918 for( i = 0; toTest[ i ] != 0; i++ )
919 {
920 log_info( " %d...\n", toTest[ i ] );
921
922 #if defined(_WIN32)
923 clock_t start = clock();
924 #elif defined(__linux__) || defined(__APPLE__)
925 timeval time1, time2;
926 gettimeofday(&time1, NULL);
927 #endif
928
929 if( test_large_multiple_files_multiple_libraries( context, deviceID, queue, toTest[ i ] ) != 0 )
930 {
931 log_error( "ERROR: multiple files, multiple libraries program test failed for %d lines! (in %s:%d)\n\n", toTest[ i ], __FILE__, __LINE__ );
932 return -1;
933 }
934
935 #if defined(_WIN32)
936 clock_t end = clock();
937 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
938 #elif defined(__linux__) || defined(__APPLE__)
939 gettimeofday(&time2, NULL);
940 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
941 #endif
942 }
943
944 return 0;
945 }
946
test_large_multiple_files(cl_context context,cl_device_id deviceID,cl_command_queue queue,unsigned int numLines)947 int test_large_multiple_files(cl_context context, cl_device_id deviceID, cl_command_queue queue, unsigned int numLines)
948 {
949 int error;
950 const char **lines;
951 unsigned int i;
952 char buffer[MAX_LINE_SIZE_IN_PROGRAM];
953 cl_program *my_programs = (cl_program*)malloc((1+numLines)*sizeof(cl_program));
954
955 if (my_programs == NULL) {
956 log_error( "ERROR: Unable to allocate my_programs array with %d programs! (in %s:%d)\n", (1+numLines), __FILE__, __LINE__);
957 return -1;
958 }
959 /* First, allocate the array for our line pointers */
960 lines = (const char **)malloc( (2*numLines + 2) * sizeof( const char * ) );
961 if (lines == NULL) {
962 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", (2*numLines + 2), __FILE__, __LINE__);
963 return -1;
964 }
965
966 for(i = 0; i < numLines; i++)
967 {
968 sprintf(buffer, composite_kernel_extern_template, i);
969 lines[i] = _strdup(buffer);
970 }
971 /* First and last lines are easy */
972 lines[ numLines ] = composite_kernel_start;
973 lines[ 2* numLines + 1] = composite_kernel_end;
974
975 /* Fill the rest with templated kernels */
976 for(i = numLines + 1; i < 2*numLines + 1; i++ )
977 {
978 sprintf(buffer, composite_kernel_template, i - numLines - 1);
979 lines[ i ] = _strdup(buffer);
980 }
981
982 /* Try to create a program with these lines */
983 error = create_single_kernel_helper_create_program(context, &my_programs[0], 2 * numLines + 2, lines);
984 if( my_programs[0] == NULL || error != CL_SUCCESS )
985 {
986 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
987 return -1;
988 }
989
990 /* Compile it */
991 error = clCompileProgram(my_programs[0], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
992 test_error( error, "Unable to compile a simple program" );
993
994 /* Create and compile templated kernels */
995 for( i = 0; i < numLines; i++)
996 {
997 sprintf(buffer, simple_kernel_template, i);
998 const char* kernel_source = _strdup(buffer);
999 error = create_single_kernel_helper_create_program(context, &my_programs[i + 1], 1, &kernel_source);
1000 if( my_programs[i+1] == NULL || error != CL_SUCCESS )
1001 {
1002 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
1003 return -1;
1004 }
1005
1006 /* Compile it */
1007 error = clCompileProgram(my_programs[i+1], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1008 test_error( error, "Unable to compile a simple program" );
1009
1010 free((void*)kernel_source);
1011 }
1012
1013 /* Link the program that calls the kernels and the library that contains them */
1014 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 1+numLines, my_programs, NULL, NULL, &error);
1015 test_error( error, "Unable to link a program with a library" );
1016
1017 // Create the composite kernel
1018 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CompositeKernel", &error);
1019 test_error( error, "Unable to create a composite kernel" );
1020
1021 // Run the composite kernel and verify the results
1022 error = verifyCopyBuffer(context, queue, kernel);
1023 if (error != CL_SUCCESS)
1024 return error;
1025
1026 /* All done! */
1027 for(i = 0; i < 1+numLines; i++) {
1028 error = clReleaseProgram( my_programs[i] );
1029 test_error( error, "Unable to release program object" );
1030 }
1031 free( my_programs );
1032 for(i = 0; i < numLines; i++)
1033 {
1034 free( (void*)lines[i] );
1035 }
1036 for(i = numLines + 1; i < 2*numLines + 1; i++ )
1037 {
1038 free( (void*)lines[i] );
1039 }
1040 free( lines );
1041
1042 error = clReleaseKernel( kernel );
1043 test_error( error, "Unable to release kernel object" );
1044
1045 error = clReleaseProgram( my_newly_linked_program );
1046 test_error( error, "Unable to release program object" );
1047
1048 return 0;
1049 }
1050
test_multiple_files(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1051 int test_multiple_files(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1052 {
1053 unsigned int toTest[] = { 8, 32, 128, 256, 0 }; // 512, 2048, 8192, 32768, 0 };
1054 unsigned int i;
1055
1056 log_info( "Testing multiple files compilation and linking into a single executable ...this might take awhile...\n" );
1057
1058 for( i = 0; toTest[ i ] != 0; i++ )
1059 {
1060 log_info( " %d...\n", toTest[ i ] );
1061
1062 #if defined(_WIN32)
1063 clock_t start = clock();
1064 #elif defined(__linux__) || defined(__APPLE__)
1065 timeval time1, time2;
1066 gettimeofday(&time1, NULL);
1067 #endif
1068
1069 if( test_large_multiple_files( context, deviceID, queue, toTest[ i ] ) != 0 )
1070 {
1071 log_error( "ERROR: multiple files program test failed for %d lines! (in %s:%d)\n\n", toTest[ i ], __FILE__, __LINE__ );
1072 return -1;
1073 }
1074
1075 #if defined(_WIN32)
1076 clock_t end = clock();
1077 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
1078 #elif defined(__linux__) || defined(__APPLE__)
1079 gettimeofday(&time2, NULL);
1080 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
1081 #endif
1082 }
1083
1084 return 0;
1085 }
1086
test_simple_compile_only(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1087 int test_simple_compile_only(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1088 {
1089 int error;
1090 cl_program program;
1091
1092 log_info("Testing a simple compilation only...\n");
1093 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1094 if( program == NULL || error != CL_SUCCESS )
1095 {
1096 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1097 return -1;
1098 }
1099
1100 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1101 test_error( error, "Unable to compile a simple program" );
1102
1103 /* All done! */
1104 error = clReleaseProgram( program );
1105 test_error( error, "Unable to release program object" );
1106
1107 return 0;
1108 }
1109
test_simple_static_compile_only(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1110 int test_simple_static_compile_only(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1111 {
1112 int error;
1113 cl_program program;
1114
1115 log_info("Testing a simple static compilations only...\n");
1116
1117 error = create_single_kernel_helper_create_program(context, &program, 1, &compile_static_var);
1118 if( program == NULL || error != CL_SUCCESS )
1119 {
1120 log_error( "ERROR: Unable to create a simple static variable test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1121 return -1;
1122 }
1123
1124 log_info("Compiling a static variable...\n");
1125 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1126 test_error( error, "Unable to compile a simple static variable program" );
1127
1128 /* All done! */
1129 error = clReleaseProgram( program );
1130 test_error( error, "Unable to release program object" );
1131
1132 error = create_single_kernel_helper_create_program(context, &program, 1, &compile_static_struct);
1133 if( program == NULL || error != CL_SUCCESS )
1134 {
1135 log_error( "ERROR: Unable to create a simple static struct test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1136 return -1;
1137 }
1138
1139 log_info("Compiling a static struct...\n");
1140 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1141 test_error( error, "Unable to compile a simple static variable program" );
1142
1143 /* All done! */
1144 error = clReleaseProgram( program );
1145 test_error( error, "Unable to release program object" );
1146
1147 error = create_single_kernel_helper_create_program(context, &program, 1, &compile_static_function);
1148 if( program == NULL || error != CL_SUCCESS )
1149 {
1150 log_error( "ERROR: Unable to create a simple static function test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1151 return -1;
1152 }
1153
1154 log_info("Compiling a static function...\n");
1155 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1156 test_error( error, "Unable to compile a simple static function program" );
1157
1158 /* All done! */
1159 error = clReleaseProgram( program );
1160 test_error( error, "Unable to release program object" );
1161
1162 return 0;
1163 }
1164
test_simple_extern_compile_only(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1165 int test_simple_extern_compile_only(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1166 {
1167 int error;
1168 cl_program program;
1169
1170 log_info("Testing a simple extern compilations only...\n");
1171 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_header);
1172 if( program == NULL || error != CL_SUCCESS )
1173 {
1174 log_error( "ERROR: Unable to create a simple extern kernel test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1175 return -1;
1176 }
1177
1178 log_info("Compiling an extern kernel...\n");
1179 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1180 test_error( error, "Unable to compile a simple extern kernel program" );
1181
1182 /* All done! */
1183 error = clReleaseProgram( program );
1184 test_error( error, "Unable to release program object" );
1185
1186 error = create_single_kernel_helper_create_program(context, &program, 1, &compile_extern_var);
1187 if( program == NULL || error != CL_SUCCESS )
1188 {
1189 log_error( "ERROR: Unable to create a simple extern variable test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1190 return -1;
1191 }
1192
1193 log_info("Compiling an extern variable...\n");
1194 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1195 test_error( error, "Unable to compile a simple extern variable program" );
1196
1197 /* All done! */
1198 error = clReleaseProgram( program );
1199 test_error( error, "Unable to release program object" );
1200
1201 error = create_single_kernel_helper_create_program(context, &program, 1, &compile_extern_struct);
1202 if( program == NULL || error != CL_SUCCESS )
1203 {
1204 log_error( "ERROR: Unable to create a simple extern struct test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1205 return -1;
1206 }
1207
1208 log_info("Compiling an extern struct...\n");
1209 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1210 test_error( error, "Unable to compile a simple extern variable program" );
1211
1212 /* All done! */
1213 error = clReleaseProgram( program );
1214 test_error( error, "Unable to release program object" );
1215
1216 error = create_single_kernel_helper_create_program(context, &program, 1, &compile_extern_function);
1217 if( program == NULL || error != CL_SUCCESS )
1218 {
1219 log_error( "ERROR: Unable to create a simple extern function test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1220 return -1;
1221 }
1222
1223 log_info("Compiling an extern function...\n");
1224 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1225 test_error( error, "Unable to compile a simple extern function program" );
1226
1227 /* All done! */
1228 error = clReleaseProgram( program );
1229 test_error( error, "Unable to release program object" );
1230
1231 return 0;
1232 }
1233
1234 struct simple_user_data {
1235 const char* m_message;
1236 cl_event m_event;
1237 };
1238
1239 const char* once_upon_a_midnight_dreary = "Once upon a midnight dreary!";
1240
simple_compile_callback(cl_program program,void * user_data)1241 static void CL_CALLBACK simple_compile_callback(cl_program program, void* user_data)
1242 {
1243 simple_user_data* simple_compile_user_data = (simple_user_data*)user_data;
1244 log_info("in the simple_compile_callback: program %p just completed compiling with '%s'\n", program, simple_compile_user_data->m_message);
1245 if (strcmp(once_upon_a_midnight_dreary, simple_compile_user_data->m_message) != 0)
1246 {
1247 log_error("ERROR: in the simple_compile_callback: Expected '%s' and got %s (in %s:%d)!\n", once_upon_a_midnight_dreary, simple_compile_user_data->m_message, __FILE__, __LINE__);
1248 }
1249
1250 int error;
1251 log_info("in the simple_compile_callback: program %p just completed compiling with '%p'\n", program, simple_compile_user_data->m_event);
1252
1253 error = clSetUserEventStatus(simple_compile_user_data->m_event, CL_COMPLETE);
1254 if (error != CL_SUCCESS)
1255 {
1256 log_error( "ERROR: in the simple_compile_callback: Unable to set user event status to CL_COMPLETE! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1257 exit(-1);
1258 }
1259 log_info("in the simple_compile_callback: Successfully signaled compile_program_completion_event!\n");
1260 }
1261
test_simple_compile_with_callback(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1262 int test_simple_compile_with_callback(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1263 {
1264 int error;
1265 cl_program program;
1266 cl_event compile_program_completion_event;
1267
1268 log_info("Testing a simple compilation with callback...\n");
1269 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1270 if( program == NULL || error != CL_SUCCESS )
1271 {
1272 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1273 return -1;
1274 }
1275
1276 compile_program_completion_event = clCreateUserEvent(context, &error);
1277 test_error( error, "Unable to create a user event");
1278
1279 simple_user_data simple_compile_user_data = {once_upon_a_midnight_dreary, compile_program_completion_event};
1280
1281 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, simple_compile_callback, (void*)&simple_compile_user_data);
1282 test_error( error, "Unable to compile a simple program with a callback" );
1283
1284 error = clWaitForEvents(1, &compile_program_completion_event);
1285 test_error( error, "clWaitForEvents failed when waiting on compile_program_completion_event");
1286
1287 /* All done! */
1288 error = clReleaseEvent(compile_program_completion_event);
1289 test_error( error, "Unable to release event object" );
1290
1291 error = clReleaseProgram( program );
1292 test_error( error, "Unable to release program object" );
1293
1294 return 0;
1295 }
1296
test_simple_embedded_header_compile(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1297 int test_simple_embedded_header_compile(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1298 {
1299 int error;
1300 cl_program program, header;
1301
1302 log_info("Testing a simple embedded header compile only...\n");
1303 program = clCreateProgramWithSource(context, 1, &another_simple_kernel_with_header, NULL, &error);
1304 if( program == NULL || error != CL_SUCCESS )
1305 {
1306 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1307 return -1;
1308 }
1309
1310 header = clCreateProgramWithSource(context, 1, &simple_header, NULL, &error);
1311 if( header == NULL || error != CL_SUCCESS )
1312 {
1313 log_error( "ERROR: Unable to create a simple header program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1314 return -1;
1315 }
1316
1317 error = clCompileProgram(program, 1, &deviceID, NULL, 1, &header, &simple_header_name, NULL, NULL);
1318 test_error( error, "Unable to compile a simple program with embedded header" );
1319
1320 /* All done! */
1321 error = clReleaseProgram( program );
1322 test_error( error, "Unable to release program object" );
1323
1324 error = clReleaseProgram( header );
1325 test_error( error, "Unable to release program object" );
1326
1327 return 0;
1328 }
1329
test_simple_link_only(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1330 int test_simple_link_only(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1331 {
1332 int error;
1333 cl_program program;
1334
1335 log_info("Testing a simple linking only...\n");
1336 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1337 if( program == NULL || error != CL_SUCCESS )
1338 {
1339 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1340 return -1;
1341 }
1342
1343 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1344 test_error( error, "Unable to compile a simple program" );
1345
1346 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 1, &program, NULL, NULL, &error);
1347 test_error( error, "Unable to link a simple program" );
1348
1349 /* All done! */
1350 error = clReleaseProgram( program );
1351 test_error( error, "Unable to release program object" );
1352
1353 error = clReleaseProgram( my_newly_linked_program );
1354 test_error( error, "Unable to release program object" );
1355
1356 return 0;
1357 }
1358
test_two_file_regular_variable_access(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1359 int test_two_file_regular_variable_access(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1360 {
1361 int error;
1362 cl_program program, second_program, my_newly_linked_program;
1363
1364 const char* sources[2] = {simple_kernel, compile_regular_var}; // here we want to avoid linking error due to lack of kernels
1365 log_info("Compiling and linking two program objects, where one tries to access regular variable from another...\n");
1366 error = create_single_kernel_helper_create_program(context, &program, 2, sources);
1367 if( program == NULL || error != CL_SUCCESS )
1368 {
1369 log_error( "ERROR: Unable to create a test program with regular variable! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1370 return -1;
1371 }
1372
1373 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1374 test_error( error, "Unable to compile a simple program with regular function" );
1375
1376 error = create_single_kernel_helper_create_program(context, &second_program, 1, &link_static_var_access);
1377 if( program == NULL || error != CL_SUCCESS )
1378 {
1379 log_error( "ERROR: Unable to create a test program that tries to access a regular variable! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1380 return -1;
1381 }
1382
1383 error = clCompileProgram(second_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1384 test_error( error, "Unable to compile a program that tries to access a regular variable" );
1385
1386 cl_program two_programs[2] = { program, second_program };
1387 my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, two_programs, NULL, NULL, &error);
1388 test_error( error, "clLinkProgram: Expected a different error code while linking a program that tries to access a regular variable" );
1389
1390 /* All done! */
1391 error = clReleaseProgram( program );
1392 test_error( error, "Unable to release program object" );
1393
1394 error = clReleaseProgram( second_program );
1395 test_error( error, "Unable to release program object" );
1396
1397 error = clReleaseProgram( my_newly_linked_program );
1398 test_error( error, "Unable to release program object" );
1399
1400 return 0;
1401 }
1402
test_two_file_regular_struct_access(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1403 int test_two_file_regular_struct_access(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1404 {
1405 int error;
1406 cl_program program, second_program, my_newly_linked_program;
1407
1408 const char* sources[2] = {simple_kernel, compile_regular_struct}; // here we want to avoid linking error due to lack of kernels
1409 log_info("Compiling and linking two program objects, where one tries to access regular struct from another...\n");
1410 error = create_single_kernel_helper_create_program(context, &program, 2, sources);
1411 if( program == NULL || error != CL_SUCCESS )
1412 {
1413 log_error( "ERROR: Unable to create a test program with regular struct! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1414 return -1;
1415 }
1416
1417 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1418 test_error( error, "Unable to compile a simple program with regular struct" );
1419
1420 error = create_single_kernel_helper_create_program(context, &second_program, 1, &link_static_struct_access);
1421 if( second_program == NULL || error != CL_SUCCESS )
1422 {
1423 log_error( "ERROR: Unable to create a test program that tries to access a regular struct! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1424 return -1;
1425 }
1426
1427 error = clCompileProgram(second_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1428 test_error( error, "Unable to compile a program that tries to access a regular struct" );
1429
1430 cl_program two_programs[2] = { program, second_program };
1431 my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, two_programs, NULL, NULL, &error);
1432 test_error( error, "clLinkProgram: Expected a different error code while linking a program that tries to access a regular struct" );
1433
1434 /* All done! */
1435 error = clReleaseProgram( program );
1436 test_error( error, "Unable to release program object" );
1437
1438 error = clReleaseProgram( second_program );
1439 test_error( error, "Unable to release program object" );
1440
1441 error = clReleaseProgram( my_newly_linked_program );
1442 test_error( error, "Unable to release program object" );
1443
1444 return 0;
1445 }
1446
1447
test_two_file_regular_function_access(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1448 int test_two_file_regular_function_access(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1449 {
1450 int error;
1451 cl_program program, second_program, my_newly_linked_program;
1452
1453 const char* sources[2] = {simple_kernel, compile_regular_function}; // here we want to avoid linking error due to lack of kernels
1454 log_info("Compiling and linking two program objects, where one tries to access regular function from another...\n");
1455 error = create_single_kernel_helper_create_program(context, &program, 2, sources);
1456 if( program == NULL || error != CL_SUCCESS )
1457 {
1458 log_error( "ERROR: Unable to create a test program with regular function! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1459 return -1;
1460 }
1461
1462 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1463 test_error( error, "Unable to compile a simple program with regular function" );
1464
1465 error = create_single_kernel_helper_create_program(context, &second_program, 1, &link_static_function_access);
1466 if( second_program == NULL || error != CL_SUCCESS )
1467 {
1468 log_error( "ERROR: Unable to create a test program that tries to access a regular function! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1469 return -1;
1470 }
1471
1472 error = clCompileProgram(second_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1473 test_error( error, "Unable to compile a program that tries to access a regular function" );
1474
1475 cl_program two_programs[2] = { program, second_program };
1476 my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, two_programs, NULL, NULL, &error);
1477 test_error( error, "clLinkProgram: Expected a different error code while linking a program that tries to access a regular function" );
1478
1479 /* All done! */
1480 error = clReleaseProgram( program );
1481 test_error( error, "Unable to release program object" );
1482
1483 error = clReleaseProgram( second_program );
1484 test_error( error, "Unable to release program object" );
1485
1486 error = clReleaseProgram( my_newly_linked_program );
1487 test_error( error, "Unable to release program object" );
1488
1489 return 0;
1490 }
1491
test_simple_embedded_header_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1492 int test_simple_embedded_header_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1493 {
1494 int error;
1495 cl_program program, header, simple_program;
1496
1497 log_info("Testing a simple embedded header link...\n");
1498 program = clCreateProgramWithSource(context, 1, &another_simple_kernel_with_header, NULL, &error);
1499 if( program == NULL || error != CL_SUCCESS )
1500 {
1501 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1502 return -1;
1503 }
1504
1505 header = clCreateProgramWithSource(context, 1, &simple_header, NULL, &error);
1506 if( header == NULL || error != CL_SUCCESS )
1507 {
1508 log_error( "ERROR: Unable to create a simple header program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1509 return -1;
1510 }
1511
1512 error = clCompileProgram(program, 1, &deviceID, NULL, 1, &header, &simple_header_name, NULL, NULL);
1513 test_error( error, "Unable to compile a simple program with embedded header" );
1514
1515 error = create_single_kernel_helper_create_program(context, &simple_program, 1, &simple_kernel);
1516 if( simple_program == NULL || error != CL_SUCCESS )
1517 {
1518 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1519 return -1;
1520 }
1521
1522 error = clCompileProgram(simple_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1523 test_error( error, "Unable to compile a simple program" );
1524
1525 cl_program two_programs[2] = { program, simple_program };
1526 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error);
1527 test_error( error, "Unable to create an executable from two binaries, one compiled with embedded header" );
1528
1529 /* All done! */
1530 error = clReleaseProgram( program );
1531 test_error( error, "Unable to release program object" );
1532
1533 error = clReleaseProgram( header );
1534 test_error( error, "Unable to release program object" );
1535
1536 error = clReleaseProgram( simple_program );
1537 test_error( error, "Unable to release program object" );
1538
1539 error = clReleaseProgram( fully_linked_program );
1540 test_error( error, "Unable to release program object" );
1541
1542 return 0;
1543 }
1544
1545 const char* when_i_pondered_weak_and_weary = "When I pondered weak and weary!";
1546
simple_link_callback(cl_program program,void * user_data)1547 static void CL_CALLBACK simple_link_callback(cl_program program, void* user_data)
1548 {
1549 simple_user_data* simple_link_user_data = (simple_user_data*)user_data;
1550 log_info("in the simple_link_callback: program %p just completed linking with '%s'\n", program, (const char*)simple_link_user_data->m_message);
1551 if (strcmp(when_i_pondered_weak_and_weary, simple_link_user_data->m_message) != 0)
1552 {
1553 log_error("ERROR: in the simple_compile_callback: Expected '%s' and got %s! (in %s:%d)\n", when_i_pondered_weak_and_weary, simple_link_user_data->m_message, __FILE__, __LINE__);
1554 }
1555
1556 int error;
1557 log_info("in the simple_link_callback: program %p just completed linking with '%p'\n", program, simple_link_user_data->m_event);
1558
1559 error = clSetUserEventStatus(simple_link_user_data->m_event, CL_COMPLETE);
1560 if (error != CL_SUCCESS)
1561 {
1562 log_error( "ERROR: simple_link_callback: Unable to set user event status to CL_COMPLETE! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1563 exit(-1);
1564 }
1565 log_info("in the simple_link_callback: Successfully signaled link_program_completion_event event!\n");
1566 }
1567
test_simple_link_with_callback(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1568 int test_simple_link_with_callback(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1569 {
1570 int error;
1571 cl_program program;
1572 cl_event link_program_completion_event;
1573
1574 log_info("Testing a simple linking with callback...\n");
1575 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1576 if( program == NULL || error != CL_SUCCESS )
1577 {
1578 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1579 return -1;
1580 }
1581
1582 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1583 test_error( error, "Unable to compile a simple program" );
1584
1585 link_program_completion_event = clCreateUserEvent(context, &error);
1586 test_error( error, "Unable to create a user event");
1587
1588 simple_user_data simple_link_user_data = {when_i_pondered_weak_and_weary, link_program_completion_event};
1589
1590 cl_program my_linked_library = clLinkProgram(context, 1, &deviceID, NULL, 1, &program, simple_link_callback, (void*)&simple_link_user_data, &error);
1591 test_error( error, "Unable to link a simple program" );
1592
1593 error = clWaitForEvents(1, &link_program_completion_event);
1594 test_error( error, "clWaitForEvents failed when waiting on link_program_completion_event");
1595
1596 /* All done! */
1597 error = clReleaseEvent(link_program_completion_event);
1598 test_error( error, "Unable to release event object" );
1599
1600 error = clReleaseProgram( program );
1601 test_error( error, "Unable to release program object" );
1602
1603 error = clReleaseProgram( my_linked_library );
1604 test_error( error, "Unable to release program object" );
1605
1606 return 0;
1607 }
1608
initBuffer(float * & srcBuffer,unsigned int cnDimension)1609 static void initBuffer(float* & srcBuffer, unsigned int cnDimension)
1610 {
1611 float num = 0.0f;
1612
1613 for( unsigned int i = 0; i < cnDimension; i++ )
1614 {
1615 if( ( i % 10 ) == 0 )
1616 {
1617 num = 0.0f;
1618 }
1619
1620 srcBuffer[ i ] = num;
1621 num = num + 1.0f;
1622 }
1623 }
1624
verifyCopyBuffer(cl_context context,cl_command_queue queue,cl_kernel kernel)1625 static int verifyCopyBuffer(cl_context context, cl_command_queue queue, cl_kernel kernel)
1626 {
1627 int error, result = CL_SUCCESS;
1628 const size_t cnDimension = 32;
1629
1630 // Allocate source buffer
1631 float * srcBuffer = (float*)malloc(cnDimension * sizeof(float));
1632 float * dstBuffer = (float*)malloc(cnDimension * sizeof(float));
1633
1634 if (srcBuffer == NULL) {
1635 log_error( "ERROR: Unable to allocate srcBuffer float array with %lu floats! (in %s:%d)\n", cnDimension, __FILE__, __LINE__);
1636 return -1;
1637 }
1638 if (dstBuffer == NULL) {
1639 log_error( "ERROR: Unable to allocate dstBuffer float array with %lu floats! (in %s:%d)\n", cnDimension, __FILE__, __LINE__);
1640 return -1;
1641 }
1642
1643 if( srcBuffer && dstBuffer )
1644 {
1645 // initialize host memory
1646 initBuffer(srcBuffer, cnDimension );
1647
1648 // Allocate device memory
1649 cl_mem deviceMemSrc = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
1650 cnDimension * sizeof( cl_float ), srcBuffer, &error);
1651 test_error( error, "Unable to create a source memory buffer" );
1652
1653 cl_mem deviceMemDst = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
1654 cnDimension * sizeof( cl_float ), 0, &error);
1655 test_error( error, "Unable to create a destination memory buffer" );
1656
1657 // Set kernel args
1658 // Set parameter 0 to be the source buffer
1659 error = clSetKernelArg(kernel, 0, sizeof( cl_mem ), ( void * )&deviceMemSrc );
1660 test_error( error, "Unable to set the first kernel argument" );
1661
1662 // Set parameter 1 to be the destination buffer
1663 error = clSetKernelArg(kernel, 1, sizeof( cl_mem ), ( void * )&deviceMemDst );
1664 test_error( error, "Unable to set the second kernel argument" );
1665
1666 // Execute kernel
1667 error = clEnqueueNDRangeKernel(queue, kernel, 1, NULL,
1668 &cnDimension, 0, 0, NULL, NULL );
1669 test_error( error, "Unable to enqueue kernel" );
1670
1671 error = clFlush( queue );
1672 test_error( error, "Unable to flush the queue" );
1673
1674 // copy results from device back to host
1675 error = clEnqueueReadBuffer(queue, deviceMemDst, CL_TRUE, 0, cnDimension * sizeof( cl_float ),
1676 dstBuffer, 0, NULL, NULL );
1677 test_error( error, "Unable to read the destination buffer" );
1678
1679 error = clFlush( queue );
1680 test_error( error, "Unable to flush the queue" );
1681
1682 // Compare the source and destination buffers
1683 const int* pSrc = (int*)srcBuffer;
1684 const int* pDst = (int*)dstBuffer;
1685 int mismatch = 0;
1686
1687 for( size_t i = 0; i < cnDimension; i++ )
1688 {
1689 if( pSrc[i] != pDst[i] )
1690 {
1691 if( mismatch < 4 )
1692 {
1693 log_info("Offset %08lX: Expected %08X, Got %08X\n", i * 4, pSrc[i], pDst[i] );
1694 }
1695 else
1696 {
1697 log_info(".");
1698 }
1699 mismatch++;
1700 }
1701 }
1702
1703 if( mismatch )
1704 {
1705 log_info("*** %d mismatches found, TEST FAILS! ***\n", mismatch );
1706 result = -1;
1707 }
1708 else
1709 {
1710 log_info("Buffers match, test passes.\n");
1711 }
1712
1713 free( srcBuffer );
1714 srcBuffer = NULL;
1715 free( dstBuffer );
1716 dstBuffer = NULL;
1717
1718 if( deviceMemSrc )
1719 {
1720 error = clReleaseMemObject( deviceMemSrc );
1721 test_error( error, "Unable to release memory object" );
1722 }
1723
1724 if( deviceMemDst )
1725 {
1726 error = clReleaseMemObject( deviceMemDst );
1727 test_error( error, "Unable to release memory object" );
1728 }
1729 }
1730 return result;
1731 }
1732
test_execute_after_simple_compile_and_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1733 int test_execute_after_simple_compile_and_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1734 {
1735 int error;
1736 cl_program program;
1737
1738 log_info("Testing execution after a simple compile and link...\n");
1739 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1740 if( program == NULL || error != CL_SUCCESS )
1741 {
1742 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1743 return -1;
1744 }
1745
1746 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1747 test_error( error, "Unable to compile a simple program" );
1748
1749 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 1, &program, NULL, NULL, &error);
1750 test_error( error, "Unable to link a simple program" );
1751
1752 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CopyBuffer", &error);
1753 test_error( error, "Unable to create a simple kernel" );
1754
1755 error = verifyCopyBuffer(context, queue, kernel);
1756 if (error != CL_SUCCESS)
1757 return error;
1758
1759 /* All done! */
1760 error = clReleaseKernel( kernel );
1761 test_error( error, "Unable to release kernel object" );
1762
1763 error = clReleaseProgram( program );
1764 test_error( error, "Unable to release program object" );
1765
1766 error = clReleaseProgram( my_newly_linked_program );
1767 test_error( error, "Unable to release program object" );
1768
1769 return 0;
1770 }
1771
test_execute_after_simple_compile_and_link_no_device_info(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1772 int test_execute_after_simple_compile_and_link_no_device_info(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1773 {
1774 int error;
1775 cl_program program;
1776
1777 log_info("Testing execution after a simple compile and link with no device information provided...\n");
1778 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1779 if( program == NULL || error != CL_SUCCESS )
1780 {
1781 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1782 return -1;
1783 }
1784
1785 error = clCompileProgram(program, 0, NULL, NULL, 0, NULL, NULL, NULL, NULL);
1786 test_error( error, "Unable to compile a simple program" );
1787
1788 cl_program my_newly_linked_program = clLinkProgram(context, 0, NULL, NULL, 1, &program, NULL, NULL, &error);
1789 test_error( error, "Unable to link a simple program" );
1790
1791 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CopyBuffer", &error);
1792 test_error( error, "Unable to create a simple kernel" );
1793
1794 error = verifyCopyBuffer(context, queue, kernel);
1795 if (error != CL_SUCCESS)
1796 return error;
1797
1798 /* All done! */
1799 error = clReleaseKernel( kernel );
1800 test_error( error, "Unable to release kernel object" );
1801
1802 error = clReleaseProgram( program );
1803 test_error( error, "Unable to release program object" );
1804
1805 error = clReleaseProgram( my_newly_linked_program );
1806 test_error( error, "Unable to release program object" );
1807
1808 return 0;
1809 }
1810
test_execute_after_simple_compile_and_link_with_defines(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1811 int test_execute_after_simple_compile_and_link_with_defines(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1812 {
1813 int error;
1814 cl_program program;
1815
1816 log_info("Testing execution after a simple compile and link with defines...\n");
1817 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel_with_defines, "-DFIRST=5 -DSECOND=37");
1818 if( program == NULL || error != CL_SUCCESS )
1819 {
1820 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1821 return -1;
1822 }
1823
1824 error = clCompileProgram(program, 1, &deviceID, "-DFIRST=5 -DSECOND=37", 0, NULL, NULL, NULL, NULL);
1825 test_error( error, "Unable to compile a simple program" );
1826
1827 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 1, &program, NULL, NULL, &error);
1828 test_error( error, "Unable to link a simple program" );
1829
1830 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CopyBuffer", &error);
1831 test_error( error, "Unable to create a simple kernel" );
1832
1833 error = verifyCopyBuffer(context, queue, kernel);
1834 if (error != CL_SUCCESS)
1835 return error;
1836
1837 /* All done! */
1838 error = clReleaseKernel( kernel );
1839 test_error( error, "Unable to release kernel object" );
1840
1841 error = clReleaseProgram( program );
1842 test_error( error, "Unable to release program object" );
1843
1844 error = clReleaseProgram( my_newly_linked_program );
1845 test_error( error, "Unable to release program object" );
1846
1847 return 0;
1848 }
1849
test_execute_after_serialize_reload_object(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1850 int test_execute_after_serialize_reload_object(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1851 {
1852 int error;
1853 cl_program program;
1854 size_t binarySize;
1855 unsigned char *binary;
1856
1857 log_info("Testing execution after serialization and reloading of the object...\n");
1858 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1859 if( program == NULL || error != CL_SUCCESS )
1860 {
1861 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1862 return -1;
1863 }
1864
1865 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1866 test_error( error, "Unable to compile a simple program" );
1867
1868 // Get the size of the resulting binary (only one device)
1869 error = clGetProgramInfo( program, CL_PROGRAM_BINARY_SIZES, sizeof( binarySize ), &binarySize, NULL );
1870 test_error( error, "Unable to get binary size" );
1871
1872 // Sanity check
1873 if( binarySize == 0 )
1874 {
1875 log_error( "ERROR: Binary size of program is zero (in %s:%d)\n", __FILE__, __LINE__ );
1876 return -1;
1877 }
1878
1879 // Create a buffer and get the actual binary
1880 binary = (unsigned char*)malloc(sizeof(unsigned char)*binarySize);
1881 if (binary == NULL) {
1882 log_error( "ERROR: Unable to allocate binary character array with %lu characters! (in %s:%d)\n", binarySize, __FILE__, __LINE__ );
1883 return -1;
1884 }
1885
1886 unsigned char *buffers[ 1 ] = { binary };
1887 cl_int loadErrors[ 1 ];
1888
1889 // Do another sanity check here first
1890 size_t size;
1891 error = clGetProgramInfo( program, CL_PROGRAM_BINARIES, 0, NULL, &size );
1892 test_error( error, "Unable to get expected size of binaries array" );
1893 if( size != sizeof( buffers ) )
1894 {
1895 log_error( "ERROR: Expected size of binaries array in clGetProgramInfo is incorrect (should be %d, got %d) (in %s:%d)\n", (int)sizeof( buffers ), (int)size, __FILE__, __LINE__ );
1896 free(binary);
1897 return -1;
1898 }
1899
1900 error = clGetProgramInfo( program, CL_PROGRAM_BINARIES, sizeof( buffers ), &buffers, NULL );
1901 test_error( error, "Unable to get program binary" );
1902
1903 // use clCreateProgramWithBinary
1904 cl_program program_with_binary = clCreateProgramWithBinary(context, 1, &deviceID, &binarySize, (const unsigned char**)buffers, loadErrors, &error);
1905 test_error( error, "Unable to create program with binary" );
1906
1907 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 1, &program_with_binary, NULL, NULL, &error);
1908 test_error( error, "Unable to link a simple program" );
1909
1910 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CopyBuffer", &error);
1911 test_error( error, "Unable to create a simple kernel" );
1912
1913 error = verifyCopyBuffer(context, queue, kernel);
1914 if (error != CL_SUCCESS)
1915 return error;
1916
1917 /* All done! */
1918 error = clReleaseKernel( kernel );
1919 test_error( error, "Unable to release kernel object" );
1920
1921 error = clReleaseProgram( program );
1922 test_error( error, "Unable to release program object" );
1923
1924 error = clReleaseProgram( my_newly_linked_program );
1925 test_error( error, "Unable to release program object" );
1926
1927 error = clReleaseProgram( program_with_binary );
1928 test_error( error, "Unable to release program object" );
1929
1930 free(binary);
1931
1932 return 0;
1933 }
1934
test_execute_after_serialize_reload_library(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)1935 int test_execute_after_serialize_reload_library(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
1936 {
1937 int error;
1938 cl_program program, another_program;
1939 size_t binarySize;
1940 unsigned char *binary;
1941
1942 log_info("Testing execution after linking a binary with a simple library...\n");
1943 // we will test creation of a simple library from one file
1944 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
1945 if( program == NULL || error != CL_SUCCESS )
1946 {
1947 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
1948 return -1;
1949 }
1950
1951 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
1952 test_error( error, "Unable to compile a simple program" );
1953
1954 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, NULL, NULL, &error);
1955 test_error( error, "Unable to create a simple library" );
1956
1957
1958 // Get the size of the resulting library (only one device)
1959 error = clGetProgramInfo( my_newly_minted_library, CL_PROGRAM_BINARY_SIZES, sizeof( binarySize ), &binarySize, NULL );
1960 test_error( error, "Unable to get binary size" );
1961
1962 // Sanity check
1963 if( binarySize == 0 )
1964 {
1965 log_error( "ERROR: Binary size of program is zero (in %s:%d)\n", __FILE__, __LINE__ );
1966 return -1;
1967 }
1968
1969 // Create a buffer and get the actual binary
1970 binary = (unsigned char*)malloc(sizeof(unsigned char)*binarySize);
1971 if (binary == NULL) {
1972 log_error( "ERROR: Unable to allocate binary character array with %lu characters (in %s:%d)!", binarySize, __FILE__, __LINE__);
1973 return -1;
1974 }
1975 unsigned char *buffers[ 1 ] = { binary };
1976 cl_int loadErrors[ 1 ];
1977
1978 // Do another sanity check here first
1979 size_t size;
1980 error = clGetProgramInfo( my_newly_minted_library, CL_PROGRAM_BINARIES, 0, NULL, &size );
1981 test_error( error, "Unable to get expected size of binaries array" );
1982 if( size != sizeof( buffers ) )
1983 {
1984 log_error( "ERROR: Expected size of binaries array in clGetProgramInfo is incorrect (should be %d, got %d) (in %s:%d)\n", (int)sizeof( buffers ), (int)size, __FILE__, __LINE__ );
1985 free(binary);
1986 return -1;
1987 }
1988
1989 error = clGetProgramInfo( my_newly_minted_library, CL_PROGRAM_BINARIES, sizeof( buffers ), &buffers, NULL );
1990 test_error( error, "Unable to get program binary" );
1991
1992 // use clCreateProgramWithBinary
1993 cl_program library_with_binary = clCreateProgramWithBinary(context, 1, &deviceID, &binarySize, (const unsigned char**)buffers, loadErrors, &error);
1994 test_error( error, "Unable to create program with binary" );
1995
1996 error = create_single_kernel_helper_create_program(context, &another_program, 1, &another_simple_kernel);
1997 if( another_program == NULL || error != CL_SUCCESS )
1998 {
1999 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2000 return -1;
2001 }
2002
2003 error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2004 test_error( error, "Unable to compile a simple program" );
2005
2006 cl_program program_and_archive[2] = { another_program, library_with_binary };
2007 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error);
2008 test_error( error, "Unable to create an executable from a binary and a library" );
2009
2010 cl_kernel kernel = clCreateKernel(fully_linked_program, "CopyBuffer", &error);
2011 test_error( error, "Unable to create a simple kernel" );
2012
2013 error = verifyCopyBuffer(context, queue, kernel);
2014 if (error != CL_SUCCESS)
2015 return error;
2016
2017 cl_kernel another_kernel = clCreateKernel(fully_linked_program, "AnotherCopyBuffer", &error);
2018 test_error( error, "Unable to create another simple kernel" );
2019
2020 error = verifyCopyBuffer(context, queue, another_kernel);
2021 if (error != CL_SUCCESS)
2022 return error;
2023
2024 /* All done! */
2025 error = clReleaseKernel( kernel );
2026 test_error( error, "Unable to release kernel object" );
2027
2028 error = clReleaseKernel( another_kernel );
2029 test_error( error, "Unable to release another kernel object" );
2030
2031 error = clReleaseProgram( program );
2032 test_error( error, "Unable to release program object" );
2033
2034 error = clReleaseProgram( another_program );
2035 test_error( error, "Unable to release program object" );
2036
2037 error = clReleaseProgram( my_newly_minted_library );
2038 test_error( error, "Unable to release program object" );
2039
2040 error = clReleaseProgram( library_with_binary );
2041 test_error( error, "Unable to release program object" );
2042
2043 error = clReleaseProgram( fully_linked_program );
2044 test_error( error, "Unable to release program object" );
2045
2046 free(binary);
2047
2048 return 0;
2049 }
2050
program_compile_completion_callback(cl_program program,void * user_data)2051 static void CL_CALLBACK program_compile_completion_callback(cl_program program, void* user_data)
2052 {
2053 int error;
2054 cl_event compile_program_completion_event = (cl_event)user_data;
2055 log_info("in the program_compile_completion_callback: program %p just completed compiling with '%p'\n", program, compile_program_completion_event);
2056
2057 error = clSetUserEventStatus(compile_program_completion_event, CL_COMPLETE);
2058 if (error != CL_SUCCESS)
2059 {
2060 log_error( "ERROR: in the program_compile_completion_callback: Unable to set user event status to CL_COMPLETE! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2061 exit(-1);
2062 }
2063 log_info("in the program_compile_completion_callback: Successfully signaled compile_program_completion_event event!\n");
2064 }
2065
program_link_completion_callback(cl_program program,void * user_data)2066 static void CL_CALLBACK program_link_completion_callback(cl_program program, void* user_data)
2067 {
2068 int error;
2069 cl_event link_program_completion_event = (cl_event)user_data;
2070 log_info("in the program_link_completion_callback: program %p just completed linking with '%p'\n", program, link_program_completion_event);
2071
2072 error = clSetUserEventStatus(link_program_completion_event, CL_COMPLETE);
2073 if (error != CL_SUCCESS)
2074 {
2075 log_error( "ERROR: in the program_link_completion_callback: Unable to set user event status to CL_COMPLETE! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2076 exit(-1);
2077 }
2078 log_info("in the program_link_completion_callback: Successfully signaled link_program_completion_event event!\n");
2079 }
2080
test_execute_after_simple_compile_and_link_with_callbacks(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2081 int test_execute_after_simple_compile_and_link_with_callbacks(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2082 {
2083 int error;
2084 cl_program program;
2085 cl_event compile_program_completion_event, link_program_completion_event;
2086
2087 log_info("Testing execution after a simple compile and link with callbacks...\n");
2088 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2089 if( program == NULL || error != CL_SUCCESS )
2090 {
2091 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2092 return -1;
2093 }
2094
2095 compile_program_completion_event = clCreateUserEvent(context, &error);
2096 test_error( error, "Unable to create a user event");
2097
2098 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL,
2099 program_compile_completion_callback, (void*)compile_program_completion_event);
2100 test_error( error, "Unable to compile a simple program" );
2101
2102 error = clWaitForEvents(1, &compile_program_completion_event);
2103 test_error( error, "clWaitForEvents failed when waiting on compile_program_completion_event");
2104
2105 error = clReleaseEvent(compile_program_completion_event);
2106 test_error( error, "Unable to release event object" );
2107
2108 link_program_completion_event = clCreateUserEvent(context, &error);
2109 test_error( error, "Unable to create a user event");
2110
2111 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 1, &program,
2112 program_link_completion_callback, (void*)link_program_completion_event, &error);
2113 test_error( error, "Unable to link a simple program" );
2114
2115 error = clWaitForEvents(1, &link_program_completion_event);
2116 test_error( error, "clWaitForEvents failed when waiting on link_program_completion_event");
2117
2118 error = clReleaseEvent(link_program_completion_event);
2119 test_error( error, "Unable to release event object" );
2120
2121 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CopyBuffer", &error);
2122 test_error( error, "Unable to create a simple kernel" );
2123
2124 error = verifyCopyBuffer(context, queue, kernel);
2125 if (error != CL_SUCCESS)
2126 return error;
2127
2128 /* All done! */
2129 error = clReleaseKernel( kernel );
2130 test_error( error, "Unable to release kernel object" );
2131
2132 error = clReleaseProgram( program );
2133 test_error( error, "Unable to release program object" );
2134
2135 error = clReleaseProgram( my_newly_linked_program );
2136 test_error( error, "Unable to release program object" );
2137
2138 return 0;
2139 }
2140
test_simple_library_only(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2141 int test_simple_library_only(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2142 {
2143 int error;
2144 cl_program program;
2145
2146 log_info("Testing creation of a simple library...\n");
2147 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2148 if( program == NULL || error != CL_SUCCESS )
2149 {
2150 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2151 return -1;
2152 }
2153
2154 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2155 test_error( error, "Unable to compile a simple program" );
2156
2157 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, NULL, NULL, &error);
2158 test_error( error, "Unable to create a simple library" );
2159
2160 /* All done! */
2161 error = clReleaseProgram( program );
2162 test_error( error, "Unable to release program object" );
2163
2164 error = clReleaseProgram( my_newly_minted_library );
2165 test_error( error, "Unable to release program object" );
2166
2167 return 0;
2168 }
2169
test_simple_library_with_callback(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2170 int test_simple_library_with_callback(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2171 {
2172 int error;
2173 cl_program program;
2174 cl_event link_program_completion_event;
2175
2176 log_info("Testing creation of a simple library with a callback...\n");
2177 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2178 if( program == NULL || error != CL_SUCCESS )
2179 {
2180 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2181 return -1;
2182 }
2183
2184 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2185 test_error( error, "Unable to compile a simple program" );
2186
2187 link_program_completion_event = clCreateUserEvent(context, &error);
2188 test_error( error, "Unable to create a user event");
2189
2190 simple_user_data simple_link_user_data = {when_i_pondered_weak_and_weary, link_program_completion_event};
2191
2192 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program,
2193 simple_link_callback, (void*)&simple_link_user_data, &error);
2194 test_error( error, "Unable to create a simple library" );
2195
2196 error = clWaitForEvents(1, &link_program_completion_event);
2197 test_error( error, "clWaitForEvents failed when waiting on link_program_completion_event");
2198
2199 /* All done! */
2200 error = clReleaseEvent(link_program_completion_event);
2201 test_error( error, "Unable to release event object" );
2202
2203 error = clReleaseProgram( program );
2204 test_error( error, "Unable to release program object" );
2205
2206 error = clReleaseProgram( my_newly_minted_library );
2207 test_error( error, "Unable to release program object" );
2208
2209 return 0;
2210 }
2211
test_simple_library_with_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2212 int test_simple_library_with_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2213 {
2214 int error;
2215 cl_program program, another_program;
2216
2217 log_info("Testing creation and linking with a simple library...\n");
2218 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2219 if( program == NULL || error != CL_SUCCESS )
2220 {
2221 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2222 return -1;
2223 }
2224
2225 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2226 test_error( error, "Unable to compile a simple program" );
2227
2228 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, NULL, NULL, &error);
2229 test_error( error, "Unable to create a simple library" );
2230
2231 error = create_single_kernel_helper_create_program(context, &another_program, 1, &another_simple_kernel);
2232 if( another_program == NULL || error != CL_SUCCESS )
2233 {
2234 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2235 return -1;
2236 }
2237
2238 error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2239 test_error( error, "Unable to compile a simple program" );
2240
2241 cl_program program_and_archive[2] = { another_program, my_newly_minted_library };
2242 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error);
2243 test_error( error, "Unable to create an executable from a binary and a library" );
2244
2245 /* All done! */
2246 error = clReleaseProgram( program );
2247 test_error( error, "Unable to release program object" );
2248
2249 error = clReleaseProgram( another_program );
2250 test_error( error, "Unable to release program object" );
2251
2252 error = clReleaseProgram( my_newly_minted_library );
2253 test_error( error, "Unable to release program object" );
2254
2255 error = clReleaseProgram( fully_linked_program );
2256 test_error( error, "Unable to release program object" );
2257
2258 return 0;
2259 }
2260
test_execute_after_simple_library_with_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2261 int test_execute_after_simple_library_with_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2262 {
2263 int error;
2264 cl_program program, another_program;
2265
2266 log_info("Testing execution after linking a binary with a simple library...\n");
2267 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2268 if( program == NULL || error != CL_SUCCESS )
2269 {
2270 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2271 return -1;
2272 }
2273
2274 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2275 test_error( error, "Unable to compile a simple program" );
2276
2277 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program, NULL, NULL, &error);
2278 test_error( error, "Unable to create a simple library" );
2279
2280 error = create_single_kernel_helper_create_program(context, &another_program, 1, &another_simple_kernel);
2281 if( another_program == NULL || error != CL_SUCCESS )
2282 {
2283 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2284 return -1;
2285 }
2286
2287 error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2288 test_error( error, "Unable to compile a simple program" );
2289
2290 cl_program program_and_archive[2] = { another_program, my_newly_minted_library };
2291 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error);
2292 test_error( error, "Unable to create an executable from a binary and a library" );
2293
2294 cl_kernel kernel = clCreateKernel(fully_linked_program, "CopyBuffer", &error);
2295 test_error( error, "Unable to create a simple kernel" );
2296
2297 error = verifyCopyBuffer(context, queue, kernel);
2298 if (error != CL_SUCCESS)
2299 return error;
2300
2301 cl_kernel another_kernel = clCreateKernel(fully_linked_program, "AnotherCopyBuffer", &error);
2302 test_error( error, "Unable to create another simple kernel" );
2303
2304 error = verifyCopyBuffer(context, queue, another_kernel);
2305 if (error != CL_SUCCESS)
2306 return error;
2307
2308 /* All done! */
2309 error = clReleaseKernel( kernel );
2310 test_error( error, "Unable to release kernel object" );
2311
2312 error = clReleaseKernel( another_kernel );
2313 test_error( error, "Unable to release another kernel object" );
2314
2315 error = clReleaseProgram( program );
2316 test_error( error, "Unable to release program object" );
2317
2318 error = clReleaseProgram( another_program );
2319 test_error( error, "Unable to release program object" );
2320
2321 error = clReleaseProgram( my_newly_minted_library );
2322 test_error( error, "Unable to release program object" );
2323
2324 error = clReleaseProgram( fully_linked_program );
2325 test_error( error, "Unable to release program object" );
2326
2327 return 0;
2328 }
2329
test_two_file_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2330 int test_two_file_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2331 {
2332 int error;
2333 cl_program program, another_program;
2334
2335 log_info("Testing two file compiling and linking...\n");
2336 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2337 if( program == NULL || error != CL_SUCCESS )
2338 {
2339 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2340 return -1;
2341 }
2342
2343 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2344 test_error( error, "Unable to compile a simple program" );
2345
2346
2347 error = create_single_kernel_helper_create_program(context, &another_program, 1, &another_simple_kernel);
2348 if( another_program == NULL || error != CL_SUCCESS )
2349 {
2350 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2351 return -1;
2352 }
2353
2354 error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2355 test_error( error, "Unable to compile a simple program" );
2356
2357 cl_program two_programs[2] = { program, another_program };
2358 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error);
2359 test_error( error, "Unable to create an executable from two binaries" );
2360
2361 /* All done! */
2362 error = clReleaseProgram( program );
2363 test_error( error, "Unable to release program object" );
2364
2365 error = clReleaseProgram( another_program );
2366 test_error( error, "Unable to release program object" );
2367
2368 error = clReleaseProgram( fully_linked_program );
2369 test_error( error, "Unable to release program object" );
2370
2371 return 0;
2372 }
2373
test_execute_after_two_file_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2374 int test_execute_after_two_file_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2375 {
2376 int error;
2377 cl_program program, another_program;
2378
2379 log_info("Testing two file compiling and linking and execution of two kernels afterwards ...\n");
2380 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2381 if( program == NULL || error != CL_SUCCESS )
2382 {
2383 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2384 return -1;
2385 }
2386
2387 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2388 test_error( error, "Unable to compile a simple program" );
2389
2390 error = create_single_kernel_helper_create_program(context, &another_program, 1, &another_simple_kernel);
2391 if( another_program == NULL || error != CL_SUCCESS )
2392 {
2393 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2394 return -1;
2395 }
2396
2397 error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2398 test_error( error, "Unable to compile a simple program" );
2399
2400 cl_program two_programs[2] = { program, another_program };
2401 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error);
2402 test_error( error, "Unable to create an executable from two binaries" );
2403
2404 cl_kernel kernel = clCreateKernel(fully_linked_program, "CopyBuffer", &error);
2405 test_error( error, "Unable to create a simple kernel" );
2406
2407 error = verifyCopyBuffer(context, queue, kernel);
2408 if (error != CL_SUCCESS)
2409 return error;
2410
2411 cl_kernel another_kernel = clCreateKernel(fully_linked_program, "AnotherCopyBuffer", &error);
2412 test_error( error, "Unable to create another simple kernel" );
2413
2414 error = verifyCopyBuffer(context, queue, another_kernel);
2415 if (error != CL_SUCCESS)
2416 return error;
2417
2418 /* All done! */
2419 error = clReleaseKernel( kernel );
2420 test_error( error, "Unable to release kernel object" );
2421
2422 error = clReleaseKernel( another_kernel );
2423 test_error( error, "Unable to release another kernel object" );
2424
2425 error = clReleaseProgram( program );
2426 test_error( error, "Unable to release program object" );
2427
2428 error = clReleaseProgram( another_program );
2429 test_error( error, "Unable to release program object" );
2430
2431 error = clReleaseProgram( fully_linked_program );
2432 test_error( error, "Unable to release program object" );
2433
2434 return 0;
2435 }
2436
test_execute_after_embedded_header_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2437 int test_execute_after_embedded_header_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2438 {
2439 int error;
2440 cl_program program, header, simple_program;
2441
2442 log_info("Testing execution after embedded header link...\n");
2443 // we will test execution after compiling and linking with embedded headers
2444 program = clCreateProgramWithSource(context, 1, &another_simple_kernel_with_header, NULL, &error);
2445 if( program == NULL || error != CL_SUCCESS )
2446 {
2447 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2448 return -1;
2449 }
2450
2451 header = clCreateProgramWithSource(context, 1, &simple_header, NULL, &error);
2452 if( header == NULL || error != CL_SUCCESS )
2453 {
2454 log_error( "ERROR: Unable to create a simple header program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2455 return -1;
2456 }
2457
2458 error = clCompileProgram(program, 1, &deviceID, NULL, 1, &header, &simple_header_name, NULL, NULL);
2459 test_error( error, "Unable to compile a simple program with embedded header" );
2460
2461 simple_program = clCreateProgramWithSource(context, 1, &simple_kernel, NULL, &error);
2462 if( simple_program == NULL || error != CL_SUCCESS )
2463 {
2464 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2465 return -1;
2466 }
2467
2468 error = clCompileProgram(simple_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2469 test_error( error, "Unable to compile a simple program" );
2470
2471 cl_program two_programs[2] = { program, simple_program };
2472 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error);
2473 test_error( error, "Unable to create an executable from two binaries, one compiled with embedded header" );
2474
2475 cl_kernel kernel = clCreateKernel(fully_linked_program, "CopyBuffer", &error);
2476 test_error( error, "Unable to create a simple kernel" );
2477
2478 error = verifyCopyBuffer(context, queue, kernel);
2479 if (error != CL_SUCCESS)
2480 return error;
2481
2482 cl_kernel another_kernel = clCreateKernel(fully_linked_program, "AnotherCopyBuffer", &error);
2483 test_error( error, "Unable to create another simple kernel" );
2484
2485 error = verifyCopyBuffer(context, queue, another_kernel);
2486 if (error != CL_SUCCESS)
2487 return error;
2488
2489 /* All done! */
2490 error = clReleaseKernel( kernel );
2491 test_error( error, "Unable to release kernel object" );
2492
2493 error = clReleaseKernel( another_kernel );
2494 test_error( error, "Unable to release another kernel object" );
2495
2496 error = clReleaseProgram( program );
2497 test_error( error, "Unable to release program object" );
2498
2499 error = clReleaseProgram( header );
2500 test_error( error, "Unable to release program object" );
2501
2502 error = clReleaseProgram( simple_program );
2503 test_error( error, "Unable to release program object" );
2504
2505 error = clReleaseProgram( fully_linked_program );
2506 test_error( error, "Unable to release program object" );
2507
2508 return 0;
2509 }
2510
2511 #if defined(__APPLE__) || defined(__linux)
2512 #define _mkdir(x) mkdir(x,S_IRWXU)
2513 #define _chdir chdir
2514 #define _rmdir rmdir
2515 #define _unlink unlink
2516 #else
2517 #include <direct.h>
2518 #endif
2519
test_execute_after_included_header_link(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2520 int test_execute_after_included_header_link(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2521 {
2522 int error;
2523 cl_program program, simple_program;
2524
2525 log_info("Testing execution after included header link...\n");
2526 // we will test execution after compiling and linking with included headers
2527 program = clCreateProgramWithSource(context, 1, &another_simple_kernel_with_header, NULL, &error);
2528 if( program == NULL || error != CL_SUCCESS )
2529 {
2530 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2531 return -1;
2532 }
2533
2534 /* setup */
2535 #if (defined(__linux__) || defined(__APPLE__)) && (!defined( __ANDROID__ ))
2536 /* Some tests systems doesn't allow one to write in the test directory */
2537 if (_chdir("/tmp") != 0) {
2538 log_error( "ERROR: Unable to remove directory foo/bar! (in %s:%d)\n", __FILE__, __LINE__ );
2539 return -1;
2540 }
2541 #endif
2542 if (_mkdir("foo") != 0) {
2543 log_error( "ERROR: Unable to create directory foo! (in %s:%d)\n", __FILE__, __LINE__ );
2544 return -1;
2545 }
2546 if (_mkdir("foo/bar") != 0) {
2547 log_error( "ERROR: Unable to create directory foo/bar! (in %s:%d)\n", __FILE__, __LINE__ );
2548 return -1;
2549 }
2550 if (_chdir("foo/bar") != 0) {
2551 log_error( "ERROR: Unable to change to directory foo/bar! (in %s:%d)\n", __FILE__, __LINE__ );
2552 return -1;
2553 }
2554 FILE* simple_header_file = fopen(simple_header_name, "w");
2555 if (simple_header_file == NULL) {
2556 log_error( "ERROR: Unable to create simple header file %s! (in %s:%d)\n", simple_header_name, __FILE__, __LINE__ );
2557 return -1;
2558 }
2559 if (fprintf(simple_header_file, "%s", simple_header) < 0) {
2560 log_error( "ERROR: Unable to write to simple header file %s! (in %s:%d)\n", simple_header_name, __FILE__, __LINE__);
2561 return -1;
2562 }
2563 if (fclose(simple_header_file) != 0) {
2564 log_error( "ERROR: Unable to close simple header file %s! (in %s:%d)\n", simple_header_name, __FILE__, __LINE__);
2565 return -1;
2566 }
2567 if (_chdir("../..") != 0) {
2568 log_error( "ERROR: Unable to change to original working directory! (in %s:%d)\n", __FILE__, __LINE__);
2569 return -1;
2570 }
2571 #if (defined(__linux__) || defined(__APPLE__)) && (!defined( __ANDROID__ ))
2572 error = clCompileProgram(program, 1, &deviceID, "-I/tmp/foo/bar", 0, NULL, NULL, NULL, NULL);
2573 #else
2574 error = clCompileProgram(program, 1, &deviceID, "-Ifoo/bar", 0, NULL, NULL, NULL, NULL);
2575 #endif
2576 test_error( error, "Unable to compile a simple program with included header" );
2577
2578 /* cleanup */
2579 if (_chdir("foo/bar") != 0) {
2580 log_error( "ERROR: Unable to change to directory foo/bar! (in %s:%d)\n", __FILE__, __LINE__ );
2581 return -1;
2582 }
2583 if (_unlink(simple_header_name) != 0) {
2584 log_error( "ERROR: Unable to remove simple header file %s! (in %s:%d)\n", simple_header_name, __FILE__, __LINE__ );
2585 return -1;
2586 }
2587 if (_chdir("../..") != 0) {
2588 log_error( "ERROR: Unable to change to original working directory! (in %s:%d)\n", __FILE__, __LINE__ );
2589 return -1;
2590 }
2591 if (_rmdir("foo/bar") != 0) {
2592 log_error( "ERROR: Unable to remove directory foo/bar! (in %s:%d)\n", __FILE__, __LINE__ );
2593 return -1;
2594 }
2595 if (_rmdir("foo") != 0) {
2596 log_error( "ERROR: Unable to remove directory foo! (in %s:%d)\n", __FILE__, __LINE__ );
2597 return -1;
2598 }
2599
2600 simple_program = clCreateProgramWithSource(context, 1, &simple_kernel, NULL, &error);
2601 if( simple_program == NULL || error != CL_SUCCESS )
2602 {
2603 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2604 return -1;
2605 }
2606
2607 error = clCompileProgram(simple_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2608 test_error( error, "Unable to compile a simple program" );
2609
2610 cl_program two_programs[2] = { program, simple_program };
2611 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, two_programs, NULL, NULL, &error);
2612 test_error( error, "Unable to create an executable from two binaries, one compiled with embedded header" );
2613
2614 cl_kernel kernel = clCreateKernel(fully_linked_program, "CopyBuffer", &error);
2615 test_error( error, "Unable to create a simple kernel" );
2616
2617 error = verifyCopyBuffer(context, queue, kernel);
2618 if (error != CL_SUCCESS)
2619 return error;
2620
2621 cl_kernel another_kernel = clCreateKernel(fully_linked_program, "AnotherCopyBuffer", &error);
2622 test_error( error, "Unable to create another simple kernel" );
2623
2624 error = verifyCopyBuffer(context, queue, another_kernel);
2625 if (error != CL_SUCCESS)
2626 return error;
2627
2628 /* All done! */
2629 error = clReleaseKernel( kernel );
2630 test_error( error, "Unable to release kernel object" );
2631
2632 error = clReleaseKernel( another_kernel );
2633 test_error( error, "Unable to release another kernel object" );
2634
2635 error = clReleaseProgram( program );
2636 test_error( error, "Unable to release program object" );
2637
2638 error = clReleaseProgram( simple_program );
2639 test_error( error, "Unable to release program object" );
2640
2641 error = clReleaseProgram( fully_linked_program );
2642 test_error( error, "Unable to release program object" );
2643
2644 return 0;
2645 }
2646
test_program_binary_type(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)2647 int test_program_binary_type(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
2648 {
2649 int error;
2650 cl_program program, another_program, program_with_binary, fully_linked_program_with_binary;
2651 cl_program_binary_type program_type = -1;
2652 size_t size;
2653 size_t binarySize;
2654 unsigned char *binary;
2655
2656 log_info("Testing querying of program binary type...\n");
2657 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2658 if( program == NULL || error != CL_SUCCESS )
2659 {
2660 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2661 return -1;
2662 }
2663
2664 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2665 test_error( error, "Unable to compile a simple program" );
2666
2667 error = clGetProgramBuildInfo (program, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2668 test_error( error, "Unable to get program binary type" );
2669 if (program_type != CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT)
2670 {
2671 log_error( "ERROR: Expected program type of a just compiled program to be CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT (in %s:%d)\n", __FILE__, __LINE__ );
2672 return -1;
2673 }
2674 program_type = -1;
2675
2676 // Get the size of the resulting binary (only one device)
2677 error = clGetProgramInfo( program, CL_PROGRAM_BINARY_SIZES, sizeof( binarySize ), &binarySize, NULL );
2678 test_error( error, "Unable to get binary size" );
2679
2680 // Sanity check
2681 if( binarySize == 0 )
2682 {
2683 log_error( "ERROR: Binary size of program is zero (in %s:%d)\n", __FILE__, __LINE__ );
2684 return -1;
2685 }
2686
2687 // Create a buffer and get the actual binary
2688 {
2689 binary = (unsigned char*)malloc(sizeof(unsigned char)*binarySize);
2690 if (binary == NULL) {
2691 log_error( "ERROR: Unable to allocate binary character array with %lu characters! (in %s:%d)\n", binarySize, __FILE__, __LINE__ );
2692 return -1;
2693 }
2694 unsigned char *buffers[ 1 ] = { binary };
2695 cl_int loadErrors[ 1 ];
2696
2697 // Do another sanity check here first
2698 size_t size;
2699 error = clGetProgramInfo( program, CL_PROGRAM_BINARIES, 0, NULL, &size );
2700 test_error( error, "Unable to get expected size of binaries array" );
2701 if( size != sizeof( buffers ) )
2702 {
2703 log_error( "ERROR: Expected size of binaries array in clGetProgramInfo is incorrect (should be %d, got %d) (in %s:%d)\n", (int)sizeof( buffers ), (int)size, __FILE__, __LINE__ );
2704 free(binary);
2705 return -1;
2706 }
2707
2708 error = clGetProgramInfo( program, CL_PROGRAM_BINARIES, sizeof( buffers ), &buffers, NULL );
2709 test_error( error, "Unable to get program binary" );
2710
2711 // use clCreateProgramWithBinary
2712 program_with_binary = clCreateProgramWithBinary(context, 1, &deviceID, &binarySize, (const unsigned char**)buffers, loadErrors, &error);
2713 test_error( error, "Unable to create program with binary" );
2714
2715 error = clGetProgramBuildInfo (program_with_binary, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2716 test_error( error, "Unable to get program binary type" );
2717 if (program_type != CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT)
2718 {
2719 log_error( "ERROR: Expected program type of a program created from compiled object to be CL_PROGRAM_BINARY_TYPE_COMPILED_OBJECT (in %s:%d)\n", __FILE__, __LINE__ );
2720 return -1;
2721 }
2722 program_type = -1;
2723 free(binary);
2724 }
2725
2726 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", 1, &program_with_binary, NULL, NULL, &error);
2727 test_error( error, "Unable to create a simple library" );
2728 error = clGetProgramBuildInfo (my_newly_minted_library, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2729 test_error( error, "Unable to get program binary type" );
2730 if (program_type != CL_PROGRAM_BINARY_TYPE_LIBRARY)
2731 {
2732 log_error( "ERROR: Expected program type of a just linked library to be CL_PROGRAM_BINARY_TYPE_LIBRARY (in %s:%d)\n", __FILE__, __LINE__ );
2733 return -1;
2734 }
2735 program_type = -1;
2736
2737 // Get the size of the resulting library (only one device)
2738 error = clGetProgramInfo( my_newly_minted_library, CL_PROGRAM_BINARY_SIZES, sizeof( binarySize ), &binarySize, NULL );
2739 test_error( error, "Unable to get binary size" );
2740
2741 // Sanity check
2742 if( binarySize == 0 )
2743 {
2744 log_error( "ERROR: Binary size of program is zero (in %s:%d)\n", __FILE__, __LINE__ );
2745 return -1;
2746 }
2747
2748 // Create a buffer and get the actual binary
2749 binary = (unsigned char*)malloc(sizeof(unsigned char)*binarySize);
2750 if (binary == NULL) {
2751 log_error( "ERROR: Unable to allocate binary character array with %lu characters! (in %s:%d)\n", binarySize, __FILE__, __LINE__);
2752 return -1;
2753 }
2754
2755 unsigned char *buffers[ 1 ] = { binary };
2756 cl_int loadErrors[ 1 ];
2757
2758 // Do another sanity check here first
2759 error = clGetProgramInfo( my_newly_minted_library, CL_PROGRAM_BINARIES, 0, NULL, &size );
2760 test_error( error, "Unable to get expected size of binaries array" );
2761 if( size != sizeof( buffers ) )
2762 {
2763 log_error( "ERROR: Expected size of binaries array in clGetProgramInfo is incorrect (should be %d, got %d) (in %s:%d)\n", (int)sizeof( buffers ), (int)size, __FILE__, __LINE__ );
2764 free(binary);
2765 return -1;
2766 }
2767
2768 error = clGetProgramInfo( my_newly_minted_library, CL_PROGRAM_BINARIES, sizeof( buffers ), &buffers, NULL );
2769 test_error( error, "Unable to get program binary" );
2770
2771 // use clCreateProgramWithBinary
2772 cl_program library_with_binary = clCreateProgramWithBinary(context, 1, &deviceID, &binarySize, (const unsigned char**)buffers, loadErrors, &error);
2773 test_error( error, "Unable to create program with binary" );
2774 error = clGetProgramBuildInfo (library_with_binary, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2775 test_error( error, "Unable to get program binary type" );
2776 if (program_type != CL_PROGRAM_BINARY_TYPE_LIBRARY)
2777 {
2778 log_error( "ERROR: Expected program type of a library loaded with binary to be CL_PROGRAM_BINARY_TYPE_LIBRARY (in %s:%d)\n", __FILE__, __LINE__ );
2779 return -1;
2780 }
2781 program_type = -1;
2782 free(binary);
2783
2784 error = create_single_kernel_helper_create_program(context, &another_program, 1, &another_simple_kernel);
2785 if( another_program == NULL || error != CL_SUCCESS )
2786 {
2787 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2788 return -1;
2789 }
2790
2791 error = clCompileProgram(another_program, 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
2792 test_error( error, "Unable to compile a simple program" );
2793
2794 cl_program program_and_archive[2] = { another_program, library_with_binary };
2795 cl_program fully_linked_program = clLinkProgram(context, 1, &deviceID, "", 2, program_and_archive, NULL, NULL, &error);
2796 test_error( error, "Unable to create an executable from a binary and a library" );
2797
2798 error = clGetProgramBuildInfo (fully_linked_program, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2799 test_error( error, "Unable to get program binary type" );
2800 if (program_type != CL_PROGRAM_BINARY_TYPE_EXECUTABLE)
2801 {
2802 log_error( "ERROR: Expected program type of a newly build executable to be CL_PROGRAM_BINARY_TYPE_EXECUTABLE (in %s:%d)\n", __FILE__, __LINE__ );
2803 return -1;
2804 }
2805 program_type = -1;
2806
2807 // Get the size of the resulting binary (only one device)
2808 error = clGetProgramInfo( fully_linked_program, CL_PROGRAM_BINARY_SIZES, sizeof( binarySize ), &binarySize, NULL );
2809 test_error( error, "Unable to get binary size" );
2810
2811 // Sanity check
2812 if( binarySize == 0 )
2813 {
2814 log_error( "ERROR: Binary size of program is zero (in %s:%d)\n", __FILE__, __LINE__ );
2815 return -1;
2816 }
2817
2818 // Create a buffer and get the actual binary
2819 {
2820 binary = (unsigned char*)malloc(sizeof(unsigned char)*binarySize);
2821 if (binary == NULL) {
2822 log_error( "ERROR: Unable to allocate binary character array with %lu characters! (in %s:%d)\n", binarySize, __FILE__, __LINE__ );
2823 return -1;
2824 }
2825 unsigned char *buffers[ 1 ] = { binary };
2826 cl_int loadErrors[ 1 ];
2827
2828 // Do another sanity check here first
2829 size_t size;
2830 error = clGetProgramInfo( fully_linked_program, CL_PROGRAM_BINARIES, 0, NULL, &size );
2831 test_error( error, "Unable to get expected size of binaries array" );
2832 if( size != sizeof( buffers ) )
2833 {
2834 log_error( "ERROR: Expected size of binaries array in clGetProgramInfo is incorrect (should be %d, got %d) (in %s:%d)\n", (int)sizeof( buffers ), (int)size, __FILE__, __LINE__ );
2835 free(binary);
2836 return -1;
2837 }
2838
2839 error = clGetProgramInfo( fully_linked_program, CL_PROGRAM_BINARIES, sizeof( buffers ), &buffers, NULL );
2840 test_error( error, "Unable to get program binary" );
2841
2842 // use clCreateProgramWithBinary
2843 fully_linked_program_with_binary = clCreateProgramWithBinary(context, 1, &deviceID, &binarySize, (const unsigned char**)buffers, loadErrors, &error);
2844 test_error( error, "Unable to create program with binary" );
2845
2846 error = clGetProgramBuildInfo (fully_linked_program_with_binary, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2847 test_error( error, "Unable to get program binary type" );
2848 if (program_type != CL_PROGRAM_BINARY_TYPE_EXECUTABLE)
2849 {
2850 log_error( "ERROR: Expected program type of a program created from a fully linked executable binary to be CL_PROGRAM_BINARY_TYPE_EXECUTABLE (in %s:%d)\n", __FILE__, __LINE__ );
2851 return -1;
2852 }
2853 program_type = -1;
2854 free(binary);
2855 }
2856
2857 error = clBuildProgram(fully_linked_program_with_binary, 1, &deviceID, NULL, NULL, NULL);
2858 test_error( error, "Unable to build a simple program" );
2859
2860 cl_kernel kernel = clCreateKernel(fully_linked_program_with_binary, "CopyBuffer", &error);
2861 test_error( error, "Unable to create a simple kernel" );
2862
2863 error = verifyCopyBuffer(context, queue, kernel);
2864 if (error != CL_SUCCESS)
2865 return error;
2866
2867 cl_kernel another_kernel = clCreateKernel(fully_linked_program_with_binary, "AnotherCopyBuffer", &error);
2868 test_error( error, "Unable to create another simple kernel" );
2869
2870 error = verifyCopyBuffer(context, queue, another_kernel);
2871 if (error != CL_SUCCESS)
2872 return error;
2873
2874 /* All done! */
2875 error = clReleaseKernel( kernel );
2876 test_error( error, "Unable to release kernel object" );
2877
2878 error = clReleaseKernel( another_kernel );
2879 test_error( error, "Unable to release another kernel object" );
2880
2881 error = clReleaseProgram( program );
2882 test_error( error, "Unable to release program object" );
2883
2884 /* Oh, one more thing. Steve Jobs and apparently Herb Sutter. The question is "Who is copying whom?" */
2885 error = create_single_kernel_helper_create_program(context, &program, 1, &simple_kernel);
2886 if( program == NULL || error != CL_SUCCESS )
2887 {
2888 log_error( "ERROR: Unable to create a simple test program! (%s in %s:%d)\n", IGetErrorString( error ), __FILE__, __LINE__ );
2889 return -1;
2890 }
2891
2892 error = clBuildProgram(program, 1, &deviceID, NULL, NULL, NULL);
2893 test_error( error, "Unable to build a simple program" );
2894 error = clGetProgramBuildInfo (program, deviceID, CL_PROGRAM_BINARY_TYPE, sizeof(cl_program_binary_type), &program_type, NULL);
2895 test_error( error, "Unable to get program binary type" );
2896 if (program_type != CL_PROGRAM_BINARY_TYPE_EXECUTABLE)
2897 {
2898 log_error( "ERROR: Expected program type of a program created from compiled object to be CL_PROGRAM_BINARY_TYPE_EXECUTABLE (in %s:%d)\n", __FILE__, __LINE__ );
2899 return -1;
2900 }
2901 program_type = -1;
2902
2903 /* All's well that ends well. William Shakespeare */
2904 error = clReleaseProgram( program );
2905 test_error( error, "Unable to release program object" );
2906
2907 error = clReleaseProgram( another_program );
2908 test_error( error, "Unable to release program object" );
2909
2910 error = clReleaseProgram( my_newly_minted_library );
2911 test_error( error, "Unable to release program object" );
2912
2913 error = clReleaseProgram( library_with_binary );
2914 test_error( error, "Unable to release program object" );
2915
2916 error = clReleaseProgram( fully_linked_program );
2917 test_error( error, "Unable to release program object" );
2918
2919 error = clReleaseProgram( fully_linked_program_with_binary );
2920 test_error( error, "Unable to release program object" );
2921
2922 error = clReleaseProgram( program_with_binary );
2923 test_error( error, "Unable to release program object" );
2924
2925 return 0;
2926 }
2927
2928 volatile int compileNotificationSent;
2929
test_notify_compile_complete(cl_program program,void * userData)2930 void CL_CALLBACK test_notify_compile_complete( cl_program program, void *userData )
2931 {
2932 if( userData == NULL || strcmp( (char *)userData, "compilation" ) != 0 )
2933 {
2934 log_error( "ERROR: User data passed in to compile notify function was not correct! (in %s:%d)\n", __FILE__, __LINE__ );
2935 compileNotificationSent = -1;
2936 }
2937 else
2938 compileNotificationSent = 1;
2939 log_info( "\n <-- program successfully compiled\n" );
2940 }
2941
2942 volatile int libraryCreationNotificationSent;
2943
test_notify_create_library_complete(cl_program program,void * userData)2944 void CL_CALLBACK test_notify_create_library_complete( cl_program program, void *userData )
2945 {
2946 if( userData == NULL || strcmp( (char *)userData, "create library" ) != 0 )
2947 {
2948 log_error( "ERROR: User data passed in to library creation notify function was not correct! (in %s:%d)\n", __FILE__, __LINE__ );
2949 libraryCreationNotificationSent = -1;
2950 }
2951 else
2952 libraryCreationNotificationSent = 1;
2953 log_info( "\n <-- library successfully created\n" );
2954 }
2955
2956 volatile int linkNotificationSent;
2957
test_notify_link_complete(cl_program program,void * userData)2958 void CL_CALLBACK test_notify_link_complete( cl_program program, void *userData )
2959 {
2960 if( userData == NULL || strcmp( (char *)userData, "linking" ) != 0 )
2961 {
2962 log_error( "ERROR: User data passed in to link notify function was not correct! (in %s:%d)\n", __FILE__, __LINE__ );
2963 linkNotificationSent = -1;
2964 }
2965 else
2966 linkNotificationSent = 1;
2967 log_info( "\n <-- program successfully linked\n" );
2968 }
2969
test_large_compile_and_link_status_options_log(cl_context context,cl_device_id deviceID,cl_command_queue queue,unsigned int numLines)2970 int test_large_compile_and_link_status_options_log(cl_context context, cl_device_id deviceID, cl_command_queue queue, unsigned int numLines)
2971 {
2972 int error;
2973 cl_program program;
2974 cl_program * simple_kernels;
2975 const char **lines;
2976 unsigned int i;
2977 char buffer[MAX_LINE_SIZE_IN_PROGRAM];
2978 char *compile_log;
2979 char *compile_options;
2980 char *library_log;
2981 char *library_options;
2982 char *linking_log;
2983 char *linking_options;
2984 cl_build_status status;
2985 size_t size_ret;
2986
2987 compileNotificationSent = libraryCreationNotificationSent = linkNotificationSent = 0;
2988
2989 simple_kernels = (cl_program*)malloc(numLines*sizeof(cl_program));
2990 if (simple_kernels == NULL) {
2991 log_error( "ERROR: Unable to allocate kernels array with %d kernels! (in %s:%d)\n", numLines, __FILE__, __LINE__);
2992 return -1;
2993 }
2994 /* First, allocate the array for our line pointers */
2995 lines = (const char **)malloc( (2*numLines + 2) * sizeof( const char * ) );
2996 if (lines == NULL) {
2997 log_error( "ERROR: Unable to allocate lines array with %d lines! (in %s:%d)\n", (2*numLines + 2), __FILE__, __LINE__);
2998 return -1;
2999 }
3000
3001 for(i = 0; i < numLines; i++)
3002 {
3003 sprintf(buffer, composite_kernel_extern_template, i);
3004 lines[i] = _strdup(buffer);
3005 }
3006 /* First and last lines are easy */
3007 lines[ numLines ] = composite_kernel_start;
3008 lines[ 2*numLines + 1] = composite_kernel_end;
3009
3010 /* Fill the rest with templated kernels */
3011 for(i = numLines + 1; i < 2*numLines + 1; i++ )
3012 {
3013 sprintf(buffer, composite_kernel_template, i - numLines - 1);
3014 lines[ i ] = _strdup(buffer);
3015 }
3016
3017 /* Try to create a program with these lines */
3018 error = create_single_kernel_helper_create_program(context, &program, 2 * numLines + 2, lines);
3019 if( program == NULL || error != CL_SUCCESS )
3020 {
3021 log_error( "ERROR: Unable to create long test program with %d lines! (%s) (in %s:%d)\n", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
3022 return -1;
3023 }
3024
3025 /* Lets check that the compilation status is CL_BUILD_NONE */
3026 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3027 test_error( error, "Unable to get program compile status" );
3028 if (status != CL_BUILD_NONE)
3029 {
3030 log_error( "ERROR: Expected compile status to be CL_BUILD_NONE prior to the beginning of the compilation! (status: %d in %s:%d)\n", (int)status, __FILE__, __LINE__ );
3031 return -1;
3032 }
3033
3034 /* Compile it */
3035 error = clCompileProgram(program, 1, &deviceID, NULL, 0, NULL, NULL, test_notify_compile_complete, (void *)"compilation");
3036 test_error( error, "Unable to compile a simple program" );
3037
3038 /* Wait for compile to complete (just keep polling, since we're just a test */
3039 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3040 test_error( error, "Unable to get program compile status" );
3041
3042 while( (int)status == CL_BUILD_IN_PROGRESS )
3043 {
3044 log_info( "\n -- still waiting for compile... (status is %d)", status );
3045 sleep( 1 );
3046 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3047 test_error( error, "Unable to get program compile status" );
3048 }
3049 if( status != CL_BUILD_SUCCESS )
3050 {
3051 log_error( "ERROR: compile failed! (status: %d in %s:%d)\n", (int)status, __FILE__, __LINE__ );
3052 return -1;
3053 }
3054
3055 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &size_ret );
3056 test_error( error, "Device failed to return compile log size" );
3057 compile_log = (char *)malloc(size_ret);
3058 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_LOG, size_ret, compile_log, NULL );
3059 if (error != CL_SUCCESS){
3060 log_error("Device failed to return a compile log (in %s:%d)\n", __FILE__, __LINE__);
3061 test_error(error, "clGetProgramBuildInfo CL_PROGRAM_BUILD_LOG failed");
3062 }
3063 log_info("BUILD LOG: %s\n", compile_log);
3064 free(compile_log);
3065
3066 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &size_ret );
3067 test_error(error, "Device failed to return compile options size");
3068 compile_options = (char *)malloc(size_ret);
3069 error = clGetProgramBuildInfo( program, deviceID, CL_PROGRAM_BUILD_OPTIONS, size_ret, compile_options, NULL );
3070 test_error(error, "Device failed to return compile options.\nclGetProgramBuildInfo CL_PROGRAM_BUILD_OPTIONS failed");
3071
3072 log_info("BUILD OPTIONS: %s\n", compile_options);
3073 free(compile_options);
3074
3075 /* Create and compile templated kernels */
3076 for( i = 0; i < numLines; i++)
3077 {
3078 sprintf(buffer, simple_kernel_template, i);
3079 const char* kernel_source = _strdup(buffer);
3080 error = create_single_kernel_helper_create_program(context, &simple_kernels[i], 1, &kernel_source);
3081 if( simple_kernels[i] == NULL || error != CL_SUCCESS )
3082 {
3083 log_error( "ERROR: Unable to create long test program with %d lines! (%s in %s:%d)", numLines, IGetErrorString( error ), __FILE__, __LINE__ );
3084 return -1;
3085 }
3086
3087 /* Compile it */
3088 error = clCompileProgram(simple_kernels[i], 1, &deviceID, NULL, 0, NULL, NULL, NULL, NULL);
3089 test_error( error, "Unable to compile a simple program" );
3090
3091 free((void*)kernel_source);
3092 }
3093
3094 /* Create library out of compiled templated kernels */
3095 cl_program my_newly_minted_library = clLinkProgram(context, 1, &deviceID, "-create-library", numLines, simple_kernels, test_notify_create_library_complete, (void *)"create library", &error);
3096 test_error( error, "Unable to create a multi-line library" );
3097
3098 /* Wait for library creation to complete (just keep polling, since we're just a test */
3099 error = clGetProgramBuildInfo( my_newly_minted_library, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3100 test_error( error, "Unable to get library creation link status" );
3101
3102 while( (int)status == CL_BUILD_IN_PROGRESS )
3103 {
3104 log_info( "\n -- still waiting for library creation... (status is %d)", status );
3105 sleep( 1 );
3106 error = clGetProgramBuildInfo( my_newly_minted_library, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3107 test_error( error, "Unable to get library creation link status" );
3108 }
3109 if( status != CL_BUILD_SUCCESS )
3110 {
3111 log_error( "ERROR: library creation failed! (status: %d in %s:%d)\n", (int)status, __FILE__, __LINE__ );
3112 return -1;
3113 }
3114 error = clGetProgramBuildInfo( my_newly_minted_library, deviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &size_ret );
3115 test_error( error, "Device failed to return a library creation log size" );
3116 library_log = (char *)malloc(size_ret);
3117 error = clGetProgramBuildInfo( my_newly_minted_library, deviceID, CL_PROGRAM_BUILD_LOG, size_ret, library_log, NULL );
3118 if (error != CL_SUCCESS) {
3119 log_error("Device failed to return a library creation log (in %s:%d)\n", __FILE__, __LINE__);
3120 test_error(error, "clGetProgramBuildInfo CL_PROGRAM_BUILD_LOG failed");
3121 }
3122 log_info("CREATE LIBRARY LOG: %s\n", library_log);
3123 free(library_log);
3124
3125 error = clGetProgramBuildInfo( my_newly_minted_library, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &size_ret );
3126 test_error(error, "Device failed to return library creation options size");
3127 library_options = (char *)malloc(size_ret);
3128 error = clGetProgramBuildInfo( my_newly_minted_library, deviceID, CL_PROGRAM_BUILD_OPTIONS, size_ret, library_options, NULL );
3129 test_error(error, "Device failed to return library creation options.\nclGetProgramBuildInfo CL_PROGRAM_BUILD_OPTIONS failed");
3130
3131 log_info("CREATE LIBRARY OPTIONS: %s\n", library_options);
3132 free(library_options);
3133
3134 /* Link the program that calls the kernels and the library that contains them */
3135 cl_program programs[2] = { program, my_newly_minted_library };
3136 cl_program my_newly_linked_program = clLinkProgram(context, 1, &deviceID, NULL, 2, programs, test_notify_link_complete, (void *)"linking", &error);
3137 test_error( error, "Unable to link a program with a library" );
3138
3139 /* Wait for linking to complete (just keep polling, since we're just a test */
3140 error = clGetProgramBuildInfo( my_newly_linked_program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3141 test_error( error, "Unable to get program link status" );
3142
3143 while( (int)status == CL_BUILD_IN_PROGRESS )
3144 {
3145 log_info( "\n -- still waiting for program linking... (status is %d)", status );
3146 sleep( 1 );
3147 error = clGetProgramBuildInfo( my_newly_linked_program, deviceID, CL_PROGRAM_BUILD_STATUS, sizeof( status ), &status, NULL );
3148 test_error( error, "Unable to get program link status" );
3149 }
3150 if( status != CL_BUILD_SUCCESS )
3151 {
3152 log_error( "ERROR: program linking failed! (status: %d in %s:%d)\n", (int)status, __FILE__, __LINE__ );
3153 return -1;
3154 }
3155 error = clGetProgramBuildInfo( my_newly_linked_program, deviceID, CL_PROGRAM_BUILD_LOG, 0, NULL, &size_ret );
3156 test_error( error, "Device failed to return a linking log size" );
3157 linking_log = (char *)malloc(size_ret);
3158 error = clGetProgramBuildInfo( my_newly_linked_program, deviceID, CL_PROGRAM_BUILD_LOG, size_ret, linking_log, NULL );
3159 if (error != CL_SUCCESS){
3160 log_error("Device failed to return a linking log (in %s:%d).\n", __FILE__, __LINE__);
3161 test_error(error, "clGetProgramBuildInfo CL_PROGRAM_BUILD_LOG failed");
3162 }
3163 log_info("BUILDING LOG: %s\n", linking_log);
3164 free(linking_log);
3165
3166 error = clGetProgramBuildInfo( my_newly_linked_program, deviceID, CL_PROGRAM_BUILD_OPTIONS, 0, NULL, &size_ret );
3167 test_error(error, "Device failed to return linking options size");
3168 linking_options = (char *)malloc(size_ret);
3169 error = clGetProgramBuildInfo( my_newly_linked_program, deviceID, CL_PROGRAM_BUILD_OPTIONS, size_ret, linking_options, NULL );
3170 test_error(error, "Device failed to return linking options.\nclGetProgramBuildInfo CL_PROGRAM_BUILD_OPTIONS failed");
3171
3172 log_info("BUILDING OPTIONS: %s\n", linking_options);
3173 free(linking_options);
3174
3175 // Create the composite kernel
3176 cl_kernel kernel = clCreateKernel(my_newly_linked_program, "CompositeKernel", &error);
3177 test_error( error, "Unable to create a composite kernel" );
3178
3179 // Run the composite kernel and verify the results
3180 error = verifyCopyBuffer(context, queue, kernel);
3181 if (error != CL_SUCCESS)
3182 return error;
3183
3184 /* All done! */
3185 error = clReleaseKernel( kernel );
3186 test_error( error, "Unable to release kernel object" );
3187
3188 error = clReleaseProgram( program );
3189 test_error( error, "Unable to release program object" );
3190
3191 for(i = 0; i < numLines; i++)
3192 {
3193 free( (void*)lines[i] );
3194 free( (void*)lines[i+numLines+1] );
3195 }
3196 free( lines );
3197
3198 for(i = 0; i < numLines; i++)
3199 {
3200 error = clReleaseProgram( simple_kernels[i] );
3201 test_error( error, "Unable to release program object" );
3202 }
3203 free( simple_kernels );
3204
3205 error = clReleaseProgram( my_newly_minted_library );
3206 test_error( error, "Unable to release program object" );
3207
3208 error = clReleaseProgram( my_newly_linked_program );
3209 test_error( error, "Unable to release program object" );
3210
3211 return 0;
3212 }
3213
test_compile_and_link_status_options_log(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)3214 int test_compile_and_link_status_options_log(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
3215 {
3216 unsigned int toTest[] = { 256, 0 }; //512, 1024, 8192, 16384, 32768, 0 };
3217 unsigned int i;
3218
3219 log_info( "Testing Compile and Link Status, Options and Logging ...this might take awhile...\n" );
3220
3221 for( i = 0; toTest[ i ] != 0; i++ )
3222 {
3223 log_info( " %d...\n", toTest[ i ] );
3224
3225 #if defined(_WIN32)
3226 clock_t start = clock();
3227 #elif defined(__linux__) || defined(__APPLE__)
3228 timeval time1, time2;
3229 gettimeofday(&time1, NULL);
3230 #endif
3231
3232 if( test_large_compile_and_link_status_options_log( context, deviceID, queue, toTest[ i ] ) != 0 )
3233 {
3234 log_error( "ERROR: large program compilation, linking, status, options and logging test failed for %d lines! (in %s:%d)\n", toTest[ i ], __FILE__, __LINE__ );
3235 return -1;
3236 }
3237
3238 #if defined(_WIN32)
3239 clock_t end = clock();
3240 log_perf( (float)( end - start ) / (float)CLOCKS_PER_SEC, false, "clock() time in secs", "%d lines", toTest[i] );
3241 #elif defined(__linux__) || defined(__APPLE__)
3242 gettimeofday(&time2, NULL);
3243 log_perf( (float)(float)(time2.tv_sec - time1.tv_sec) + 1.0e-6 * (time2.tv_usec - time1.tv_usec) , false, "wall time in secs", "%d lines", toTest[i] );
3244 #endif
3245 }
3246
3247 return 0;
3248 }
3249