diff --git a/test_common/harness/kernelHelpers.cpp b/test_common/harness/kernelHelpers.cpp index 566700eb..bca058c8 100644 --- a/test_common/harness/kernelHelpers.cpp +++ b/test_common/harness/kernelHelpers.cpp @@ -1693,3 +1693,23 @@ Version get_max_OpenCL_C_for_context(cl_context context) }); return current_version; } + +bool poll_until(unsigned timeout_ms, unsigned interval_ms, + std::function fn) +{ + unsigned time_spent_ms = 0; + bool ret = false; + + while (time_spent_ms < timeout_ms) + { + ret = fn(); + if (ret) + { + break; + } + usleep(interval_ms * 1000); + time_spent_ms += interval_ms; + } + + return ret; +} diff --git a/test_common/harness/kernelHelpers.h b/test_common/harness/kernelHelpers.h index c4a82431..7f972070 100644 --- a/test_common/harness/kernelHelpers.h +++ b/test_common/harness/kernelHelpers.h @@ -40,6 +40,8 @@ #include "deviceInfo.h" #include "harness/alloc.h" +#include + /* * The below code is intended to be used at the top of kernels that appear inline in files to set line and file info for the kernel: * @@ -181,4 +183,8 @@ Version get_device_cl_c_version(cl_device_id device); // the OpenCL C version supported by all devices in a context. Version get_max_OpenCL_C_for_context(cl_context context); +// Poll fn every interval_ms until timeout_ms or it returns true +bool poll_until(unsigned timeout_ms, unsigned interval_ms, + std::function fn); + #endif // _kernelHelpers_h diff --git a/test_conformance/api/CMakeLists.txt b/test_conformance/api/CMakeLists.txt index 4d44f959..20d69830 100644 --- a/test_conformance/api/CMakeLists.txt +++ b/test_conformance/api/CMakeLists.txt @@ -23,6 +23,7 @@ set(${MODULE_NAME}_SOURCES test_kernel_arg_info_compatibility.cpp test_null_buffer_arg.cpp test_mem_object_info.cpp + test_queue.cpp test_queue_hint.cpp test_queue_properties.cpp test_sub_group_dispatch.cpp diff --git a/test_conformance/api/main.cpp b/test_conformance/api/main.cpp index 5fc5db4c..9308fa8d 100644 --- a/test_conformance/api/main.cpp +++ b/test_conformance/api/main.cpp @@ -117,6 +117,7 @@ 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(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 index 73b96eea..b55f5a3b 100644 --- a/test_conformance/api/procs.h +++ b/test_conformance/api/procs.h @@ -121,6 +121,10 @@ extern int test_sub_group_dispatch(cl_device_id deviceID, cl_context contex extern int test_clone_kernel(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); extern int test_zero_sized_enqueue(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements); extern int test_queue_properties( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements ); +extern int test_queue_flush_on_release(cl_device_id deviceID, + cl_context context, + cl_command_queue queue, + 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 new file mode 100644 index 00000000..0599784a --- /dev/null +++ b/test_conformance/api/test_queue.cpp @@ -0,0 +1,61 @@ +// +// Copyright (c) 2020 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. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. +// + + +#include "testBase.h" +#include "harness/typeWrappers.h" + +int test_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 = clCreateCommandQueue(context, deviceID, 0, &err); + test_error(err, "Could not create command queue"); + + // 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 + size_t gws = 1; + clEventWrapper event; + err = clEnqueueNDRangeKernel(queue, kernel, 1, nullptr, &gws, nullptr, 0, + nullptr, &event); + test_error(err, "Could not enqueue kernel"); + + // Release the queue + err = clReleaseCommandQueue(queue); + + // Wait for kernel to execute since the queue must flush on release + bool success = poll_until(2000, 50, [event]() { + cl_int status; + cl_int err = clGetEventInfo(event, 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; +}