1 //
2 // Copyright (c) 2017 The Khronos Group Inc.
3 //
4 // Licensed under the Apache License, Version 2.0 (the "License");
5 // you may not use this file except in compliance with the License.
6 // You may obtain a copy of the License at
7 //
8 // http://www.apache.org/licenses/LICENSE-2.0
9 //
10 // Unless required by applicable law or agreed to in writing, software
11 // distributed under the License is distributed on an "AS IS" BASIS,
12 // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13 // See the License for the specific language governing permissions and
14 // limitations under the License.
15 //
16 #include "testBase.h"
17 #include "harness/testHarness.h"
18
19 const char *write_kernels[] = {
20 "__kernel void write_up(__global int *dst, int length)\n"
21 "{\n"
22 "\n"
23 " dst[get_global_id(0)] *= 2;\n"
24 "\n"
25 "}\n"
26 "__kernel void write_down(__global int *dst, int length)\n"
27 "{\n"
28 "\n"
29 " dst[get_global_id(0)]--;\n"
30 "\n"
31 "}\n"
32 };
33
34 #define TEST_SIZE 10000
35 #define TEST_COUNT 100
36 #define RANDOMIZE 1
37 #define DEBUG_OUT 0
38
39 /*
40 Tests event dependencies by running two kernels that use the same buffer.
41 If two_queues is set they are run in separate queues.
42 If test_enqueue_wait_for_events is set then clEnqueueWaitForEvent is called
43 between them. If test_barrier is set then clEnqueueBarrier is called between
44 them (only for single queue). If neither are set, nothing is done to prevent
45 them from executing in the wrong order. This can be used for verification.
46 */
test_event_enqueue_wait_for_events_run_test(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements,int two_queues,int two_devices,int test_enqueue_wait_for_events,int test_barrier,int use_waitlist,int use_marker)47 int test_event_enqueue_wait_for_events_run_test(
48 cl_device_id deviceID, cl_context context, cl_command_queue queue,
49 int num_elements, int two_queues, int two_devices,
50 int test_enqueue_wait_for_events, int test_barrier, int use_waitlist,
51 int use_marker)
52 {
53 cl_int error = CL_SUCCESS;
54 size_t threads[3] = { TEST_SIZE, 0, 0 };
55 int i, loop_count, event_count, expected_value, failed;
56 int expected_if_only_queue[2];
57 int max_count = TEST_SIZE;
58
59 cl_platform_id platform;
60 cl_command_queue
61 queues[2]; // Not a wrapper so we don't autorelease if they are the same
62 clCommandQueueWrapper queueWrappers[2]; // If they are different, we use the
63 // wrapper so it will auto release
64 clContextWrapper context_to_use;
65 clMemWrapper data;
66 clProgramWrapper program;
67 clKernelWrapper kernel1[TEST_COUNT], kernel2[TEST_COUNT];
68 clEventWrapper event[TEST_COUNT * 4 + 2]; // If we usemarkers we get 2 more
69 // events per iteration
70
71 if (test_enqueue_wait_for_events)
72 log_info("\tTesting with clEnqueueBarrierWithWaitList as barrier "
73 "function.\n");
74 if (test_barrier)
75 log_info("\tTesting with clEnqueueBarrierWithWaitList as barrier "
76 "function.\n");
77 if (use_waitlist)
78 log_info(
79 "\tTesting with waitlist-based depenednecies between kernels.\n");
80 if (use_marker)
81 log_info("\tTesting with clEnqueueMarker as a barrier function.\n");
82 if (test_barrier && (two_queues || two_devices))
83 {
84 log_error("\tTest requested with clEnqueueBarrier across two queues. "
85 "This is not a valid combination.\n");
86 return -1;
87 }
88
89 error = clGetPlatformIDs(1, &platform, NULL);
90 test_error(error, "clGetPlatformIDs failed.");
91
92 // If we are to use two devices, then get them and create a context with
93 // both.
94 cl_device_id *two_device_ids;
95 if (two_devices)
96 {
97 two_device_ids = (cl_device_id *)malloc(sizeof(cl_device_id) * 2);
98 cl_uint number_returned;
99 error = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 2, two_device_ids,
100 &number_returned);
101 test_error(error, "clGetDeviceIDs for CL_DEVICE_TYPE_ALL failed.");
102 if (number_returned != 2)
103 {
104 log_info("Failed to obtain two devices. Test can not run.\n");
105 free(two_device_ids);
106 return 0;
107 }
108
109 for (i = 0; i < 2; i++)
110 {
111 cl_device_type type;
112 error = clGetDeviceInfo(two_device_ids[i], CL_DEVICE_TYPE,
113 sizeof(cl_device_type), &type, NULL);
114 test_error(error, "clGetDeviceInfo failed.");
115 if (type & CL_DEVICE_TYPE_CPU)
116 log_info("\tDevice %d is CL_DEVICE_TYPE_CPU.\n", i);
117 if (type & CL_DEVICE_TYPE_GPU)
118 log_info("\tDevice %d is CL_DEVICE_TYPE_GPU.\n", i);
119 if (type & CL_DEVICE_TYPE_ACCELERATOR)
120 log_info("\tDevice %d is CL_DEVICE_TYPE_ACCELERATOR.\n", i);
121 if (type & CL_DEVICE_TYPE_DEFAULT)
122 log_info("\tDevice %d is CL_DEVICE_TYPE_DEFAULT.\n", i);
123 }
124
125 context_to_use = clCreateContext(NULL, 2, two_device_ids,
126 notify_callback, NULL, &error);
127 test_error(error, "clCreateContext failed for two devices.");
128
129 log_info("\tTesting with two devices.\n");
130 }
131 else
132 {
133 context_to_use =
134 clCreateContext(NULL, 1, &deviceID, NULL, NULL, &error);
135 test_error(error, "clCreateContext failed for one device.");
136
137 log_info("\tTesting with one device.\n");
138 }
139
140 // If we are using two queues then create them
141 cl_command_queue_properties props = CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE;
142 if (two_queues)
143 {
144 // Get a second queue
145 if (two_devices)
146 {
147 if (!checkDeviceForQueueSupport(two_device_ids[0], props)
148 || !checkDeviceForQueueSupport(two_device_ids[1], props))
149 {
150 log_info(
151 "WARNING: One or more device for multi-device test does "
152 "not support out-of-order exec mode; skipping test.\n");
153 return -1942;
154 }
155
156 queueWrappers[0] = clCreateCommandQueue(
157 context_to_use, two_device_ids[0], props, &error);
158 test_error(
159 error,
160 "clCreateCommandQueue for first queue on first device failed.");
161 queueWrappers[1] = clCreateCommandQueue(
162 context_to_use, two_device_ids[1], props, &error);
163 test_error(error,
164 "clCreateCommandQueue for second queue on second device "
165 "failed.");
166 }
167 else
168 {
169 // Single device has already been checked for out-of-order exec
170 // support
171 queueWrappers[0] =
172 clCreateCommandQueue(context_to_use, deviceID, props, &error);
173 test_error(error, "clCreateCommandQueue for first queue failed.");
174 queueWrappers[1] =
175 clCreateCommandQueue(context_to_use, deviceID, props, &error);
176 test_error(error, "clCreateCommandQueue for second queue failed.");
177 }
178 // Ugly hack to make sure we only have the wrapper auto-release if they
179 // are different queues
180 queues[0] = queueWrappers[0];
181 queues[1] = queueWrappers[1];
182 log_info("\tTesting with two queues.\n");
183 }
184 else
185 {
186 // (Note: single device has already been checked for out-of-order exec
187 // support) Otherwise create one queue and have the second one be the
188 // same
189 queueWrappers[0] =
190 clCreateCommandQueue(context_to_use, deviceID, props, &error);
191 test_error(error, "clCreateCommandQueue for first queue failed.");
192 queues[0] = queueWrappers[0];
193 queues[1] = (cl_command_queue)queues[0];
194 log_info("\tTesting with one queue.\n");
195 }
196
197
198 // Setup - create a buffer and the two kernels
199 data = clCreateBuffer(context_to_use, CL_MEM_READ_WRITE,
200 TEST_SIZE * sizeof(cl_int), NULL, &error);
201 test_error(error, "clCreateBuffer failed");
202
203
204 // Initialize the values to zero
205 cl_int *values = (cl_int *)malloc(TEST_SIZE * sizeof(cl_int));
206 for (i = 0; i < (int)TEST_SIZE; i++) values[i] = 0;
207 error =
208 clEnqueueWriteBuffer(queues[0], data, CL_TRUE, 0,
209 TEST_SIZE * sizeof(cl_int), values, 0, NULL, NULL);
210 test_error(error, "clEnqueueWriteBuffer failed");
211 expected_value = 0;
212
213 // Build the kernels
214 if (create_single_kernel_helper(context_to_use, &program, &kernel1[0], 1,
215 write_kernels, "write_up"))
216 return -1;
217
218 error = clSetKernelArg(kernel1[0], 0, sizeof(data), &data);
219 error |= clSetKernelArg(kernel1[0], 1, sizeof(max_count), &max_count);
220 test_error(error, "clSetKernelArg 1 failed");
221
222 for (i = 1; i < TEST_COUNT; i++)
223 {
224 kernel1[i] = clCreateKernel(program, "write_up", &error);
225 test_error(error, "clCreateKernel 1 failed");
226
227 error = clSetKernelArg(kernel1[i], 0, sizeof(data), &data);
228 error |= clSetKernelArg(kernel1[i], 1, sizeof(max_count), &max_count);
229 test_error(error, "clSetKernelArg 1 failed");
230 }
231
232 for (i = 0; i < TEST_COUNT; i++)
233 {
234 kernel2[i] = clCreateKernel(program, "write_down", &error);
235 test_error(error, "clCreateKernel 2 failed");
236
237 error = clSetKernelArg(kernel2[i], 0, sizeof(data), &data);
238 error |= clSetKernelArg(kernel2[i], 1, sizeof(max_count), &max_count);
239 test_error(error, "clSetKernelArg 2 failed");
240 }
241
242 // Execution - run the first kernel, then enqueue the wait on the events,
243 // then the second kernel If clEnqueueBarrierWithWaitList works, the buffer
244 // will be filled with 1s, then multiplied by 4s, then incremented to 5s,
245 // repeatedly. Otherwise the values may be 2s (if the first one doesn't
246 // work) or 8s (if the second one doesn't work).
247 if (RANDOMIZE)
248 log_info("Queues chosen randomly for each kernel execution.\n");
249 else
250 log_info("Queues chosen alternatily for each kernel execution.\n");
251
252 event_count = 0;
253 for (i = 0; i < (int)TEST_SIZE; i++) values[i] = 1;
254 error = clEnqueueWriteBuffer(queues[0], data, CL_FALSE, 0,
255 TEST_SIZE * sizeof(cl_int), values, 0, NULL,
256 &event[event_count]);
257 test_error(error, "clEnqueueWriteBuffer 2 failed");
258 expected_value = 1;
259 expected_if_only_queue[0] = 1;
260 expected_if_only_queue[1] = 1;
261
262 int queue_to_use = 1;
263 if (test_enqueue_wait_for_events)
264 {
265 error = clEnqueueBarrierWithWaitList(queues[queue_to_use], 1,
266 &event[event_count], NULL);
267 test_error(error, "Unable to queue wait for events");
268 }
269 else if (test_barrier)
270 {
271 error =
272 clEnqueueBarrierWithWaitList(queues[queue_to_use], 0, NULL, NULL);
273 test_error(error, "Unable to queue barrier");
274 }
275
276 for (loop_count = 0; loop_count < TEST_COUNT; loop_count++)
277 {
278 // Execute kernel 1
279 event_count++;
280 if (use_waitlist | use_marker)
281 {
282 if (DEBUG_OUT)
283 log_info("clEnqueueNDRangeKernel(queues[%d], kernel1[%d], 1, "
284 "NULL, threads, NULL, 1, &event[%d], &event[%d])\n",
285 queue_to_use, loop_count, event_count - 1,
286 event_count);
287 error = clEnqueueNDRangeKernel(
288 queues[queue_to_use], kernel1[loop_count], 1, NULL, threads,
289 NULL, 1, &event[event_count - 1], &event[event_count]);
290 }
291 else
292 {
293 if (DEBUG_OUT)
294 log_info("clEnqueueNDRangeKernel(queues[%d], kernel1[%d], 1, "
295 "NULL, threads, NULL, 0, NULL, &event[%d])\n",
296 queue_to_use, loop_count, event_count);
297 error = clEnqueueNDRangeKernel(
298 queues[queue_to_use], kernel1[loop_count], 1, NULL, threads,
299 NULL, 0, NULL, &event[event_count]);
300 }
301 if (error)
302 {
303 log_info("\tLoop count %d\n", loop_count);
304 print_error(error, "clEnqueueNDRangeKernel for kernel 1 failed");
305 return error;
306 }
307 expected_value *= 2;
308 expected_if_only_queue[queue_to_use] *= 2;
309
310 // If we are using a marker, it needs to go in the same queue
311 if (use_marker)
312 {
313 event_count++;
314 if (DEBUG_OUT)
315 log_info("clEnqueueMarker(queues[%d], event[%d])\n",
316 queue_to_use, event_count);
317
318 #ifdef CL_VERSION_1_2
319 error = clEnqueueMarkerWithWaitList(queues[queue_to_use], 0, NULL,
320 &event[event_count]);
321 #else
322 error = clEnqueueMarker(queues[queue_to_use], &event[event_count]);
323 #endif
324 }
325
326 // Pick the next queue to run
327 if (RANDOMIZE)
328 queue_to_use = rand() % 2;
329 else
330 queue_to_use = (queue_to_use + 1) % 2;
331
332 // Put in a barrier if requested
333 if (test_enqueue_wait_for_events)
334 {
335 if (DEBUG_OUT)
336 log_info("clEnqueueBarrierWithWaitList(queues[%d], 1, "
337 "&event[%d], NULL)\n",
338 queue_to_use, event_count);
339 error = clEnqueueBarrierWithWaitList(queues[queue_to_use], 1,
340 &event[event_count], NULL);
341 test_error(error, "Unable to queue wait for events");
342 }
343 else if (test_barrier)
344 {
345 if (DEBUG_OUT)
346 log_info("clEnqueueBarrierWithWaitList(queues[%d])\n",
347 queue_to_use);
348 error = clEnqueueBarrierWithWaitList(queues[queue_to_use], 0, NULL,
349 NULL);
350 test_error(error, "Unable to queue barrier");
351 }
352
353 // Execute Kernel 2
354 event_count++;
355 if (use_waitlist | use_marker)
356 {
357 if (DEBUG_OUT)
358 log_info("clEnqueueNDRangeKernel(queues[%d], kernel2[%d], 1, "
359 "NULL, threads, NULL, 1, &event[%d], &event[%d])\n",
360 queue_to_use, loop_count, event_count - 1,
361 event_count);
362 error = clEnqueueNDRangeKernel(
363 queues[queue_to_use], kernel2[loop_count], 1, NULL, threads,
364 NULL, 1, &event[event_count - 1], &event[event_count]);
365 }
366 else
367 {
368 if (DEBUG_OUT)
369 log_info("clEnqueueNDRangeKernel(queues[%d], kernel2[%d], 1, "
370 "NULL, threads, NULL, 0, NULL, &event[%d])\n",
371 queue_to_use, loop_count, event_count);
372 error = clEnqueueNDRangeKernel(
373 queues[queue_to_use], kernel2[loop_count], 1, NULL, threads,
374 NULL, 0, NULL, &event[event_count]);
375 }
376 if (error)
377 {
378 log_info("\tLoop count %d\n", loop_count);
379 print_error(error, "clEnqueueNDRangeKernel for kernel 2 failed");
380 return error;
381 }
382 expected_value--;
383 expected_if_only_queue[queue_to_use]--;
384
385 // If we are using a marker, it needs to go in the same queue
386 if (use_marker)
387 {
388 event_count++;
389 if (DEBUG_OUT)
390 log_info("clEnqueueMarker(queues[%d], event[%d])\n",
391 queue_to_use, event_count);
392
393 #ifdef CL_VERSION_1_2
394 error = clEnqueueMarkerWithWaitList(queues[queue_to_use], 0, NULL,
395 &event[event_count]);
396 #else
397 error = clEnqueueMarker(queues[queue_to_use], &event[event_count]);
398 #endif
399 }
400
401 // Pick the next queue to run
402 if (RANDOMIZE)
403 queue_to_use = rand() % 2;
404 else
405 queue_to_use = (queue_to_use + 1) % 2;
406
407 // Put in a barrier if requested
408 if (test_enqueue_wait_for_events)
409 {
410 if (DEBUG_OUT)
411 log_info("clEnqueueBarrierWithWaitList(queues[%d], 1, "
412 "&event[%d], NULL)\n",
413 queue_to_use, event_count);
414 error = clEnqueueBarrierWithWaitList(queues[queue_to_use], 1,
415 &event[event_count], NULL);
416 test_error(error, "Unable to queue wait for events");
417 }
418 else if (test_barrier)
419 {
420 if (DEBUG_OUT)
421 log_info("clEnqueueBarrierWithWaitList(queues[%d])\n",
422 queue_to_use);
423 error = clEnqueueBarrierWithWaitList(queues[queue_to_use], 0, NULL,
424 NULL);
425 test_error(error, "Unable to queue barrier");
426 }
427 }
428
429 // Now finish up everything
430 if (two_queues)
431 {
432 error = clFlush(queues[1]);
433 test_error(error, "clFlush[1] failed");
434 }
435
436 error = clEnqueueReadBuffer(queues[0], data, CL_TRUE, 0,
437 TEST_SIZE * sizeof(cl_int), values, 1,
438 &event[event_count], NULL);
439
440 test_error(error, "clEnqueueReadBuffer failed");
441
442 failed = 0;
443 for (i = 0; i < (int)TEST_SIZE; i++)
444 if (values[i] != expected_value)
445 {
446 failed = 1;
447 log_info("\tvalues[%d] = %d, expected %d (If only queue 1 accessed "
448 "memory: %d only queue 2 accessed memory: %d)\n",
449 i, values[i], expected_value, expected_if_only_queue[0],
450 expected_if_only_queue[1]);
451 break;
452 }
453
454 free(values);
455 if (two_devices) free(two_device_ids);
456
457 return failed;
458 }
459
test(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements,int two_queues,int two_devices,int test_enqueue_wait_for_events,int test_barrier,int use_waitlists,int use_marker)460 int test(cl_device_id deviceID, cl_context context, cl_command_queue queue,
461 int num_elements, int two_queues, int two_devices,
462 int test_enqueue_wait_for_events, int test_barrier, int use_waitlists,
463 int use_marker)
464 {
465 if (!checkDeviceForQueueSupport(deviceID,
466 CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE))
467 {
468 log_info("WARNING: Device does not support out-of-order exec mode; "
469 "skipping test.\n");
470 return 0;
471 }
472
473 log_info("Running test for baseline results to determine if out-of-order "
474 "execution can be detected...\n");
475 int baseline_results = test_event_enqueue_wait_for_events_run_test(
476 deviceID, context, queue, num_elements, two_queues, two_devices, 0, 0,
477 0, 0);
478 if (baseline_results == 0)
479 {
480 if (test_enqueue_wait_for_events)
481 log_info(
482 "WARNING: could not detect any out-of-order execution without "
483 "using clEnqueueBarrierWithWaitList, so this test is not a "
484 "valid test of out-of-order event dependencies.\n");
485 if (test_barrier)
486 log_info(
487 "WARNING: could not detect any out-of-order execution without "
488 "using clEnqueueBarrierWithWaitList, so this test is not a "
489 "valid test of out-of-order event dependencies.\n");
490 if (use_waitlists)
491 log_info("WARNING: could not detect any out-of-order execution "
492 "without using waitlists, so this test is not a valid "
493 "test of out-of-order event dependencies.\n");
494 if (use_marker)
495 log_info("WARNING: could not detect any out-of-order execution "
496 "without using clEnqueueMarker, so this test is not a "
497 "valid test of out-of-order event dependencies.\n");
498 }
499 else if (baseline_results == 1)
500 {
501 if (test_enqueue_wait_for_events)
502 log_info("Detected incorrect execution (possibly out-of-order) "
503 "without clEnqueueBarrierWithWaitList. Test can be a "
504 "valid test of out-of-order event dependencies.\n");
505 if (test_barrier)
506 log_info("Detected incorrect execution (possibly out-of-order) "
507 "without clEnqueueBarrierWithWaitList. Test can be a "
508 "valid test of out-of-order event dependencies.\n");
509 if (use_waitlists)
510 log_info("Detected incorrect execution (possibly out-of-order) "
511 "without waitlists. Test can be a valid test of "
512 "out-of-order event dependencies.\n");
513 if (use_marker)
514 log_info("Detected incorrect execution (possibly out-of-order) "
515 "without clEnqueueMarker. Test can be a valid test of "
516 "out-of-order event dependencies.\n");
517 }
518 else if (baseline_results == -1942)
519 {
520 // Just ignore and return (out-of-order exec mode not supported)
521 return 0;
522 }
523 else
524 {
525 print_error(baseline_results, "Baseline run failed");
526 return baseline_results;
527 }
528 log_info("Running test for actual results...\n");
529 return test_event_enqueue_wait_for_events_run_test(
530 deviceID, context, queue, num_elements, two_queues, two_devices,
531 test_enqueue_wait_for_events, test_barrier, use_waitlists, use_marker);
532 }
533
534
test_out_of_order_event_waitlist_single_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)535 int test_out_of_order_event_waitlist_single_queue(cl_device_id deviceID,
536 cl_context context,
537 cl_command_queue queue,
538 int num_elements)
539 {
540 int two_queues = 0;
541 int two_devices = 0;
542 int test_enqueue_wait_for_events = 0;
543 int test_barrier = 0;
544 int use_waitlists = 1;
545 int use_marker = 0;
546 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
547 test_enqueue_wait_for_events, test_barrier, use_waitlists,
548 use_marker);
549 }
550
test_out_of_order_event_waitlist_multi_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)551 int test_out_of_order_event_waitlist_multi_queue(cl_device_id deviceID,
552 cl_context context,
553 cl_command_queue queue,
554 int num_elements)
555 {
556 int two_queues = 1;
557 int two_devices = 0;
558 int test_enqueue_wait_for_events = 0;
559 int test_barrier = 0;
560 int use_waitlists = 1;
561 int use_marker = 0;
562 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
563 test_enqueue_wait_for_events, test_barrier, use_waitlists,
564 use_marker);
565 }
566
test_out_of_order_event_waitlist_multi_queue_multi_device(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)567 int test_out_of_order_event_waitlist_multi_queue_multi_device(
568 cl_device_id deviceID, cl_context context, cl_command_queue queue,
569 int num_elements)
570 {
571 int two_queues = 1;
572 int two_devices = 1;
573 int test_enqueue_wait_for_events = 0;
574 int test_barrier = 0;
575 int use_waitlists = 1;
576 int use_marker = 0;
577 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
578 test_enqueue_wait_for_events, test_barrier, use_waitlists,
579 use_marker);
580 }
581
582
test_out_of_order_event_enqueue_wait_for_events_single_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)583 int test_out_of_order_event_enqueue_wait_for_events_single_queue(
584 cl_device_id deviceID, cl_context context, cl_command_queue queue,
585 int num_elements)
586 {
587 int two_queues = 0;
588 int two_devices = 0;
589 int test_enqueue_wait_for_events = 1;
590 int test_barrier = 0;
591 int use_waitlists = 0;
592 int use_marker = 0;
593 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
594 test_enqueue_wait_for_events, test_barrier, use_waitlists,
595 use_marker);
596 }
597
test_out_of_order_event_enqueue_wait_for_events_multi_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)598 int test_out_of_order_event_enqueue_wait_for_events_multi_queue(
599 cl_device_id deviceID, cl_context context, cl_command_queue queue,
600 int num_elements)
601 {
602 int two_queues = 1;
603 int two_devices = 0;
604 int test_enqueue_wait_for_events = 1;
605 int test_barrier = 0;
606 int use_waitlists = 0;
607 int use_marker = 0;
608 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
609 test_enqueue_wait_for_events, test_barrier, use_waitlists,
610 use_marker);
611 }
612
613
test_out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)614 int test_out_of_order_event_enqueue_wait_for_events_multi_queue_multi_device(
615 cl_device_id deviceID, cl_context context, cl_command_queue queue,
616 int num_elements)
617 {
618 int two_queues = 1;
619 int two_devices = 1;
620 int test_enqueue_wait_for_events = 1;
621 int test_barrier = 0;
622 int use_waitlists = 0;
623 int use_marker = 0;
624 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
625 test_enqueue_wait_for_events, test_barrier, use_waitlists,
626 use_marker);
627 }
628
629
test_out_of_order_event_enqueue_barrier_single_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)630 int test_out_of_order_event_enqueue_barrier_single_queue(cl_device_id deviceID,
631 cl_context context,
632 cl_command_queue queue,
633 int num_elements)
634 {
635 int two_queues = 0;
636 int two_devices = 0;
637 int test_enqueue_wait_for_events = 0;
638 int test_barrier = 1;
639 int use_waitlists = 0;
640 int use_marker = 0;
641 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
642 test_enqueue_wait_for_events, test_barrier, use_waitlists,
643 use_marker);
644 }
645
646
test_out_of_order_event_enqueue_marker_single_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)647 int test_out_of_order_event_enqueue_marker_single_queue(cl_device_id deviceID,
648 cl_context context,
649 cl_command_queue queue,
650 int num_elements)
651 {
652 int two_queues = 0;
653 int two_devices = 0;
654 int test_enqueue_wait_for_events = 0;
655 int test_barrier = 0;
656 int use_waitlists = 0;
657 int use_marker = 1;
658 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
659 test_enqueue_wait_for_events, test_barrier, use_waitlists,
660 use_marker);
661 }
662
test_out_of_order_event_enqueue_marker_multi_queue(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)663 int test_out_of_order_event_enqueue_marker_multi_queue(cl_device_id deviceID,
664 cl_context context,
665 cl_command_queue queue,
666 int num_elements)
667 {
668 int two_queues = 1;
669 int two_devices = 0;
670 int test_enqueue_wait_for_events = 0;
671 int test_barrier = 0;
672 int use_waitlists = 0;
673 int use_marker = 1;
674 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
675 test_enqueue_wait_for_events, test_barrier, use_waitlists,
676 use_marker);
677 }
678
679
test_out_of_order_event_enqueue_marker_multi_queue_multi_device(cl_device_id deviceID,cl_context context,cl_command_queue queue,int num_elements)680 int test_out_of_order_event_enqueue_marker_multi_queue_multi_device(
681 cl_device_id deviceID, cl_context context, cl_command_queue queue,
682 int num_elements)
683 {
684 int two_queues = 1;
685 int two_devices = 1;
686 int test_enqueue_wait_for_events = 0;
687 int test_barrier = 0;
688 int use_waitlists = 0;
689 int use_marker = 1;
690 return test(deviceID, context, queue, num_elements, two_queues, two_devices,
691 test_enqueue_wait_for_events, test_barrier, use_waitlists,
692 use_marker);
693 }
694