[Piglit] [PATCH 1/2] cl: add support for global offset
Serge Martin
edb+piglit at sigluy.net
Tue May 17 17:25:16 UTC 2016
On Sunday 15 May 2016 17:39:24 Jan Vesely wrote:
> Signed-off-by: Jan Vesely <jan.vesely at rutgers.edu>
Review-by: Serge Martin <edb+piglit at sigluy.net>
> ---
> tests/cl/api/create-program-with-binary.c | 2 +-
> tests/cl/custom/buffer-flags.c | 2 +-
> tests/cl/custom/flush-after-enqueue-kernel.c | 2 +-
> tests/cl/custom/r600-create-release-buffer-bug.c | 12 ++++-----
> tests/cl/custom/run-simple-kernel.c | 2 +-
> tests/cl/custom/use-sub-buffer-in-kernel.c | 4 +--
> tests/cl/program/bitcoin-phatk.c | 2 +-
> tests/cl/program/max-work-item-sizes.c | 1 +
> tests/cl/program/program-tester.c | 33
> ++++++++++++++++++++++++ tests/util/piglit-util-cl.c |
> 7 +++--
> tests/util/piglit-util-cl.h | 4 +++
> 11 files changed, 56 insertions(+), 15 deletions(-)
>
> diff --git a/tests/cl/api/create-program-with-binary.c
> b/tests/cl/api/create-program-with-binary.c index 4873b76..4238331 100644
> --- a/tests/cl/api/create-program-with-binary.c
> +++ b/tests/cl/api/create-program-with-binary.c
> @@ -183,7 +183,7 @@ piglit_cl_test(const int argc,
> size_t global_work_size = 1;
> size_t local_work_size = 1;
> cl_command_queue queue = ctx->command_queues[i];
> - if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1,
> + if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, NULL,
> &global_work_size, &local_work_size)) {
> fprintf(stderr, "Failed to execute binary kernel.");
> piglit_merge_result(&result, PIGLIT_FAIL);
> diff --git a/tests/cl/custom/buffer-flags.c b/tests/cl/custom/buffer-flags.c
> index 6bf988f..115fb92 100644
> --- a/tests/cl/custom/buffer-flags.c
> +++ b/tests/cl/custom/buffer-flags.c
> @@ -130,7 +130,7 @@ buffer_test(piglit_cl_context *ctx,
>
> printf("Running the kernel...\n");
> if (!piglit_cl_enqueue_ND_range_kernel(context->command_queues[0],
> - kernel, 1, &global, &local)) {
> + kernel, 1, NULL, &global, &local)) {
> ret = PIGLIT_FAIL;
> goto cleanup;
> }
> diff --git a/tests/cl/custom/flush-after-enqueue-kernel.c
> b/tests/cl/custom/flush-after-enqueue-kernel.c index 00a0516..8d3fc8c
> 100644
> --- a/tests/cl/custom/flush-after-enqueue-kernel.c
> +++ b/tests/cl/custom/flush-after-enqueue-kernel.c
> @@ -92,7 +92,7 @@ piglit_cl_test(const int argc,
> }
>
> if (!piglit_cl_enqueue_ND_range_kernel(context->command_queues[0],
> - kernel, 3, global_size, local_size)) {
> + kernel, 3, NULL, global_size, local_size)) {
> return PIGLIT_FAIL;
> }
>
> diff --git a/tests/cl/custom/r600-create-release-buffer-bug.c
> b/tests/cl/custom/r600-create-release-buffer-bug.c index 7c3a214..535e57d
> 100644
> --- a/tests/cl/custom/r600-create-release-buffer-bug.c
> +++ b/tests/cl/custom/r600-create-release-buffer-bug.c
> @@ -74,8 +74,8 @@ piglit_cl_test(const int argc,
> return PIGLIT_FAIL;
> }
>
> - if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, &global_size,
> - &local_size)) {
> + if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, NULL,
> + &global_size, &local_size)) {
> return PIGLIT_FAIL;
> }
>
> @@ -84,8 +84,8 @@ piglit_cl_test(const int argc,
> return PIGLIT_FAIL;
> }
>
> - if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, &global_size,
> - &local_size)) {
> + if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, NULL,
> + &global_size, &local_size)) {
> return PIGLIT_FAIL;
> }
>
> @@ -98,8 +98,8 @@ piglit_cl_test(const int argc,
> return PIGLIT_FAIL;
> }
>
> - if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, &global_size,
> - &local_size)) {
> + if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, NULL,
> + &global_size, &local_size)) {
> return PIGLIT_FAIL;
> }
>
> diff --git a/tests/cl/custom/run-simple-kernel.c
> b/tests/cl/custom/run-simple-kernel.c index 48d7725..db529d2 100644
> --- a/tests/cl/custom/run-simple-kernel.c
> +++ b/tests/cl/custom/run-simple-kernel.c
> @@ -61,7 +61,7 @@ piglit_cl_test(const int argc,
> sizeof(cl_int), &data);
> piglit_cl_set_kernel_buffer_arg(kernel, 0, &buffer);
> piglit_cl_execute_ND_range_kernel(context->command_queues[0], kernel, 1,
> - &global_size, &local_size);
> + NULL, &global_size, &local_size);
>
> /* Read the buffer and check the result */
> piglit_cl_read_buffer(context->command_queues[0], buffer, 0,
> diff --git a/tests/cl/custom/use-sub-buffer-in-kernel.c
> b/tests/cl/custom/use-sub-buffer-in-kernel.c index 10cfe01..74c0858 100644
> --- a/tests/cl/custom/use-sub-buffer-in-kernel.c
> +++ b/tests/cl/custom/use-sub-buffer-in-kernel.c
> @@ -98,8 +98,8 @@ piglit_cl_test(const int argc,
> return PIGLIT_FAIL;
> }
>
> - if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, &global_size,
> - &local_size)) {
> + if (!piglit_cl_enqueue_ND_range_kernel(queue, kernel, 1, NULL,
> + &global_size, &local_size)) {
> return PIGLIT_FAIL;
> }
> clFinish(queue);
> diff --git a/tests/cl/program/bitcoin-phatk.c
> b/tests/cl/program/bitcoin-phatk.c index 422a3ce..f57d41b 100644
> --- a/tests/cl/program/bitcoin-phatk.c
> +++ b/tests/cl/program/bitcoin-phatk.c
> @@ -514,7 +514,7 @@ piglit_cl_test(const int argc,
> piglit_cl_set_kernel_buffer_arg(env->kernel, 23, &buffer);
>
> piglit_cl_execute_ND_range_kernel(env->context->command_queues[0],
> - env->kernel, 1,
> + env->kernel, 1, NULL,
> &global_size, &local_size);
> piglit_cl_read_whole_buffer(env->context->command_queues[0], buffer,
> data);
>
> diff --git a/tests/cl/program/max-work-item-sizes.c
> b/tests/cl/program/max-work-item-sizes.c index 6fdd18e..5f160a5 100644
> --- a/tests/cl/program/max-work-item-sizes.c
> +++ b/tests/cl/program/max-work-item-sizes.c
> @@ -116,6 +116,7 @@ piglit_cl_test(const int argc,
> piglit_cl_execute_ND_range_kernel(env->context->command_queues[0],
> env->kernel,
> i+1,
> + NULL,
> global_size,
> global_size);
>
> diff --git a/tests/cl/program/program-tester.c
> b/tests/cl/program/program-tester.c index 0a3b011..8e3cb4d 100644
> --- a/tests/cl/program/program-tester.c
> +++ b/tests/cl/program/program-tester.c
> @@ -279,7 +279,9 @@ bool expect_test_fail = false;
> cl_uint work_dimensions = 1;
> size_t global_work_size[3] = {1, 1, 1};
> size_t local_work_size[3] = {1, 1, 1};
> +size_t global_offset[3] = {0, 0, 0};
> bool local_work_size_null = false;
> +bool global_offset_null = true;
>
> /* Helper functions */
>
> @@ -384,6 +386,8 @@ struct test {
> size_t global_work_size[3];
> size_t local_work_size[3];
> bool local_work_size_null;
> + size_t global_offset[3];
> + bool global_offset_null;
>
> bool expect_test_fail;
>
> @@ -407,6 +411,7 @@ struct test create_test()
> //.global_work_size = global_work_size,
> //.local_work_size = local_work_size,
> .local_work_size_null = local_work_size_null,
> + .global_offset_null = global_offset_null,
>
> .expect_test_fail = expect_test_fail,
>
> @@ -419,6 +424,7 @@ struct test create_test()
>
> memcpy(t.global_work_size, global_work_size, sizeof(global_work_size));
> memcpy(t.local_work_size, local_work_size, sizeof(local_work_size));
> + memcpy(t.global_offset, global_offset, sizeof(global_offset));
>
> return t;
> }
> @@ -1736,6 +1742,19 @@ parse_config(const char* config_str,
> } else {
> local_work_size_null = true;
> }
> + } else if(regex_match(key, "^global_offset$")) {
> + if(!regex_match(value, REGEX_FULL_MATCH(REGEX_NULL))) {
> + int i;
> + uint64_t* int_global_offset;
> + get_uint_array(value, &int_global_offset, 3);
> + for(i = 0; i < 3; i++) {
> + global_offset[i] = int_global_offset[i];
> + }
> + global_offset_null = false;
> + free(int_global_offset);
> + } else {
> + global_offset_null = true;
> + }
> } else {
> fprintf(stderr,
> "Invalid configuration, key '%s' does not belong
to a [config]
> section: %s\n", @@ -1778,6 +1797,19 @@ parse_config(const char* config_str,
> } else {
> test->local_work_size_null = true;
> }
> + } else if(regex_match(key, "^global_offset$")) {
> + if(!regex_match(value, REGEX_FULL_MATCH(REGEX_NULL))) {
> + int i;
> + uint64_t* int_global_offset;
> + get_uint_array(value, &int_global_offset, 3);
> + for(i = 0; i < 3; i++) {
> + test->global_offset[i] = int_global_offset[i];
> + }
> + test->global_offset_null = false;
> + free(int_global_offset);
> + } else {
> + test->global_offset_null = true;
> + }
> } else if(regex_match(key, "^arg_in$")) {
> get_test_arg(value, test, true);
> } else if(regex_match(key, "^arg_out$")) {
> @@ -2396,6 +2428,7 @@ test_kernel(const struct
> piglit_cl_program_test_config* config,
> if(!piglit_cl_execute_ND_range_kernel(env->context->command_queues[0],
> kernel,
> test.work_dimensions,
> + test.global_offset_null ? NULL :
> test.global_offset, test.global_work_size,
> test.local_work_size_null ? NULL :
> test.local_work_size)) { printf("Failed to enqueue the kernel\n");
> diff --git a/tests/util/piglit-util-cl.c b/tests/util/piglit-util-cl.c
> index 51d6808..7e7d985 100644
> --- a/tests/util/piglit-util-cl.c
> +++ b/tests/util/piglit-util-cl.c
> @@ -1263,14 +1263,15 @@ piglit_cl_set_kernel_buffer_arg(cl_kernel kernel,
> cl_uint arg_index, bool
> piglit_cl_enqueue_ND_range_kernel(cl_command_queue command_queue,
> cl_kernel kernel, cl_uint work_dim,
> + const size_t* global_offset,
> const size_t* global_work_size,
> const size_t* local_work_size)
> {
> cl_int errNo;
>
> errNo = clEnqueueNDRangeKernel(command_queue, kernel, work_dim,
> - NULL, global_work_size, local_work_size,
> - 0, NULL, NULL);
> + global_offset, global_work_size,
> + local_work_size, 0, NULL, NULL);
> if(!piglit_cl_check_error(errNo, CL_SUCCESS)) {
> fprintf(stderr,
> "Could not enqueue ND range kernel: %s\n",
> @@ -1284,6 +1285,7 @@ piglit_cl_enqueue_ND_range_kernel(cl_command_queue
> command_queue, bool
> piglit_cl_execute_ND_range_kernel(cl_command_queue command_queue,
> cl_kernel kernel, cl_uint work_dim,
> + const size_t* global_offset,
> const size_t* global_work_size,
> const size_t* local_work_size)
> {
> @@ -1292,6 +1294,7 @@ piglit_cl_execute_ND_range_kernel(cl_command_queue
> command_queue, if(!piglit_cl_enqueue_ND_range_kernel(command_queue,
> kernel,
> work_dim,
> + global_offset,
> global_work_size,
> local_work_size)) {
> return false;
> diff --git a/tests/util/piglit-util-cl.h b/tests/util/piglit-util-cl.h
> index e4730cc..9fb4f9f 100644
> --- a/tests/util/piglit-util-cl.h
> +++ b/tests/util/piglit-util-cl.h
> @@ -704,6 +704,7 @@ piglit_cl_set_kernel_buffer_arg(cl_kernel kernel,
> * @param command_queue Command queue to enqueue operation on.
> * @param kernel Kernel to be enqueued.
> * @param work_dim Work dimensions.
> + * @param global_offset Global offset.
> * @param global_work_size Global work sizes.
> * @param local_work_size Local work sizes.
> * @return \c true on succes, \c false otherwise.
> @@ -712,6 +713,7 @@ bool
> piglit_cl_enqueue_ND_range_kernel(cl_command_queue command_queue,
> cl_kernel kernel,
> cl_uint work_dim,
> + const size_t* global_offset,
> const size_t* global_work_size,
> const size_t* local_work_size);
>
> @@ -721,6 +723,7 @@ piglit_cl_enqueue_ND_range_kernel(cl_command_queue
> command_queue, * @param command_queue Command queue to enqueue
> operation on. * @param kernel Kernel to be enqueued.
> * @param work_dim Work dimensions.
> + * @param global_offset Global offset.
> * @param global_work_size Global work sizes.
> * @param local_work_size Local work sizes.
> * @return \c true on succes, \c false otherwise.
> @@ -729,6 +732,7 @@ bool
> piglit_cl_execute_ND_range_kernel(cl_command_queue command_queue,
> cl_kernel kernel,
> cl_uint work_dim,
> + const size_t* global_offset,
> const size_t* global_work_size,
> const size_t* local_work_size);
More information about the Piglit
mailing list