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