ahmed/OpenCL-CTS
1
0
Fork 0
mirror of https://github.com/KhronosGroup/OpenCL-CTS.git synced 2026-03-25 16:29:03 +00:00
Code Packages Projects Releases Wiki Activity

Modernization of tests from test_semaphores.cpp to align with new SemaphoreTestBase infrastructure (#2029)

Browse Source
This commit is contained in:
Marcin Hajder
2024-09-03 19:23:22 +02:00
committed by GitHub
parent 7131f87974
commit eb7a30ae42
1 changed files with 603 additions and 617 deletions
Download Patch File Download Diff File Expand all files Collapse all files

678
test_conformance/extensions/cl_khr_semaphore/test_semaphores.cpp
View File

@@ -16,13 +16,14 @@
#include "harness/typeWrappers.h" #include "harness/typeWrappers.h"
#include "harness/extensionHelpers.h"
#include "harness/errorHelpers.h" #include "harness/errorHelpers.h"
#include <system_error> #include <system_error>
#include <thread> #include <thread>
#include <chrono> #include <chrono>
#include <vector> #include <vector>
#include "semaphore_base.h"
#define FLUSH_DELAY_S 5 #define FLUSH_DELAY_S 5
#define SEMAPHORE_PARAM_TEST(param_name, param_type, expected) \ #define SEMAPHORE_PARAM_TEST(param_name, param_type, expected) \
@@ -30,8 +31,8 @@
{ \ { \
param_type value; \ param_type value; \
size_t size; \ size_t size; \
cl_int error = clGetSemaphoreInfoKHR(sema, param_name, sizeof(value), \ cl_int error = clGetSemaphoreInfoKHR(semaphore, param_name, \
&value, &size); \ sizeof(value), &value, &size); \
test_error(error, "Unable to get " #param_name " from semaphore"); \ test_error(error, "Unable to get " #param_name " from semaphore"); \
if (value != expected) \ if (value != expected) \
{ \ { \
@@ -54,8 +55,8 @@
{ \ { \
param_type value[num_params]; \ param_type value[num_params]; \
size_t size; \ size_t size; \
cl_int error = clGetSemaphoreInfoKHR(sema, param_name, sizeof(value), \ cl_int error = clGetSemaphoreInfoKHR(semaphore, param_name, \
&value, &size); \ sizeof(value), &value, &size); \
test_error(error, "Unable to get " #param_name " from semaphore"); \ test_error(error, "Unable to get " #param_name " from semaphore"); \
if (size != sizeof(value)) \ if (size != sizeof(value)) \
{ \ { \
@@ -70,114 +71,46 @@
} \ } \
} while (false) } while (false)
static const char* source = "__kernel void empty() {}"; namespace {
// Helper function that signals and waits on semaphore across two different const char* source = "__kernel void empty() {}";
// queues.
static int semaphore_cross_queue_helper(cl_device_id deviceID, struct SimpleSemaphore1 : public SemaphoreTestBase
cl_context context,
cl_command_queue queue_1,
cl_command_queue queue_2)
{ {
cl_int err; SimpleSemaphore1(cl_device_id device, cl_context context,
cl_command_queue queue)
: SemaphoreTestBase(device, context, queue)
{}
if (!is_extension_available(deviceID, "cl_khr_semaphore")) cl_int Run() override
{ {
log_info("cl_khr_semaphore is not supported on this platform. " cl_int err = CL_SUCCESS;
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
// Obtain pointers to semaphore's API
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR);
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR);
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR);
GET_PFN(deviceID, clReleaseSemaphoreKHR);
// Create semaphore
cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR),
0
};
cl_semaphore_khr sema =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore");
// Signal semaphore on queue_1
clEventWrapper signal_event;
err = clEnqueueSignalSemaphoresKHR(queue_1, 1, &sema, nullptr, 0, nullptr,
&signal_event);
test_error(err, "Could not signal semaphore");
// Wait semaphore on queue_2
clEventWrapper wait_event;
err = clEnqueueWaitSemaphoresKHR(queue_2, 1, &sema, nullptr, 0, nullptr,
&wait_event);
test_error(err, "Could not wait semaphore");
// Finish queue_1 and queue_2
err = clFinish(queue_1);
test_error(err, "Could not finish queue");
err = clFinish(queue_2);
test_error(err, "Could not finish queue");
// Ensure all events are completed
test_assert_event_complete(signal_event);
test_assert_event_complete(wait_event);
// Release semaphore
err = clReleaseSemaphoreKHR(sema);
test_error(err, "Could not release semaphore");
return TEST_PASS;
}
// Confirm that a signal followed by a wait will complete successfully
int test_semaphores_simple_1(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
cl_int err;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
}
// Obtain pointers to semaphore's API
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR);
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR);
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR);
GET_PFN(deviceID, clReleaseSemaphoreKHR);
// Create ooo queue // Create ooo queue
clCommandQueueWrapper queue = clCreateCommandQueue( clCommandQueueWrapper queue = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err); context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
// Create semaphore // Create semaphore
cl_semaphore_properties_khr sema_props[] = { cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
0 0
}; };
cl_semaphore_khr sema = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
// Signal semaphore // Signal semaphore
clEventWrapper signal_event; clEventWrapper signal_event;
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema, nullptr, 0, nullptr, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&signal_event); nullptr, &signal_event);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
// Wait semaphore // Wait semaphore
clEventWrapper wait_event; clEventWrapper wait_event;
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&wait_event); nullptr, &wait_event);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Finish // Finish
@@ -188,45 +121,33 @@ int test_semaphores_simple_1(cl_device_id deviceID, cl_context context,
test_assert_event_complete(signal_event); test_assert_event_complete(signal_event);
test_assert_event_complete(wait_event); test_assert_event_complete(wait_event);
// Release semaphore return CL_SUCCESS;
err = clReleaseSemaphoreKHR(sema);
test_error(err, "Could not release semaphore");
return TEST_PASS;
}
// Confirm that signal a semaphore with no event dependencies will not result
// in an implicit dependency on everything previously submitted
int test_semaphores_simple_2(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
cl_int err;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
} }
};
// Obtain pointers to semaphore's API struct SimpleSemaphore2 : public SemaphoreTestBase
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR); {
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR); SimpleSemaphore2(cl_device_id device, cl_context context,
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR); cl_command_queue queue)
GET_PFN(deviceID, clReleaseSemaphoreKHR); : SemaphoreTestBase(device, context, queue)
{}
cl_int Run() override
{
cl_int err = CL_SUCCESS;
// Create ooo queue // Create ooo queue
clCommandQueueWrapper queue = clCreateCommandQueue( clCommandQueueWrapper queue = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err); context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
// Create semaphore // Create semaphore
cl_semaphore_properties_khr sema_props[] = { cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
0 0
}; };
cl_semaphore_khr sema = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
@@ -237,8 +158,8 @@ int test_semaphores_simple_2(cl_device_id deviceID, cl_context context,
// Create Kernel // Create Kernel
clProgramWrapper program; clProgramWrapper program;
clKernelWrapper kernel; clKernelWrapper kernel;
err = create_single_kernel_helper(context, &program, &kernel, 1, &source, err = create_single_kernel_helper(context, &program, &kernel, 1,
"empty"); &source, "empty");
test_error(err, "Could not create kernel"); test_error(err, "Could not create kernel");
// Enqueue task_1 (dependency on user_event) // Enqueue task_1 (dependency on user_event)
@@ -248,14 +169,14 @@ int test_semaphores_simple_2(cl_device_id deviceID, cl_context context,
// Signal semaphore // Signal semaphore
clEventWrapper signal_event; clEventWrapper signal_event;
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema, nullptr, 0, nullptr, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&signal_event); nullptr, &signal_event);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
// Wait semaphore // Wait semaphore
clEventWrapper wait_event; clEventWrapper wait_event;
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&wait_event); nullptr, &wait_event);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Flush and delay // Flush and delay
@@ -281,52 +202,41 @@ int test_semaphores_simple_2(cl_device_id deviceID, cl_context context,
test_assert_event_complete(signal_event); test_assert_event_complete(signal_event);
test_assert_event_complete(wait_event); test_assert_event_complete(wait_event);
// Release semaphore return CL_SUCCESS;
err = clReleaseSemaphoreKHR(sema);
test_error(err, "Could not release semaphore");
return TEST_PASS;
}
// Confirm that a semaphore can be reused multiple times
int test_semaphores_reuse(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
cl_int err;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
} }
};
// Obtain pointers to semaphore's API struct SemaphoreReuse : public SemaphoreTestBase
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR); {
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR); SemaphoreReuse(cl_device_id device, cl_context context,
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR); cl_command_queue queue)
GET_PFN(deviceID, clReleaseSemaphoreKHR); : SemaphoreTestBase(device, context, queue)
{}
cl_int Run() override
{
cl_int err = CL_SUCCESS;
// Create ooo queue // Create ooo queue
clCommandQueueWrapper queue = clCreateCommandQueue( clCommandQueueWrapper queue = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err); context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
// Create semaphore // Create semaphore
cl_semaphore_properties_khr sema_props[] = { cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
0 0
}; };
cl_semaphore_khr sema = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
// Create Kernel // Create Kernel
clProgramWrapper program; clProgramWrapper program;
clKernelWrapper kernel; clKernelWrapper kernel;
err = create_single_kernel_helper(context, &program, &kernel, 1, &source, err = create_single_kernel_helper(context, &program, &kernel, 1,
"empty"); &source, "empty");
test_error(err, "Could not create kernel"); test_error(err, "Could not create kernel");
constexpr size_t loop_count = 10; constexpr size_t loop_count = 10;
@@ -339,7 +249,7 @@ int test_semaphores_reuse(cl_device_id deviceID, cl_context context,
test_error(err, "Unable to enqueue task_1"); test_error(err, "Unable to enqueue task_1");
// Signal semaphore (dependency on task_1) // Signal semaphore (dependency on task_1)
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema, nullptr, 1, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore, nullptr, 1,
&task_events[0], &signal_events[0]); &task_events[0], &signal_events[0]);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
@@ -348,8 +258,8 @@ int test_semaphores_reuse(cl_device_id deviceID, cl_context context,
for (loop = 1; loop < loop_count; ++loop) for (loop = 1; loop < loop_count; ++loop)
{ {
// Wait semaphore // Wait semaphore
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&wait_events[loop - 1]); nullptr, &wait_events[loop - 1]);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Enqueue task_loop (dependency on wait) // Enqueue task_loop (dependency on wait)
@@ -362,15 +272,15 @@ int test_semaphores_reuse(cl_device_id deviceID, cl_context context,
test_error(err, "Unable to wait for wait semaphore to complete"); test_error(err, "Unable to wait for wait semaphore to complete");
// Signal semaphore (dependency on task_loop) // Signal semaphore (dependency on task_loop)
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema, nullptr, 1, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore, nullptr, 1,
&task_events[loop], &task_events[loop],
&signal_events[loop]); &signal_events[loop]);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
} }
// Wait semaphore // Wait semaphore
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&wait_events[loop - 1]); nullptr, &wait_events[loop - 1]);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Finish // Finish
@@ -385,107 +295,147 @@ int test_semaphores_reuse(cl_device_id deviceID, cl_context context,
test_assert_event_complete(task_events[loop]); test_assert_event_complete(task_events[loop]);
} }
// Release semaphore return CL_SUCCESS;
err = clReleaseSemaphoreKHR(sema); }
test_error(err, "Could not release semaphore"); };
template <bool in_order> struct SemaphoreCrossQueue : public SemaphoreTestBase
{
SemaphoreCrossQueue(cl_device_id device, cl_context context,
cl_command_queue queue)
: SemaphoreTestBase(device, context, queue)
{}
// Helper function that signals and waits on semaphore across two different
// queues.
int semaphore_cross_queue_helper(cl_device_id deviceID, cl_context context,
cl_command_queue queue_1,
cl_command_queue queue_2)
{
cl_int err = CL_SUCCESS;
// Create semaphore
cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
0
};
semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore");
// Signal semaphore on queue_1
clEventWrapper signal_event;
err = clEnqueueSignalSemaphoresKHR(queue_1, 1, semaphore, nullptr, 0,
nullptr, &signal_event);
test_error(err, "Could not signal semaphore");
// Wait semaphore on queue_2
clEventWrapper wait_event;
err = clEnqueueWaitSemaphoresKHR(queue_2, 1, semaphore, nullptr, 0,
nullptr, &wait_event);
test_error(err, "Could not wait semaphore");
// Finish queue_1 and queue_2
err = clFinish(queue_1);
test_error(err, "Could not finish queue");
err = clFinish(queue_2);
test_error(err, "Could not finish queue");
// Ensure all events are completed
test_assert_event_complete(signal_event);
test_assert_event_complete(wait_event);
return TEST_PASS; return TEST_PASS;
} }
// Confirm that a semaphore works across different ooo queues
int test_semaphores_cross_queues_ooo(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
cl_int err;
// Create ooo queues
clCommandQueueWrapper queue_1 = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue");
clCommandQueueWrapper queue_2 = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue");
return semaphore_cross_queue_helper(deviceID, context, queue_1, queue_2);
}
// Confirm that a semaphore works across different in-order queues
int test_semaphores_cross_queues_io(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
cl_int err;
cl_int run_in_order()
{
cl_int err = CL_SUCCESS;
// Create in-order queues // Create in-order queues
clCommandQueueWrapper queue_1 = clCommandQueueWrapper queue_1 =
clCreateCommandQueue(context, deviceID, 0, &err); clCreateCommandQueue(context, device, 0, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
clCommandQueueWrapper queue_2 = clCommandQueueWrapper queue_2 =
clCreateCommandQueue(context, deviceID, 0, &err); clCreateCommandQueue(context, device, 0, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
return semaphore_cross_queue_helper(deviceID, context, queue_1, queue_2); return semaphore_cross_queue_helper(device, context, queue_1, queue_2);
}
// Confirm that we can signal multiple semaphores with one command
int test_semaphores_multi_signal(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
cl_int err;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
} }
// Obtain pointers to semaphore's API cl_int run_out_of_order()
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR); {
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR); cl_int err = CL_SUCCESS;
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR); // Create ooo queues
GET_PFN(deviceID, clReleaseSemaphoreKHR); clCommandQueueWrapper queue_1 = clCreateCommandQueue(
context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue");
clCommandQueueWrapper queue_2 = clCreateCommandQueue(
context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue");
return semaphore_cross_queue_helper(device, context, queue_1, queue_2);
}
cl_int Run() override
{
if (in_order)
return run_in_order();
else
return run_out_of_order();
}
};
struct SemaphoreMultiSignal : public SemaphoreTestBase
{
SemaphoreMultiSignal(cl_device_id device, cl_context context,
cl_command_queue queue)
: SemaphoreTestBase(device, context, queue), semaphore_second(this)
{}
cl_int Run() override
{
cl_int err = CL_SUCCESS;
// Create ooo queue // Create ooo queue
clCommandQueueWrapper queue = clCreateCommandQueue( clCommandQueueWrapper queue = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err); context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
// Create semaphore // Create semaphore
cl_semaphore_properties_khr sema_props[] = { cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
0 0
}; };
cl_semaphore_khr sema_1 = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
cl_semaphore_khr sema_2 = semaphore_second =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
// Signal semaphore 1 and 2 // Signal semaphore 1 and 2
clEventWrapper signal_event; clEventWrapper signal_event;
cl_semaphore_khr sema_list[] = { sema_1, sema_2 }; cl_semaphore_khr sema_list[] = { semaphore, semaphore_second };
err = clEnqueueSignalSemaphoresKHR(queue, 2, sema_list, nullptr, 0, nullptr, err = clEnqueueSignalSemaphoresKHR(queue, 2, sema_list, nullptr, 0,
&signal_event); nullptr, &signal_event);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
// Wait semaphore 1 // Wait semaphore 1
clEventWrapper wait_1_event; clEventWrapper wait_1_event;
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema_1, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&wait_1_event); nullptr, &wait_1_event);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Wait semaphore 2 // Wait semaphore 2
clEventWrapper wait_2_event; clEventWrapper wait_2_event;
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema_2, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore_second, nullptr, 0,
&wait_2_event); nullptr, &wait_2_event);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Finish // Finish
@@ -497,71 +447,58 @@ int test_semaphores_multi_signal(cl_device_id deviceID, cl_context context,
test_assert_event_complete(wait_1_event); test_assert_event_complete(wait_1_event);
test_assert_event_complete(wait_2_event); test_assert_event_complete(wait_2_event);
// Release semaphores return CL_SUCCESS;
err = clReleaseSemaphoreKHR(sema_1);
test_error(err, "Could not release semaphore");
err = clReleaseSemaphoreKHR(sema_2);
test_error(err, "Could not release semaphore");
return TEST_PASS;
}
// Confirm that we can wait for multiple semaphores with one command
int test_semaphores_multi_wait(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
cl_int err;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
} }
clSemaphoreWrapper semaphore_second = nullptr;
};
// Obtain pointers to semaphore's API struct SemaphoreMultiWait : public SemaphoreTestBase
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR); {
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR); SemaphoreMultiWait(cl_device_id device, cl_context context,
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR); cl_command_queue queue)
GET_PFN(deviceID, clReleaseSemaphoreKHR); : SemaphoreTestBase(device, context, queue), semaphore_second(this)
{}
cl_int Run() override
{
cl_int err = CL_SUCCESS;
// Create ooo queue // Create ooo queue
clCommandQueueWrapper queue = clCreateCommandQueue( clCommandQueueWrapper queue = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err); context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
// Create semaphores // Create semaphores
cl_semaphore_properties_khr sema_props[] = { cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
0 0
}; };
cl_semaphore_khr sema_1 = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
cl_semaphore_khr sema_2 = semaphore_second =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
// Signal semaphore 1 // Signal semaphore 1
clEventWrapper signal_1_event; clEventWrapper signal_1_event;
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema_1, nullptr, 0, nullptr, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&signal_1_event); nullptr, &signal_1_event);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
// Signal semaphore 2 // Signal semaphore 2
clEventWrapper signal_2_event; clEventWrapper signal_2_event;
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema_2, nullptr, 0, nullptr, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore_second, nullptr,
&signal_2_event); 0, nullptr, &signal_2_event);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
// Wait semaphore 1 and 2 // Wait semaphore 1 and 2
clEventWrapper wait_event; clEventWrapper wait_event;
cl_semaphore_khr sema_list[] = { sema_1, sema_2 }; cl_semaphore_khr sema_list[] = { semaphore, semaphore_second };
err = clEnqueueWaitSemaphoresKHR(queue, 2, sema_list, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 2, sema_list, nullptr, 0,
&wait_event); nullptr, &wait_event);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Finish // Finish
@@ -573,46 +510,33 @@ int test_semaphores_multi_wait(cl_device_id deviceID, cl_context context,
test_assert_event_complete(signal_2_event); test_assert_event_complete(signal_2_event);
test_assert_event_complete(wait_event); test_assert_event_complete(wait_event);
// Release semaphores return CL_SUCCESS;
err = clReleaseSemaphoreKHR(sema_1);
test_error(err, "Could not release semaphore");
err = clReleaseSemaphoreKHR(sema_2);
test_error(err, "Could not release semaphore");
return TEST_PASS;
}
// Confirm the semaphores can be successfully queried
int test_semaphores_queries(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
cl_int err = CL_SUCCESS;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
} }
clSemaphoreWrapper semaphore_second = nullptr;
};
// Obtain pointers to semaphore's API struct SemaphoreQueries : public SemaphoreTestBase
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR); {
GET_PFN(deviceID, clGetSemaphoreInfoKHR); SemaphoreQueries(cl_device_id device, cl_context context,
GET_PFN(deviceID, clRetainSemaphoreKHR); cl_command_queue queue)
GET_PFN(deviceID, clReleaseSemaphoreKHR); : SemaphoreTestBase(device, context, queue)
{}
cl_int Run() override
{
cl_int err = CL_SUCCESS;
// Create binary semaphore // Create binary semaphore
cl_semaphore_properties_khr sema_props[] = { cl_semaphore_properties_khr sema_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
static_cast<cl_semaphore_properties_khr>( static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR), CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR),
(cl_semaphore_properties_khr)deviceID, (cl_semaphore_properties_khr)device,
CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR, CL_SEMAPHORE_DEVICE_HANDLE_LIST_END_KHR,
0 0
}; };
cl_semaphore_khr sema = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
@@ -621,25 +545,26 @@ int test_semaphores_queries(cl_device_id deviceID, cl_context context,
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_TYPE_KHR, cl_semaphore_type_khr, SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_TYPE_KHR, cl_semaphore_type_khr,
CL_SEMAPHORE_TYPE_BINARY_KHR); CL_SEMAPHORE_TYPE_BINARY_KHR);
// Confirm that querying CL_SEMAPHORE_CONTEXT_KHR returns the right context // Confirm that querying CL_SEMAPHORE_CONTEXT_KHR returns the right
// context
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_CONTEXT_KHR, cl_context, context); SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_CONTEXT_KHR, cl_context, context);
// Confirm that querying CL_SEMAPHORE_REFERENCE_COUNT_KHR returns the right // Confirm that querying CL_SEMAPHORE_REFERENCE_COUNT_KHR returns the
// value // right value
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_REFERENCE_COUNT_KHR, cl_uint, 1);
err = clRetainSemaphoreKHR(semaphore);
test_error(err, "Could not retain semaphore");
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_REFERENCE_COUNT_KHR, cl_uint, 2);
err = clReleaseSemaphoreKHR(semaphore);
test_error(err, "Could not release semaphore");
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_REFERENCE_COUNT_KHR, cl_uint, 1); SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_REFERENCE_COUNT_KHR, cl_uint, 1);
// Confirm that querying CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR returns the // Confirm that querying CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR returns the
// same device id the semaphore was created with // same device id the semaphore was created with
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR, cl_device_id, SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_DEVICE_HANDLE_LIST_KHR, cl_device_id,
deviceID); device);
err = clRetainSemaphoreKHR(sema);
test_error(err, "Could not retain semaphore");
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_REFERENCE_COUNT_KHR, cl_uint, 2);
err = clReleaseSemaphoreKHR(sema);
test_error(err, "Could not release semaphore");
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_REFERENCE_COUNT_KHR, cl_uint, 1);
// Confirm that querying CL_SEMAPHORE_PROPERTIES_KHR returns the same // Confirm that querying CL_SEMAPHORE_PROPERTIES_KHR returns the same
// properties the semaphore was created with // properties the semaphore was created with
@@ -648,52 +573,42 @@ int test_semaphores_queries(cl_device_id deviceID, cl_context context,
// Confirm that querying CL_SEMAPHORE_PAYLOAD_KHR returns the unsignaled // Confirm that querying CL_SEMAPHORE_PAYLOAD_KHR returns the unsignaled
// state // state
SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_PAYLOAD_KHR, cl_semaphore_payload_khr, 0); SEMAPHORE_PARAM_TEST(CL_SEMAPHORE_PAYLOAD_KHR, cl_semaphore_payload_khr,
0);
err = clReleaseSemaphoreKHR(sema); return CL_SUCCESS;
test_error(err, "Could not release semaphore");
return TEST_PASS;
}
// Test it is possible to export a semaphore to a sync fd and import the same
// sync fd to a new semaphore
int test_semaphores_import_export_fd(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
cl_int err;
if (!is_extension_available(deviceID, "cl_khr_semaphore"))
{
log_info("cl_khr_semaphore is not supported on this platform. "
"Skipping test.\n");
return TEST_SKIPPED_ITSELF;
} }
};
if (!is_extension_available(deviceID, "cl_khr_external_semaphore_sync_fd")) struct SemaphoreImportExportFD : public SemaphoreTestBase
{
SemaphoreImportExportFD(cl_device_id device, cl_context context,
cl_command_queue queue)
: SemaphoreTestBase(device, context, queue), semaphore_second(this)
{}
cl_int Run() override
{ {
log_info("cl_khr_external_semaphore_sync_fd is not supported on this " cl_int err = CL_SUCCESS;
if (!is_extension_available(device,
"cl_khr_external_semaphore_sync_fd"))
{
log_info(
"cl_khr_external_semaphore_sync_fd is not supported on this "
"platform. Skipping test.\n"); "platform. Skipping test.\n");
return TEST_SKIPPED_ITSELF; return TEST_SKIPPED_ITSELF;
} }
// Obtain pointers to semaphore's API
GET_PFN(deviceID, clCreateSemaphoreWithPropertiesKHR);
GET_PFN(deviceID, clEnqueueSignalSemaphoresKHR);
GET_PFN(deviceID, clEnqueueWaitSemaphoresKHR);
GET_PFN(deviceID, clGetSemaphoreHandleForTypeKHR);
GET_PFN(deviceID, clReleaseSemaphoreKHR);
// Create ooo queue // Create ooo queue
clCommandQueueWrapper queue = clCreateCommandQueue( clCommandQueueWrapper queue = clCreateCommandQueue(
context, deviceID, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err); context, device, CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE, &err);
test_error(err, "Could not create command queue"); test_error(err, "Could not create command queue");
// Create semaphore // Create semaphore
cl_semaphore_properties_khr sema_1_props[] = { cl_semaphore_properties_khr sema_1_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
static_cast<cl_semaphore_properties_khr>( static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR), CL_SEMAPHORE_EXPORT_HANDLE_TYPES_KHR),
static_cast<cl_semaphore_properties_khr>( static_cast<cl_semaphore_properties_khr>(
@@ -702,22 +617,22 @@ int test_semaphores_import_export_fd(cl_device_id deviceID, cl_context context,
CL_SEMAPHORE_EXPORT_HANDLE_TYPES_LIST_END_KHR), CL_SEMAPHORE_EXPORT_HANDLE_TYPES_LIST_END_KHR),
0 0
}; };
cl_semaphore_khr sema_1 = semaphore =
clCreateSemaphoreWithPropertiesKHR(context, sema_1_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_1_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
// Signal semaphore // Signal semaphore
clEventWrapper signal_event; clEventWrapper signal_event;
err = clEnqueueSignalSemaphoresKHR(queue, 1, &sema_1, nullptr, 0, nullptr, err = clEnqueueSignalSemaphoresKHR(queue, 1, semaphore, nullptr, 0,
&signal_event); nullptr, &signal_event);
test_error(err, "Could not signal semaphore"); test_error(err, "Could not signal semaphore");
// Extract sync fd // Extract sync fd
int handle = -1; int handle = -1;
size_t handle_size; size_t handle_size;
err = clGetSemaphoreHandleForTypeKHR(sema_1, deviceID, err = clGetSemaphoreHandleForTypeKHR(
CL_SEMAPHORE_HANDLE_SYNC_FD_KHR, semaphore, device, CL_SEMAPHORE_HANDLE_SYNC_FD_KHR, sizeof(handle),
sizeof(handle), &handle, &handle_size); &handle, &handle_size);
test_error(err, "Could not extract semaphore handle"); test_error(err, "Could not extract semaphore handle");
test_assert_error(sizeof(handle) == handle_size, "Invalid handle size"); test_assert_error(sizeof(handle) == handle_size, "Invalid handle size");
test_assert_error(handle >= 0, "Invalid handle"); test_assert_error(handle >= 0, "Invalid handle");
@@ -725,19 +640,20 @@ int test_semaphores_import_export_fd(cl_device_id deviceID, cl_context context,
// Create semaphore from sync fd // Create semaphore from sync fd
cl_semaphore_properties_khr sema_2_props[] = { cl_semaphore_properties_khr sema_2_props[] = {
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR), static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_KHR),
static_cast<cl_semaphore_properties_khr>(CL_SEMAPHORE_TYPE_BINARY_KHR), static_cast<cl_semaphore_properties_khr>(
CL_SEMAPHORE_TYPE_BINARY_KHR),
CL_SEMAPHORE_HANDLE_SYNC_FD_KHR, CL_SEMAPHORE_HANDLE_SYNC_FD_KHR,
static_cast<cl_semaphore_properties_khr>(handle), 0 static_cast<cl_semaphore_properties_khr>(handle), 0
}; };
cl_semaphore_khr sema_2 = semaphore_second =
clCreateSemaphoreWithPropertiesKHR(context, sema_2_props, &err); clCreateSemaphoreWithPropertiesKHR(context, sema_2_props, &err);
test_error(err, "Could not create semaphore"); test_error(err, "Could not create semaphore");
// Wait semaphore // Wait semaphore
clEventWrapper wait_event; clEventWrapper wait_event;
err = clEnqueueWaitSemaphoresKHR(queue, 1, &sema_2, nullptr, 0, nullptr, err = clEnqueueWaitSemaphoresKHR(queue, 1, semaphore_second, nullptr, 0,
&wait_event); nullptr, &wait_event);
test_error(err, "Could not wait semaphore"); test_error(err, "Could not wait semaphore");
// Finish // Finish
@@ -748,11 +664,81 @@ int test_semaphores_import_export_fd(cl_device_id deviceID, cl_context context,
test_assert_event_complete(signal_event); test_assert_event_complete(signal_event);
test_assert_event_complete(wait_event); test_assert_event_complete(wait_event);
// Release semaphore return CL_SUCCESS;
err = clReleaseSemaphoreKHR(sema_1); }
test_error(err, "Could not release semaphore"); clSemaphoreWrapper semaphore_second = nullptr;
};
} // anonymous namespace
err = clReleaseSemaphoreKHR(sema_2); // Confirm that a signal followed by a wait will complete successfully
test_error(err, "Could not release semaphore"); int test_semaphores_simple_1(cl_device_id deviceID, cl_context context,
return TEST_PASS; cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<SimpleSemaphore1>(deviceID, context, defaultQueue);
}
// Confirm that signal a semaphore with no event dependencies will not result
// in an implicit dependency on everything previously submitted
int test_semaphores_simple_2(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<SimpleSemaphore2>(deviceID, context, defaultQueue);
}
// Confirm that a semaphore can be reused multiple times
int test_semaphores_reuse(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<SemaphoreReuse>(deviceID, context, defaultQueue);
}
// Confirm that a semaphore works across different ooo queues
int test_semaphores_cross_queues_ooo(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<SemaphoreCrossQueue<false>>(deviceID, context,
defaultQueue);
}
// Confirm that a semaphore works across different in-order queues
int test_semaphores_cross_queues_io(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<SemaphoreCrossQueue<true>>(deviceID, context,
defaultQueue);
}
// Confirm that we can signal multiple semaphores with one command
int test_semaphores_multi_signal(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<SemaphoreMultiSignal>(deviceID, context,
defaultQueue);
}
// Confirm that we can wait for multiple semaphores with one command
int test_semaphores_multi_wait(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<SemaphoreMultiWait>(deviceID, context, defaultQueue);
}
// Confirm the semaphores can be successfully queried
int test_semaphores_queries(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue, int num_elements)
{
return MakeAndRunTest<SemaphoreQueries>(deviceID, context, defaultQueue);
}
// Test it is possible to export a semaphore to a sync fd and import the same
// sync fd to a new semaphore
int test_semaphores_import_export_fd(cl_device_id deviceID, cl_context context,
cl_command_queue defaultQueue,
int num_elements)
{
return MakeAndRunTest<SemaphoreImportExportFD>(deviceID, context,
defaultQueue);
} }