Added a test to verify the flush behavior of clReleaseCommandQueue wh…

…en multiple dependent queues are used
shajder · Oct 30, 2024 · 1131e19 · 1131e19
1 parent 6337d9b
commit 1131e19
Show file tree

Hide file tree

Showing 3 changed files with 68 additions and 1 deletion.
diff --git a/test_conformance/api/main.cpp b/test_conformance/api/main.cpp
@@ -120,7 +120,10 @@ test_definition test_list[] = {
     ADD_TEST(get_image1d_info),
     ADD_TEST(get_image1d_array_info),
     ADD_TEST(get_image2d_array_info),
+
     ADD_TEST(queue_flush_on_release),
+    ADD_TEST(multi_queue_flush_on_release),
+
     ADD_TEST(queue_hint),
     ADD_TEST(queue_properties),
     ADD_TEST_VERSION(sub_group_dispatch, Version(2, 1)),

diff --git a/test_conformance/api/procs.h b/test_conformance/api/procs.h
@@ -140,6 +140,10 @@ extern int test_queue_flush_on_release(cl_device_id deviceID,
                                        cl_context context,
                                        cl_command_queue queue,
                                        int num_elements);
+extern int test_multi_queue_flush_on_release(cl_device_id deviceID,
+                                             cl_context context,
+                                             cl_command_queue defaultQueue,
+                                             int num_elements);
 extern int test_buffer_properties_queries(cl_device_id deviceID,
                                           cl_context context,
                                           cl_command_queue queue,

diff --git a/test_conformance/api/test_queue.cpp b/test_conformance/api/test_queue.cpp
@@ -1,5 +1,5 @@
 //
-// Copyright (c) 2020 The Khronos Group Inc.
+// Copyright (c) 2024 The Khronos Group Inc.
 //
 // Licensed under the Apache License, Version 2.0 (the "License");
 // you may not use this file except in compliance with the License.
@@ -59,3 +59,63 @@ int test_queue_flush_on_release(cl_device_id deviceID, cl_context context,
 
     return success ? TEST_PASS : TEST_FAIL;
 }
+
+int test_multi_queue_flush_on_release(cl_device_id deviceID, cl_context context,
+                                      cl_command_queue defaultQueue,
+                                      int num_elements)
+{
+    cl_int err;
+
+    // Create A command queue
+    cl_command_queue queue_A = clCreateCommandQueue(context, deviceID, 0, &err);
+    test_error(err, "Could not create command queue A");
+
+    // Create B command queue
+    cl_command_queue queue_B = clCreateCommandQueue(context, deviceID, 0, &err);
+    test_error(err, "Could not create command queue B");
+
+    // Create a kernel
+    clProgramWrapper program;
+    clKernelWrapper kernel;
+    const char *source = "void kernel test(){}";
+    err = create_single_kernel_helper(context, &program, &kernel, 1, &source,
+                                      "test");
+    test_error(err, "Could not create kernel");
+
+    // Enqueue the kernel on queue_A and obtain event to synchronize with
+    // queue_B
+    size_t gws = num_elements;
+    clEventWrapper event_A;
+    err = clEnqueueNDRangeKernel(queue_A, kernel, 1, nullptr, &gws, nullptr, 0,
+                                 nullptr, &event_A);
+    test_error(err, "Could not enqueue kernel");
+
+    // Enqueue the kernel on queue_B using event_A for synchronization and
+    // create event_B to track completion
+    clEventWrapper event_B;
+    err = clEnqueueNDRangeKernel(queue_B, kernel, 1, nullptr, &gws, nullptr, 1,
+                                 &event_A, &event_B);
+    test_error(err, "Could not enqueue kernel");
+
+    // Release queue_A, which performs an implicit flush to issue any previously
+    // queued OpenCL commands
+    err = clReleaseCommandQueue(queue_A);
+    test_error(err, "clReleaseCommandQueue failed");
+
+    err = clFinish(queue_B);
+    test_error(err, "clFinish failed");
+
+    // Wait for kernel to execute since the queue must flush on release
+    bool success = poll_until(2000, 50, [&event_B]() {
+        cl_int status;
+        cl_int err = clGetEventInfo(event_B, CL_EVENT_COMMAND_EXECUTION_STATUS,
+                                    sizeof(cl_int), &status, nullptr);
+        if ((err != CL_SUCCESS) || (status != CL_COMPLETE))
+        {
+            return false;
+        }
+        return true;
+    });
+
+    return success ? TEST_PASS : TEST_FAIL;
+}