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
18 #if ! defined( _WIN32 )
19 #include "unistd.h" // for "sleep" used in the "while (1)" busy wait loop in
20 #endif
21 // test_event_flush
22
23 const char *sample_long_test_kernel[] = {
24 "__kernel void sample_test(__global float *src, __global int *dst)\n"
25 "{\n"
26 " int tid = get_global_id(0);\n"
27 " int i;\n"
28 "\n"
29 " for( i = 0; i < 10000; i++ )\n"
30 " {\n"
31 " dst[tid] = (int)src[tid] * 3;\n"
32 " }\n"
33 "\n"
34 "}\n" };
35
create_and_execute_kernel(cl_context inContext,cl_command_queue inQueue,cl_program * outProgram,cl_kernel * outKernel,cl_mem * streams,unsigned int lineCount,const char ** lines,const char * kernelName,cl_event * outEvent)36 int create_and_execute_kernel( cl_context inContext, cl_command_queue inQueue, cl_program *outProgram, cl_kernel *outKernel, cl_mem *streams,
37 unsigned int lineCount, const char **lines, const char *kernelName, cl_event *outEvent )
38 {
39 size_t threads[1] = { 1000 }, localThreads[1];
40 int error;
41
42 if( create_single_kernel_helper( inContext, outProgram, outKernel, lineCount, lines, kernelName ) )
43 {
44 return -1;
45 }
46
47 error = get_max_common_work_group_size( inContext, *outKernel, threads[0], &localThreads[0] );
48 test_error( error, "Unable to get work group size to use" );
49
50 streams[0] = clCreateBuffer(inContext, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1000, NULL, &error);
51 test_error( error, "Creating test array failed" );
52 streams[1] = clCreateBuffer(inContext, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_int) * 1000, NULL, &error);
53 test_error( error, "Creating test array failed" );
54
55 /* Set the arguments */
56 error = clSetKernelArg( *outKernel, 0, sizeof( streams[0] ), &streams[0] );
57 test_error( error, "Unable to set kernel arguments" );
58 error = clSetKernelArg( *outKernel, 1, sizeof( streams[1] ), &streams[1] );
59 test_error( error, "Unable to set kernel arguments" );
60
61 error = clEnqueueNDRangeKernel(inQueue, *outKernel, 1, NULL, threads, localThreads, 0, NULL, outEvent);
62 test_error( error, "Unable to execute test kernel" );
63
64 return 0;
65 }
66
67 #define SETUP_EVENT( c, q ) \
68 clProgramWrapper program; \
69 clKernelWrapper kernel; \
70 clMemWrapper streams[2]; \
71 clEventWrapper event; \
72 int error; \
73 if( create_and_execute_kernel( c, q, &program, &kernel, &streams[0], 1, sample_long_test_kernel, "sample_test", &event ) ) return -1;
74
75 #define FINISH_EVENT(_q) clFinish(_q)
76
IGetStatusString(cl_int status)77 const char *IGetStatusString( cl_int status )
78 {
79 static char tempString[ 128 ];
80 switch( status )
81 {
82 case CL_COMPLETE: return "CL_COMPLETE";
83 case CL_RUNNING: return "CL_RUNNING";
84 case CL_QUEUED: return "CL_QUEUED";
85 case CL_SUBMITTED: return "CL_SUBMITTED";
86 default:
87 sprintf( tempString, "<unknown: %d>", (int)status );
88 return tempString;
89 }
90 }
91
92 /* Note: tests clGetEventStatus and clReleaseEvent (implicitly) */
test_event_get_execute_status(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)93 int test_event_get_execute_status( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
94 {
95 cl_int status;
96 SETUP_EVENT( context, queue );
97
98 /* Now wait for it to be done */
99 error = clWaitForEvents( 1, &event );
100 test_error( error, "Unable to wait for event" );
101
102 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
103 test_error( error, "Calling clGetEventStatus to wait for event completion failed" );
104 if( status != CL_COMPLETE )
105 {
106 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after event complete (%d:%s)\n", status, IGetStatusString( status ) );
107 return -1;
108 }
109
110 FINISH_EVENT(queue);
111 return 0;
112 }
113
test_event_get_info(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)114 int test_event_get_info( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
115 {
116 SETUP_EVENT( context, queue );
117
118 /* Verify parameters of clGetEventInfo not already tested by other tests */
119 cl_command_queue otherQueue;
120 size_t size;
121
122 error = clGetEventInfo( event, CL_EVENT_COMMAND_QUEUE, sizeof( otherQueue ), &otherQueue, &size );
123 test_error( error, "Unable to get event info!" );
124 // We can not check if this is the right queue because this is an opaque object.
125 if( size != sizeof( queue ) )
126 {
127 log_error( "ERROR: Returned command queue size does not validate (expected %d, got %d)\n", (int)sizeof( queue ), (int)size );
128 return -1;
129 }
130
131 cl_command_type type;
132 error = clGetEventInfo( event, CL_EVENT_COMMAND_TYPE, sizeof( type ), &type, &size );
133 test_error( error, "Unable to get event info!" );
134 if( type != CL_COMMAND_NDRANGE_KERNEL )
135 {
136 log_error( "ERROR: Returned command type does not validate (expected %d, got %d)\n", (int)CL_COMMAND_NDRANGE_KERNEL, (int)type );
137 return -1;
138 }
139 if( size != sizeof( type ) )
140 {
141 log_error( "ERROR: Returned command type size does not validate (expected %d, got %d)\n", (int)sizeof( type ), (int)size );
142 return -1;
143 }
144
145 cl_uint count;
146 error = clGetEventInfo( event, CL_EVENT_REFERENCE_COUNT, sizeof( count ), &count, &size );
147 test_error( error, "Unable to get event info for CL_EVENT_REFERENCE_COUNT!" );
148 if( size != sizeof( count ) )
149 {
150 log_error( "ERROR: Returned command type size does not validate (expected %d, got %d)\n", (int)sizeof( type ), (int)size );
151 return -1;
152 }
153
154 cl_context testCtx;
155 error = clGetEventInfo( event, CL_EVENT_CONTEXT, sizeof( testCtx ), &testCtx, &size );
156 test_error( error, "Unable to get event context info!" );
157 if( size != sizeof( context ) )
158 {
159 log_error( "ERROR: Returned context size does not validate (expected %d, got %d)\n", (int)sizeof( context ), (int)size );
160 return -1;
161 }
162 if( testCtx != context )
163 {
164 log_error( "ERROR: Returned context does not match (expected %p, got %p)\n", (void *)context, (void *)testCtx );
165 return -1;
166 }
167
168 FINISH_EVENT(queue);
169 return 0;
170 }
171
test_event_get_write_array_status(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)172 int test_event_get_write_array_status( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
173 {
174 cl_mem stream;
175 cl_float testArray[ 1024 * 32 ];
176 cl_event event;
177 int error;
178 cl_int status;
179
180
181 stream = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1024 * 32, NULL, &error );
182 test_error( error, "Creating test array failed" );
183
184 error = clEnqueueWriteBuffer(queue, stream, CL_FALSE, 0, sizeof(cl_float)*1024*32, (void *)testArray, 0, NULL, &event);
185 test_error( error, "Unable to set testing kernel data" );
186
187 /* Now wait for it to be done */
188 error = clWaitForEvents( 1, &event );
189 test_error( error, "Unable to wait for event" );
190
191 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
192 test_error( error, "Calling clGetEventStatus to wait for event completion failed" );
193 if( status != CL_COMPLETE )
194 {
195 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after array write complete (%d:%s)\n", status, IGetStatusString( status ) );
196 return -1;
197 }
198
199
200 clReleaseMemObject( stream );
201 clReleaseEvent( event );
202
203 return 0;
204 }
205
test_event_get_read_array_status(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)206 int test_event_get_read_array_status( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
207 {
208 cl_mem stream;
209 cl_float testArray[ 1024 * 32 ];
210 cl_event event;
211 int error;
212 cl_int status;
213
214
215 stream = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1024 * 32, NULL, &error );
216 test_error( error, "Creating test array failed" );
217
218 error = clEnqueueReadBuffer(queue, stream, CL_FALSE, 0, sizeof(cl_float)*1024*32, (void *)testArray, 0, NULL, &event);
219 test_error( error, "Unable to get testing kernel data" );
220
221
222 /* It should still be running... */
223 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
224 test_error( error, "Calling clGetEventStatus didn't work!" );
225
226 if( status != CL_RUNNING && status != CL_QUEUED && status != CL_SUBMITTED && status != CL_COMPLETE)
227 {
228 log_error( "ERROR: Incorrect status returned from clGetErrorStatus during array read (%d:%s)\n", status, IGetStatusString( status ) );
229 return -1;
230 }
231
232 /* Now wait for it to be done */
233 error = clWaitForEvents( 1, &event );
234 test_error( error, "Unable to wait for event" );
235
236 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
237 test_error( error, "Calling clGetEventStatus to wait for event completion failed" );
238 if( status != CL_COMPLETE )
239 {
240 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after array read complete (%d:%s)\n", status, IGetStatusString( status ) );
241 return -1;
242 }
243
244
245 clReleaseMemObject( stream );
246 clReleaseEvent( event );
247
248 return 0;
249 }
250
251 /* clGetEventStatus not implemented yet */
252
test_event_wait_for_execute(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)253 int test_event_wait_for_execute( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
254 {
255 cl_int status;
256 SETUP_EVENT( context, queue );
257
258 /* Now we wait for it to be done, then test the status again */
259 error = clWaitForEvents( 1, &event );
260 test_error( error, "Unable to wait for execute event" );
261
262 /* Make sure it worked */
263 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
264 test_error( error, "Calling clGetEventStatus didn't work!" );
265 if( status != CL_COMPLETE )
266 {
267 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after event complete (%d:%s)\n", status, IGetStatusString( status ) );
268 return -1;
269 }
270
271 FINISH_EVENT(queue);
272 return 0;
273 }
274
test_event_wait_for_array(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)275 int test_event_wait_for_array( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
276 {
277 cl_mem streams[2];
278 cl_float readArray[ 1024 * 32 ];
279 cl_float writeArray[ 1024 * 32 ];
280 cl_event events[2];
281 int error;
282 cl_int status;
283
284
285 streams[0] = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1024 * 32, NULL, &error );
286 test_error( error, "Creating test array failed" );
287 streams[1] = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1024 * 32, NULL, &error );
288 test_error( error, "Creating test array failed" );
289
290 error = clEnqueueReadBuffer(queue, streams[0], CL_FALSE, 0, sizeof(cl_float)*1024*32, (void *)readArray, 0, NULL, &events[0]);
291 test_error( error, "Unable to read testing kernel data" );
292
293 error = clEnqueueWriteBuffer(queue, streams[1], CL_FALSE, 0, sizeof(cl_float)*1024*32, (void *)writeArray, 0, NULL, &events[1]);
294 test_error( error, "Unable to write testing kernel data" );
295
296 /* Both should still be running */
297 error = clGetEventInfo( events[0], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
298 test_error( error, "Calling clGetEventStatus didn't work!" );
299 if( status != CL_RUNNING && status != CL_QUEUED && status != CL_SUBMITTED && status != CL_COMPLETE)
300 {
301 log_error( "ERROR: Incorrect status returned from clGetErrorStatus during array read (%d:%s)\n", status, IGetStatusString( status ) );
302 return -1;
303 }
304
305 error = clGetEventInfo( events[1], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
306 test_error( error, "Calling clGetEventStatus didn't work!" );
307 if( status != CL_RUNNING && status != CL_QUEUED && status != CL_SUBMITTED && status != CL_COMPLETE)
308 {
309 log_error( "ERROR: Incorrect status returned from clGetErrorStatus during array write (%d:%s)\n", status, IGetStatusString( status ) );
310 return -1;
311 }
312
313 /* Now try waiting for both */
314 error = clWaitForEvents( 2, events );
315 test_error( error, "Unable to wait for array events" );
316
317 /* Double check status on both */
318 error = clGetEventInfo( events[0], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
319 test_error( error, "Calling clGetEventStatus didn't work!" );
320 if( status != CL_COMPLETE )
321 {
322 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after array read complete (%d:%s)\n", status, IGetStatusString( status ) );
323 return -1;
324 }
325
326 error = clGetEventInfo( events[1], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
327 test_error( error, "Calling clGetEventStatus didn't work!" );
328 if( status != CL_COMPLETE )
329 {
330 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after array write complete (%d:%s)\n", status, IGetStatusString( status ) );
331 return -1;
332 }
333
334 clReleaseMemObject( streams[0] );
335 clReleaseMemObject( streams[1] );
336 clReleaseEvent( events[0] );
337 clReleaseEvent( events[1] );
338
339 return 0;
340 }
341
test_event_flush(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)342 int test_event_flush( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
343 {
344 int loopCount = 0;
345 cl_int status;
346 SETUP_EVENT( context, queue );
347
348 /* Now flush. Note that we can't guarantee this actually lets the op finish, but we can guarantee it's no longer queued */
349 error = clFlush( queue );
350 test_error( error, "Unable to flush events" );
351
352 /* Make sure it worked */
353 while (1) {
354 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS,
355 sizeof( status ), &status, NULL );
356 test_error( error, "Calling clGetEventStatus didn't work!" );
357
358 if( status != CL_QUEUED )
359 break;
360
361 #if ! defined( _WIN32 )
362 sleep(1); // give it some time here.
363 #else // _WIN32
364 Sleep(1000);
365 #endif
366 ++loopCount;
367 }
368
369 /*
370 CL_QUEUED (command has been enqueued in the command-queue),
371 CL_SUBMITTED (enqueued command has been submitted by the host to the device associated with the command-queue),
372 CL_RUNNING (device is currently executing this command),
373 CL_COMPLETE (the command has completed), or
374 Error code given by a negative integer value. (command was abnormally terminated – this may be caused by a bad memory access etc.).
375 */
376 if(status != CL_COMPLETE && status != CL_SUBMITTED &&
377 status != CL_RUNNING && status != CL_COMPLETE)
378 {
379 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after event flush (%d:%s)\n", status, IGetStatusString( status ) );
380 return -1;
381 }
382
383 /* Now wait */
384 error = clFinish( queue );
385 test_error( error, "Unable to finish events" );
386
387 FINISH_EVENT(queue);
388 return 0;
389 }
390
391
test_event_finish_execute(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)392 int test_event_finish_execute( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
393 {
394 cl_int status;
395 SETUP_EVENT( context, queue );
396
397 /* Now flush and finish all ops */
398 error = clFinish( queue );
399 test_error( error, "Unable to finish all events" );
400
401 /* Make sure it worked */
402 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
403 test_error( error, "Calling clGetEventStatus didn't work!" );
404 if( status != CL_COMPLETE )
405 {
406 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after event complete (%d:%s)\n", status, IGetStatusString( status ) );
407 return -1;
408 }
409
410 FINISH_EVENT(queue);
411 return 0;
412 }
413
test_event_finish_array(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)414 int test_event_finish_array( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
415 {
416 cl_mem streams[2];
417 cl_float readArray[ 1024 * 32 ];
418 cl_float writeArray[ 1024 * 32 ];
419 cl_event events[2];
420 int error;
421 cl_int status;
422
423
424 streams[0] = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1024 * 32, NULL, &error );
425 test_error( error, "Creating test array failed" );
426 streams[1] = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * 1024 * 32, NULL, &error );
427 test_error( error, "Creating test array failed" );
428
429 error = clEnqueueReadBuffer(queue, streams[0], CL_FALSE, 0, sizeof(cl_float)*1024*32, (void *)readArray, 0, NULL, &events[0]);
430 test_error( error, "Unable to read testing kernel data" );
431
432 error = clEnqueueWriteBuffer(queue, streams[1], CL_FALSE, 0, sizeof(cl_float)*1024*32, (void *)writeArray, 0, NULL, &events[1]);
433 test_error( error, "Unable to write testing kernel data" );
434
435 /* Both should still be running */
436 error = clGetEventInfo( events[0], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
437 test_error( error, "Calling clGetEventStatus didn't work!" );
438 if( status != CL_RUNNING && status != CL_QUEUED && status != CL_SUBMITTED && status != CL_COMPLETE)
439 {
440 log_error( "ERROR: Incorrect status returned from clGetErrorStatus during array read (%d:%s)\n", status, IGetStatusString( status ) );
441 return -1;
442 }
443
444 error = clGetEventInfo( events[1], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
445 test_error( error, "Calling clGetEventStatus didn't work!" );
446 if( status != CL_RUNNING && status != CL_QUEUED && status != CL_SUBMITTED && status != CL_COMPLETE)
447 {
448 log_error( "ERROR: Incorrect status returned from clGetErrorStatus during array write (%d:%s)\n", status, IGetStatusString( status ) );
449 return -1;
450 }
451
452 /* Now try finishing all ops */
453 error = clFinish( queue );
454 test_error( error, "Unable to finish all events" );
455
456 /* Double check status on both */
457 error = clGetEventInfo( events[0], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
458 test_error( error, "Calling clGetEventStatus didn't work!" );
459 if( status != CL_COMPLETE )
460 {
461 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after array read complete (%d:%s)\n", status, IGetStatusString( status ) );
462 return -1;
463 }
464
465 error = clGetEventInfo( events[1], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
466 test_error( error, "Calling clGetEventStatus didn't work!" );
467 if( status != CL_COMPLETE )
468 {
469 log_error( "ERROR: Incorrect status returned from clGetErrorStatus after array write complete (%d:%s)\n", status, IGetStatusString( status ) );
470 return -1;
471 }
472
473 clReleaseMemObject( streams[0] );
474 clReleaseMemObject( streams[1] );
475 clReleaseEvent( events[0] );
476 clReleaseEvent( events[1] );
477
478 return 0;
479 }
480
481
482 #define NUM_EVENT_RUNS 100
483
test_event_release_before_done(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)484 int test_event_release_before_done( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
485 {
486 // Create a kernel to run
487 clProgramWrapper program;
488 clKernelWrapper kernel[NUM_EVENT_RUNS];
489 size_t threads[1] = { 1000 };
490 cl_event events[NUM_EVENT_RUNS];
491 cl_int status;
492 clMemWrapper streams[NUM_EVENT_RUNS][2];
493 int error, i;
494
495 // Create a kernel
496 if( create_single_kernel_helper( context, &program, &kernel[0], 1, sample_long_test_kernel, "sample_test" ) )
497 {
498 return -1;
499 }
500
501 for( i = 1; i < NUM_EVENT_RUNS; i++ ) {
502 kernel[i] = clCreateKernel(program, "sample_test", &error);
503 test_error(error, "Unable to create kernel");
504 }
505
506 error = get_max_common_work_group_size( context, kernel[0], 1024, &threads[0] );
507 test_error( error, "Unable to get work group size to use" );
508
509 // Create a set of streams to use as arguments
510 for( i = 0; i < NUM_EVENT_RUNS; i++ )
511 {
512 streams[i][0] = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_float) * threads[0], NULL, &error );
513 streams[i][1] = clCreateBuffer( context, (cl_mem_flags)(CL_MEM_READ_WRITE), sizeof(cl_int) * threads[0], NULL, &error );
514 if( ( streams[i][0] == NULL ) || ( streams[i][1] == NULL ) )
515 {
516 log_error( "ERROR: Unable to allocate testing streams" );
517 return -1;
518 }
519 }
520
521 // Execute the kernels one by one, hopefully making sure they won't be done by the time we get to the end
522 for( i = 0; i < NUM_EVENT_RUNS; i++ )
523 {
524 error = clSetKernelArg( kernel[i], 0, sizeof( cl_mem ), &streams[i][0] );
525 error |= clSetKernelArg( kernel[i], 1, sizeof( cl_mem ), &streams[i][1] );
526 test_error( error, "Unable to set kernel arguments" );
527
528 error = clEnqueueNDRangeKernel( queue, kernel[i], 1, NULL, threads, threads, 0, NULL, &events[i]);
529 test_error( error, "Unable to execute test kernel" );
530 }
531
532 // Free all but the last event
533 for( i = 0; i < NUM_EVENT_RUNS - 1; i++ )
534 {
535 clReleaseEvent( events[ i ] );
536 }
537
538 // Get status on the last one, then free it
539 error = clGetEventInfo( events[ NUM_EVENT_RUNS - 1 ], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
540 test_error( error, "Unable to get event status" );
541
542 clReleaseEvent( events[ NUM_EVENT_RUNS - 1 ] );
543
544 // Was the status still-running?
545 if( status == CL_COMPLETE )
546 {
547 log_info( "WARNING: Events completed before they could be released, so test is a null-op. Increase workload and try again." );
548 }
549 else if( status == CL_RUNNING || status == CL_QUEUED || status == CL_SUBMITTED )
550 {
551 log_info( "Note: Event status was running or queued when released, so test was good.\n" );
552 }
553
554 // If we didn't crash by now, the test succeeded
555 clFinish( queue );
556
557 return 0;
558 }
559
test_event_enqueue_marker(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)560 int test_event_enqueue_marker( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
561 {
562 cl_int status;
563 SETUP_EVENT( context, queue );
564
565 /* Now we queue a marker and wait for that, which--since it queues afterwards--should guarantee the execute finishes too */
566 clEventWrapper markerEvent;
567 //error = clEnqueueMarker( queue, &markerEvent );
568
569 #ifdef CL_VERSION_1_2
570 error = clEnqueueMarkerWithWaitList(queue, 0, NULL, &markerEvent );
571 #else
572 error = clEnqueueMarker( queue, &markerEvent );
573 #endif
574 test_error( error, "Unable to queue marker" );
575 /* Now we wait for it to be done, then test the status again */
576 error = clWaitForEvents( 1, &markerEvent );
577 test_error( error, "Unable to wait for marker event" );
578
579 /* Check the status of the first event */
580 error = clGetEventInfo( event, CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof( status ), &status, NULL );
581 test_error( error, "Calling clGetEventInfo didn't work!" );
582 if( status != CL_COMPLETE )
583 {
584 log_error( "ERROR: Incorrect status returned from clGetEventInfo after event complete (%d:%s)\n", status, IGetStatusString( status ) );
585 return -1;
586 }
587
588 FINISH_EVENT(queue);
589 return 0;
590 }
591
592 #ifdef CL_VERSION_1_2
test_event_enqueue_marker_with_event_list(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)593 int test_event_enqueue_marker_with_event_list( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
594 {
595
596 cl_int status;
597 SETUP_EVENT( context, queue );
598 cl_event event_list[3]={ NULL, NULL, NULL};
599
600 size_t threads[1] = { 10 }, localThreads[1]={1};
601 cl_uint event_count=2;
602 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[0]);
603 test_error( error, " clEnqueueMarkerWithWaitList 1 " );
604
605 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[1]);
606 test_error( error, " clEnqueueMarkerWithWaitList 2" );
607
608 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, NULL);
609 test_error( error, " clEnqueueMarkerWithWaitList 3" );
610
611 // test the case event returned
612 error =clEnqueueMarkerWithWaitList(queue, event_count, event_list, &event_list[2]);
613 test_error( error, " clEnqueueMarkerWithWaitList " );
614
615 error = clReleaseEvent(event_list[0]);
616 error |= clReleaseEvent(event_list[1]);
617 test_error( error, "clReleaseEvent" );
618
619 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[0]);
620 test_error( error, " clEnqueueMarkerWithWaitList 1 -1 " );
621
622 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[1]);
623 test_error( error, " clEnqueueMarkerWithWaitList 2-2" );
624
625 // test the case event =NULL, caused [CL_INVALID_VALUE] : OpenCL Error : clEnqueueMarkerWithWaitList failed: event is a NULL value
626 error =clEnqueueMarkerWithWaitList(queue, event_count, event_list, NULL);
627 test_error( error, " clEnqueueMarkerWithWaitList " );
628
629 error = clReleaseEvent(event_list[0]);
630 error |= clReleaseEvent(event_list[1]);
631 error |= clReleaseEvent(event_list[2]);
632 test_error( error, "clReleaseEvent" );
633
634 FINISH_EVENT(queue);
635 return 0;
636 }
637
test_event_enqueue_barrier_with_event_list(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)638 int test_event_enqueue_barrier_with_event_list( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
639 {
640
641 cl_int status;
642 SETUP_EVENT( context, queue );
643 cl_event event_list[3]={ NULL, NULL, NULL};
644
645 size_t threads[1] = { 10 }, localThreads[1]={1};
646 cl_uint event_count=2;
647 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[0]);
648 test_error( error, " clEnqueueBarrierWithWaitList 1 " );
649
650 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[1]);
651 test_error( error, " clEnqueueBarrierWithWaitList 2" );
652
653 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, NULL);
654 test_error( error, " clEnqueueBarrierWithWaitList 20" );
655
656 // test the case event returned
657 error =clEnqueueBarrierWithWaitList(queue, event_count, event_list, &event_list[2]);
658 test_error( error, " clEnqueueBarrierWithWaitList " );
659
660 clReleaseEvent(event_list[0]);
661 clReleaseEvent(event_list[1]);
662
663 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[0]);
664 test_error( error, " clEnqueueBarrierWithWaitList 1 " );
665
666 error= clEnqueueNDRangeKernel( queue,kernel,1,NULL, threads, localThreads, 0, NULL, &event_list[1]);
667 test_error( error, " clEnqueueBarrierWithWaitList 2" );
668
669 // test the case event =NULL, caused [CL_INVALID_VALUE] : OpenCL Error : clEnqueueMarkerWithWaitList failed: event is a NULL value
670 error = clEnqueueBarrierWithWaitList(queue, event_count, event_list, NULL);
671 test_error( error, " clEnqueueBarrierWithWaitList " );
672
673 clReleaseEvent(event_list[0]);
674 clReleaseEvent(event_list[1]);
675 clReleaseEvent(event_list[2]);
676
677 FINISH_EVENT(queue);
678 return 0;
679 }
680 #endif
681