diff --git a/test_conformance/extensions/cl_khr_command_buffer/CMakeLists.txt b/test_conformance/extensions/cl_khr_command_buffer/CMakeLists.txt index ac259f6d..4ba6eeeb 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/CMakeLists.txt +++ b/test_conformance/extensions/cl_khr_command_buffer/CMakeLists.txt @@ -3,6 +3,7 @@ set(MODULE_NAME CL_KHR_COMMAND_BUFFER) set(${MODULE_NAME}_SOURCES main.cpp basic_command_buffer.cpp + command_buffer_queue_substitution.cpp ) include(../../CMakeCommon.txt) diff --git a/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.cpp b/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.cpp index d38b0152..d064ea42 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.cpp +++ b/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.cpp @@ -13,158 +13,148 @@ // See the License for the specific language governing permissions and // limitations under the License. // -#include "command_buffer_test_base.h" +#include "basic_command_buffer.h" #include "procs.h" -#include "harness/typeWrappers.h" #include #include #include -#define CHECK_VERIFICATION_ERROR(reference, result, index) \ - { \ - if (reference != result) \ - { \ - log_error("Expected %d was %d at index %u\n", reference, result, \ - index); \ - return TEST_FAIL; \ - } \ + +BasicCommandBufferTest::BasicCommandBufferTest(cl_device_id device, + cl_context context, + cl_command_queue queue) + : CommandBufferTestBase(device), context(context), queue(nullptr), + num_elements(0), command_buffer(this), simultaneous_use_support(false), + out_of_order_support(false), + // try to use simultaneous path by default + simultaneous_use_requested(true), + // due to simultaneous cases extend buffer size + buffer_size_multiplier(1) + +{ + cl_int error = clRetainCommandQueue(queue); + if (error != CL_SUCCESS) + { + throw std::runtime_error("clRetainCommandQueue failed\n"); } + this->queue = queue; +} + +bool BasicCommandBufferTest::Skip() +{ + cl_command_queue_properties required_properties; + cl_int error = clGetDeviceInfo( + device, CL_DEVICE_COMMAND_BUFFER_REQUIRED_QUEUE_PROPERTIES_KHR, + sizeof(required_properties), &required_properties, NULL); + test_error(error, + "Unable to query " + "CL_DEVICE_COMMAND_BUFFER_REQUIRED_QUEUE_PROPERTIES_KHR"); + + cl_command_queue_properties queue_properties; + + error = clGetCommandQueueInfo(queue, CL_QUEUE_PROPERTIES, + sizeof(queue_properties), &queue_properties, + NULL); + test_error(error, "Unable to query CL_QUEUE_PROPERTIES"); + + + // Query if device supports simultaneous use + cl_device_command_buffer_capabilities_khr capabilities; + error = clGetDeviceInfo(device, CL_DEVICE_COMMAND_BUFFER_CAPABILITIES_KHR, + sizeof(capabilities), &capabilities, NULL); + test_error(error, + "Unable to query CL_DEVICE_COMMAND_BUFFER_CAPABILITIES_KHR"); + simultaneous_use_support = simultaneous_use_requested + && (capabilities & CL_COMMAND_BUFFER_CAPABILITY_SIMULTANEOUS_USE_KHR) + != 0; + out_of_order_support = + capabilities & CL_COMMAND_BUFFER_CAPABILITY_OUT_OF_ORDER_KHR; + + // Skip if queue properties don't contain those required + return required_properties != (required_properties & queue_properties); +} + +cl_int BasicCommandBufferTest::SetUp(int elements) +{ + cl_int error = init_extension_functions(); + if (error != CL_SUCCESS) + { + return error; + } + + if (elements <= 0) + { + return CL_INVALID_VALUE; + } + num_elements = static_cast(elements); + + // Kernel performs a parallel copy from an input buffer to output buffer + // is created. + const char *kernel_str = + R"( + __kernel void copy(__global int* in, __global int* out, __global int* offset) { + size_t id = get_global_id(0); + int ind = offset[0] + id; + out[ind] = in[ind]; + })"; + + error = create_single_kernel_helper_create_program(context, &program, 1, + &kernel_str); + test_error(error, "Failed to create program with source"); + + error = clBuildProgram(program, 1, &device, nullptr, nullptr, nullptr); + test_error(error, "Failed to build program"); + + in_mem = + clCreateBuffer(context, CL_MEM_READ_ONLY, + sizeof(cl_int) * num_elements * buffer_size_multiplier, + nullptr, &error); + test_error(error, "clCreateBuffer failed"); + + out_mem = + clCreateBuffer(context, CL_MEM_WRITE_ONLY, + sizeof(cl_int) * num_elements * buffer_size_multiplier, + nullptr, &error); + test_error(error, "clCreateBuffer failed"); + + cl_int offset = 0; + off_mem = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(cl_int), &offset, &error); + test_error(error, "clCreateBuffer failed"); + + kernel = clCreateKernel(program, "copy", &error); + test_error(error, "Failed to create copy kernel"); + + error = clSetKernelArg(kernel, 0, sizeof(in_mem), &in_mem); + test_error(error, "clSetKernelArg failed"); + + error = clSetKernelArg(kernel, 1, sizeof(out_mem), &out_mem); + test_error(error, "clSetKernelArg failed"); + + error = clSetKernelArg(kernel, 2, sizeof(off_mem), &off_mem); + test_error(error, "clSetKernelArg failed"); + + if (simultaneous_use_support) + { + cl_command_buffer_properties_khr properties[3] = { + CL_COMMAND_BUFFER_FLAGS_KHR, CL_COMMAND_BUFFER_SIMULTANEOUS_USE_KHR, + 0 + }; + command_buffer = + clCreateCommandBufferKHR(1, &queue, properties, &error); + } + else + { + command_buffer = clCreateCommandBufferKHR(1, &queue, nullptr, &error); + } + test_error(error, "clCreateCommandBufferKHR failed"); + + return CL_SUCCESS; +} namespace { -// Helper test fixture for constructing OpenCL objects used in testing -// a variety of simple command-buffer enqueue scenarios. -struct BasicCommandBufferTest : CommandBufferTestBase -{ - - BasicCommandBufferTest(cl_device_id device, cl_context context, - cl_command_queue queue) - : CommandBufferTestBase(device), context(context), queue(queue), - command_buffer(this), num_elements(0), simultaneous_use(false), - out_of_order_support(false) - {} - - virtual bool Skip() - { - cl_command_queue_properties required_properties; - cl_int error = clGetDeviceInfo( - device, CL_DEVICE_COMMAND_BUFFER_REQUIRED_QUEUE_PROPERTIES_KHR, - sizeof(required_properties), &required_properties, NULL); - test_error(error, - "Unable to query " - "CL_DEVICE_COMMAND_BUFFER_REQUIRED_QUEUE_PROPERTIES_KHR"); - - cl_command_queue_properties queue_properties; - - error = clGetCommandQueueInfo(queue, CL_QUEUE_PROPERTIES, - sizeof(queue_properties), - &queue_properties, NULL); - test_error(error, "Unable to query CL_QUEUE_PROPERTIES"); - - // Skip if queue properties don't contain those required - return required_properties != (required_properties & queue_properties); - } - - virtual cl_int SetUp(int elements) - { - cl_int error = init_extension_functions(); - if (error != CL_SUCCESS) - { - return error; - } - - // Query if device supports simultaneous use - cl_device_command_buffer_capabilities_khr capabilities; - error = - clGetDeviceInfo(device, CL_DEVICE_COMMAND_BUFFER_CAPABILITIES_KHR, - sizeof(capabilities), &capabilities, NULL); - test_error(error, - "Unable to query CL_DEVICE_COMMAND_BUFFER_CAPABILITIES_KHR"); - simultaneous_use = - capabilities & CL_COMMAND_BUFFER_CAPABILITY_SIMULTANEOUS_USE_KHR; - out_of_order_support = - capabilities & CL_COMMAND_BUFFER_CAPABILITY_OUT_OF_ORDER_KHR; - - if (elements <= 0) - { - return CL_INVALID_VALUE; - } - num_elements = static_cast(elements); - - // Kernel performs a parallel copy from an input buffer to output buffer - // is created. - const char *kernel_str = - R"( - __kernel void copy(__global int* in, __global int* out) { - size_t id = get_global_id(0); - out[id] = in[id]; - })"; - - error = create_single_kernel_helper_create_program(context, &program, 1, - &kernel_str); - test_error(error, "Failed to create program with source"); - - error = clBuildProgram(program, 1, &device, nullptr, nullptr, nullptr); - test_error(error, "Failed to build program"); - - in_mem = clCreateBuffer(context, CL_MEM_READ_ONLY, - sizeof(cl_int) * num_elements, nullptr, &error); - test_error(error, "clCreateBuffer failed"); - - out_mem = - clCreateBuffer(context, CL_MEM_WRITE_ONLY, - sizeof(cl_int) * num_elements, nullptr, &error); - test_error(error, "clCreateBuffer failed"); - - kernel = clCreateKernel(program, "copy", &error); - test_error(error, "Failed to create copy kernel"); - - error = clSetKernelArg(kernel, 0, sizeof(in_mem), &in_mem); - test_error(error, "clSetKernelArg failed"); - - error = clSetKernelArg(kernel, 1, sizeof(out_mem), &out_mem); - test_error(error, "clSetKernelArg failed"); - - if (simultaneous_use) - { - cl_command_buffer_properties_khr properties[3] = { - CL_COMMAND_BUFFER_FLAGS_KHR, - CL_COMMAND_BUFFER_SIMULTANEOUS_USE_KHR, 0 - }; - command_buffer = - clCreateCommandBufferKHR(1, &queue, properties, &error); - } - else - { - command_buffer = - clCreateCommandBufferKHR(1, &queue, nullptr, &error); - } - test_error(error, "clCreateCommandBufferKHR failed"); - - return CL_SUCCESS; - } - - // Test body returning an OpenCL error code - virtual cl_int Run() = 0; - - -protected: - size_t data_size() const { return num_elements * sizeof(cl_int); } - - cl_context context; - cl_command_queue queue; - clCommandBufferWrapper command_buffer; - clProgramWrapper program; - clKernelWrapper kernel; - clMemWrapper in_mem, out_mem; - size_t num_elements; - - // Device support query results - bool simultaneous_use; - bool out_of_order_support; -}; - // Test enqueuing a command-buffer containing a single NDRange command once struct BasicEnqueueTest : public BasicCommandBufferTest { @@ -375,7 +365,7 @@ struct ExplicitFlushTest : public BasicCommandBufferTest bool Skip() override { - return !simultaneous_use || BasicCommandBufferTest::Skip(); + return BasicCommandBufferTest::Skip() || !simultaneous_use_support; } }; @@ -431,7 +421,7 @@ struct InterleavedEnqueueTest : public BasicCommandBufferTest bool Skip() override { - return !simultaneous_use || BasicCommandBufferTest::Skip(); + return BasicCommandBufferTest::Skip() || !simultaneous_use_support; } }; @@ -495,13 +485,6 @@ struct OutOfOrderTest : public BasicCommandBufferTest cl_int error = BasicCommandBufferTest::SetUp(elements); test_error(error, "BasicCommandBufferTest::SetUp failed"); - if (!out_of_order_support) - { - // Test will skip as device doesn't support out-of-order - // command-buffers - return CL_SUCCESS; - } - out_of_order_queue = clCreateCommandQueue( context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &error); test_error(error, "Unable to create command queue to test with"); @@ -515,7 +498,7 @@ struct OutOfOrderTest : public BasicCommandBufferTest bool Skip() override { - return !out_of_order_support || BasicCommandBufferTest::Skip(); + return BasicCommandBufferTest::Skip() || !out_of_order_support; } clCommandQueueWrapper out_of_order_queue; @@ -523,28 +506,6 @@ struct OutOfOrderTest : public BasicCommandBufferTest clEventWrapper event; }; -#undef CHECK_VERIFICATION_ERROR - -template -int MakeAndRunTest(cl_device_id device, cl_context context, - cl_command_queue queue, int num_elements) -{ - CHECK_COMMAND_BUFFER_EXTENSION_AVAILABLE(device); - - auto test_fixture = T(device, context, queue); - cl_int error = test_fixture.SetUp(num_elements); - test_error_ret(error, "Error in test initialization", TEST_FAIL); - - if (test_fixture.Skip()) - { - return TEST_SKIPPED_ITSELF; - } - - error = test_fixture.Run(); - test_error_ret(error, "Test Failed", TEST_FAIL); - - return TEST_PASS; -} } // anonymous namespace int test_single_ndrange(cl_device_id device, cl_context context, diff --git a/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.h b/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.h new file mode 100644 index 00000000..4a46a141 --- /dev/null +++ b/test_conformance/extensions/cl_khr_command_buffer/basic_command_buffer.h @@ -0,0 +1,100 @@ +// +// Copyright (c) 2022 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. + +#ifndef _CL_KHR_BASIC_COMMAND_BUFFER_H +#define _CL_KHR_BASIC_COMMAND_BUFFER_H + +#include "command_buffer_test_base.h" +#include "harness/typeWrappers.h" + +#define ADD_PROP(prop) \ + { \ + prop, #prop \ + } + +#define CHECK_VERIFICATION_ERROR(reference, result, index) \ + { \ + if (reference != result) \ + { \ + log_error("Expected %d was %d at index %u\n", reference, result, \ + index); \ + return TEST_FAIL; \ + } \ + } + +// Helper test fixture for constructing OpenCL objects used in testing +// a variety of simple command-buffer enqueue scenarios. +struct BasicCommandBufferTest : CommandBufferTestBase +{ + + BasicCommandBufferTest(cl_device_id device, cl_context context, + cl_command_queue queue); + + virtual bool Skip(); + virtual cl_int SetUp(int elements); + + // Test body returning an OpenCL error code + virtual cl_int Run() = 0; + +protected: + size_t data_size() const { return num_elements * sizeof(cl_int); } + + cl_context context; + clCommandQueueWrapper queue; + clCommandBufferWrapper command_buffer; + clProgramWrapper program; + clKernelWrapper kernel; + clMemWrapper in_mem, out_mem, off_mem; + size_t num_elements; + + // Device support query results + bool simultaneous_use_support; + bool out_of_order_support; + + // user request for simultaneous use + bool simultaneous_use_requested; + unsigned buffer_size_multiplier; +}; + +template +int MakeAndRunTest(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements) +{ + CHECK_COMMAND_BUFFER_EXTENSION_AVAILABLE(device); + + try + { + auto test_fixture = T(device, context, queue); + + if (test_fixture.Skip()) + { + return TEST_SKIPPED_ITSELF; + } + + cl_int error = test_fixture.SetUp(num_elements); + test_error_ret(error, "Error in test initialization", TEST_FAIL); + + error = test_fixture.Run(); + test_error_ret(error, "Test Failed", TEST_FAIL); + } catch (const std::runtime_error &e) + { + log_error("%s", e.what()); + return TEST_FAIL; + } + + return TEST_PASS; +} + +#endif // _CL_KHR_BASIC_COMMAND_BUFFER_H diff --git a/test_conformance/extensions/cl_khr_command_buffer/command_buffer_queue_substitution.cpp b/test_conformance/extensions/cl_khr_command_buffer/command_buffer_queue_substitution.cpp new file mode 100644 index 00000000..7aa262aa --- /dev/null +++ b/test_conformance/extensions/cl_khr_command_buffer/command_buffer_queue_substitution.cpp @@ -0,0 +1,278 @@ +// +// Copyright (c) 2022 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 "basic_command_buffer.h" +#include "procs.h" + +#include + +namespace { + +//////////////////////////////////////////////////////////////////////////////// +// Command-queue substitution tests which handles below cases: +// -substitution on queue without properties +// -substitution on queue with properties +// -simultaneous use queue substitution + +template +struct SubstituteQueueTest : public BasicCommandBufferTest +{ + SubstituteQueueTest(cl_device_id device, cl_context context, + cl_command_queue queue) + : BasicCommandBufferTest(device, context, queue), + properties_use_requested(prop_use), user_event(nullptr) + { + simultaneous_use_requested = simul_use; + if (simul_use) buffer_size_multiplier = 2; + } + + //-------------------------------------------------------------------------- + bool Skip() override + { + if (properties_use_requested) + { + Version version = get_device_cl_version(device); + const cl_device_info host_queue_query = version >= Version(2, 0) + ? CL_DEVICE_QUEUE_ON_HOST_PROPERTIES + : CL_DEVICE_QUEUE_PROPERTIES; + + cl_queue_properties host_queue_props = 0; + int error = clGetDeviceInfo(device, host_queue_query, + sizeof(host_queue_props), + &host_queue_props, NULL); + test_error(error, "clGetDeviceInfo failed"); + + if ((host_queue_props & CL_QUEUE_PROFILING_ENABLE) == 0) + return true; + } + + return BasicCommandBufferTest::Skip() + || (simultaneous_use_requested && !simultaneous_use_support); + } + + //-------------------------------------------------------------------------- + cl_int SetUp(int elements) override + { + // By default command queue is created without properties, + // if test requires queue with properties default queue must be + // replaced. + if (properties_use_requested) + { + // due to the skip condition + cl_int error = CL_SUCCESS; + queue = clCreateCommandQueue(context, device, + CL_QUEUE_PROFILING_ENABLE, &error); + test_error( + error, + "clCreateCommandQueue with CL_QUEUE_PROFILING_ENABLE failed"); + } + + return BasicCommandBufferTest::SetUp(elements); + } + + //-------------------------------------------------------------------------- + cl_int Run() override + { + // record command buffer with primary queue + cl_int error = RecordCommandBuffer(); + test_error(error, "RecordCommandBuffer failed"); + + // create substitute queue + clCommandQueueWrapper new_queue; + if (properties_use_requested) + { + new_queue = clCreateCommandQueue(context, device, + CL_QUEUE_PROFILING_ENABLE, &error); + test_error( + error, + "clCreateCommandQueue with CL_QUEUE_PROFILING_ENABLE failed"); + } + else + { + const cl_command_queue_properties queue_properties = 0; + new_queue = + clCreateCommandQueue(context, device, queue_properties, &error); + test_error(error, "clCreateCommandQueue failed"); + } + + if (simultaneous_use_support) + { + // enque simultaneous command-buffers with substitute queue + error = RunSimultaneous(new_queue); + test_error(error, "RunSimultaneous failed"); + } + else + { + // enque single command-buffer with substitute queue + error = RunSingle(new_queue); + test_error(error, "RunSingle failed"); + } + + return CL_SUCCESS; + } + + //-------------------------------------------------------------------------- + cl_int RecordCommandBuffer() + { + cl_int error = clCommandNDRangeKernelKHR( + command_buffer, nullptr, nullptr, kernel, 1, nullptr, &num_elements, + nullptr, 0, nullptr, nullptr, nullptr); + test_error(error, "clCommandNDRangeKernelKHR failed"); + + error = clFinalizeCommandBufferKHR(command_buffer); + test_error(error, "clFinalizeCommandBufferKHR failed"); + return CL_SUCCESS; + } + + //-------------------------------------------------------------------------- + cl_int RunSingle(const cl_command_queue& q) + { + cl_int error = CL_SUCCESS; + std::vector output_data(num_elements); + + error = clEnqueueFillBuffer(q, in_mem, &pattern_pri, sizeof(cl_int), 0, + data_size(), 0, nullptr, nullptr); + test_error(error, "clEnqueueFillBuffer failed"); + + cl_command_queue queues[] = { q }; + error = clEnqueueCommandBufferKHR(1, queues, command_buffer, 0, nullptr, + nullptr); + test_error(error, "clEnqueueCommandBufferKHR failed"); + + error = clEnqueueReadBuffer(q, out_mem, CL_TRUE, 0, data_size(), + output_data.data(), 0, nullptr, nullptr); + test_error(error, "clEnqueueReadBuffer failed"); + + error = clFinish(q); + test_error(error, "clFinish failed"); + + for (size_t i = 0; i < num_elements; i++) + { + CHECK_VERIFICATION_ERROR(pattern_pri, output_data[i], i); + } + + return CL_SUCCESS; + } + + //-------------------------------------------------------------------------- + struct SimulPassData + { + cl_int pattern; + cl_int offset; + cl_command_queue queue; + std::vector output_buffer; + }; + + //-------------------------------------------------------------------------- + cl_int EnqueueSimultaneousPass(SimulPassData& pd) + { + cl_int error = clEnqueueFillBuffer( + pd.queue, in_mem, &pd.pattern, sizeof(cl_int), + pd.offset * sizeof(cl_int), data_size(), 0, nullptr, nullptr); + test_error(error, "clEnqueueFillBuffer failed"); + + error = + clEnqueueFillBuffer(pd.queue, off_mem, &pd.offset, sizeof(cl_int), + 0, sizeof(cl_int), 0, nullptr, nullptr); + test_error(error, "clEnqueueFillBuffer failed"); + + if (!user_event) + { + user_event = clCreateUserEvent(context, &error); + test_error(error, "clCreateUserEvent failed"); + } + + cl_command_queue queues[] = { pd.queue }; + error = clEnqueueCommandBufferKHR(1, queues, command_buffer, 1, + &user_event, nullptr); + test_error(error, "clEnqueueCommandBufferKHR failed"); + + error = clEnqueueReadBuffer( + pd.queue, out_mem, CL_FALSE, pd.offset * sizeof(cl_int), + data_size(), pd.output_buffer.data(), 0, nullptr, nullptr); + + test_error(error, "clEnqueueReadBuffer failed"); + + return CL_SUCCESS; + } + + //-------------------------------------------------------------------------- + cl_int RunSimultaneous(const cl_command_queue& q) + { + cl_int error = CL_SUCCESS; + cl_int offset = static_cast(num_elements); + + std::vector simul_passes = { + { pattern_pri, 0, q, std::vector(num_elements) }, + { pattern_sec, offset, q, std::vector(num_elements) } + }; + + for (auto&& pass : simul_passes) + { + error = EnqueueSimultaneousPass(pass); + test_error(error, "EnqueuePass failed"); + } + + error = clSetUserEventStatus(user_event, CL_COMPLETE); + test_error(error, "clSetUserEventStatus failed"); + + for (auto&& pass : simul_passes) + { + error = clFinish(pass.queue); + test_error(error, "clFinish failed"); + + auto& res_data = pass.output_buffer; + + for (size_t i = 0; i < num_elements; i++) + { + CHECK_VERIFICATION_ERROR(pass.pattern, res_data[i], i); + } + } + + return CL_SUCCESS; + } + + //-------------------------------------------------------------------------- + const cl_int pattern_pri = 0xB; + const cl_int pattern_sec = 0xC; + + bool properties_use_requested; + clEventWrapper user_event; +}; + +} // anonymous namespace + +int test_queue_substitution(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements) +{ + return MakeAndRunTest>( + device, context, queue, num_elements); +} + +int test_properties_queue_substitution(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements) +{ + return MakeAndRunTest>( + device, context, queue, num_elements); +} + +int test_simultaneous_queue_substitution(cl_device_id device, + cl_context context, + cl_command_queue queue, + int num_elements) +{ + return MakeAndRunTest>( + device, context, queue, num_elements); +} diff --git a/test_conformance/extensions/cl_khr_command_buffer/main.cpp b/test_conformance/extensions/cl_khr_command_buffer/main.cpp index 4dece455..338d4d4c 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/main.cpp +++ b/test_conformance/extensions/cl_khr_command_buffer/main.cpp @@ -15,11 +15,15 @@ #include "procs.h" #include "harness/testHarness.h" -test_definition test_list[] = { - ADD_TEST(single_ndrange), ADD_TEST(interleaved_enqueue), - ADD_TEST(mixed_commands), ADD_TEST(explicit_flush), - ADD_TEST(user_events), ADD_TEST(out_of_order) -}; +test_definition test_list[] = { ADD_TEST(single_ndrange), + ADD_TEST(interleaved_enqueue), + ADD_TEST(mixed_commands), + ADD_TEST(explicit_flush), + ADD_TEST(user_events), + ADD_TEST(out_of_order), + ADD_TEST(queue_substitution), + ADD_TEST(properties_queue_substitution), + ADD_TEST(simultaneous_queue_substitution) }; int main(int argc, const char *argv[]) diff --git a/test_conformance/extensions/cl_khr_command_buffer/procs.h b/test_conformance/extensions/cl_khr_command_buffer/procs.h index 58fd228f..f63d7847 100644 --- a/test_conformance/extensions/cl_khr_command_buffer/procs.h +++ b/test_conformance/extensions/cl_khr_command_buffer/procs.h @@ -31,5 +31,16 @@ extern int test_user_events(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements); extern int test_out_of_order(cl_device_id device, cl_context context, cl_command_queue queue, int num_elements); +extern int test_queue_substitution(cl_device_id device, cl_context context, + cl_command_queue queue, int num_elements); +extern int test_properties_queue_substitution(cl_device_id device, + cl_context context, + cl_command_queue queue, + int num_elements); +extern int test_simultaneous_queue_substitution(cl_device_id device, + cl_context context, + cl_command_queue queue, + int num_elements); + #endif /*_CL_KHR_COMMAND_BUFFER_PROCS_H*/