diff --git a/test_conformance/extensions/cl_khr_semaphore/CMakeLists.txt b/test_conformance/extensions/cl_khr_semaphore/CMakeLists.txt index 682ada5f..6f4ac812 100644 --- a/test_conformance/extensions/cl_khr_semaphore/CMakeLists.txt +++ b/test_conformance/extensions/cl_khr_semaphore/CMakeLists.txt @@ -9,6 +9,7 @@ set(${MODULE_NAME}_SOURCES test_semaphores_negative_create.cpp test_semaphores_cross_queue.cpp test_semaphores_queries.cpp + test_semaphores_payload.cpp semaphore_base.h ) diff --git a/test_conformance/extensions/cl_khr_semaphore/test_semaphores.cpp b/test_conformance/extensions/cl_khr_semaphore/test_semaphores.cpp index ce146b41..f4eaf5d1 100644 --- a/test_conformance/extensions/cl_khr_semaphore/test_semaphores.cpp +++ b/test_conformance/extensions/cl_khr_semaphore/test_semaphores.cpp @@ -14,14 +14,8 @@ // limitations under the License. // -#include - #include "semaphore_base.h" -#include "semaphore_base.h" - -#define FLUSH_DELAY_S 5 - namespace { const char* source = "__kernel void empty() {}"; diff --git a/test_conformance/extensions/cl_khr_semaphore/test_semaphores_payload.cpp b/test_conformance/extensions/cl_khr_semaphore/test_semaphores_payload.cpp new file mode 100644 index 00000000..94fb82e0 --- /dev/null +++ b/test_conformance/extensions/cl_khr_semaphore/test_semaphores_payload.cpp @@ -0,0 +1,86 @@ +// +// 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. +// 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 "semaphore_base.h" + +namespace { + +struct PayloadSemaphore : public SemaphoreTestBase +{ + clSemaphoreWrapper sema_sec; + + PayloadSemaphore(cl_device_id device, cl_context context, + cl_command_queue queue, cl_int nelems) + : SemaphoreTestBase(device, context, queue, nelems), sema_sec(this) + {} + + cl_int Run() override + { + cl_int err = CL_SUCCESS; + + // Create semaphore + cl_semaphore_properties_khr sema_props[] = { + static_cast(CL_SEMAPHORE_TYPE_KHR), + static_cast( + CL_SEMAPHORE_TYPE_BINARY_KHR), + 0 + }; + semaphore = + clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); + test_error(err, "Could not create semaphore"); + + sema_sec = + clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); + test_error(err, "Could not create semaphore"); + + { + cl_semaphore_payload_khr payload_list[] = { 1, 2 }; + cl_semaphore_khr semaphores[2] = { semaphore, sema_sec }; + + // Signal semaphore + err = clEnqueueSignalSemaphoresKHR( + queue, 2, semaphores, payload_list, 0, nullptr, nullptr); + test_error(err, "Could not signal semaphore"); + } + + { + cl_semaphore_payload_khr payload_list[] = { 3, 4 }; + cl_semaphore_khr semaphores[2] = { semaphore, sema_sec }; + + // Wait semaphore + err = clEnqueueWaitSemaphoresKHR(queue, 2, semaphores, payload_list, + 0, nullptr, nullptr); + test_error(err, "Could not wait semaphore"); + } + + // Finish + err = clFinish(queue); + test_error(err, "Could not finish queue"); + + return CL_SUCCESS; + } +}; + + +} // anonymous namespace + +// Confirm that a valid semaphore payload values list will be ignored if no +// semaphores in the list of sema_objects require a payload +REGISTER_TEST_VERSION(semaphores_payload, Version(1, 2)) +{ + return MakeAndRunTest(device, context, queue, + num_elements); +}