Please ignore this version, there is a typo. Zhu Bingbing would send a new mail.
-----Original Message----- From: Beignet [mailto:[email protected]] On Behalf Of Meng Mengmeng Sent: Tuesday, December 09, 2014 2:18 PM To: [email protected] Cc: Zhu, BingbingX Subject: [Beignet] [PATCH] add test of copy_image_1d into benchmark From: Zhu Bingbing <[email protected]> Signed-off-by: Zhu Bingbing <[email protected]> --- benchmark/CMakeLists.txt | 3 +- benchmark/benchmark_copy_image_1d.cpp | 55 ++++++++++++++++++++++++++++++ kernels/runtime_benchmark_copy_image_1d.cl | 8 +++++ 3 files changed, 65 insertions(+), 1 deletion(-) create mode 100644 benchmark/benchmark_copy_image_1d.cpp create mode 100644 kernels/runtime_benchmark_copy_image_1d.cl diff --git a/benchmark/CMakeLists.txt b/benchmark/CMakeLists.txt index ac2d8aa..86ac922 100644 --- a/benchmark/CMakeLists.txt +++ b/benchmark/CMakeLists.txt @@ -12,7 +12,8 @@ set (benchmark_sources ../utests/utest_helper.cpp ../utests/vload_bench.cpp enqueue_copy_buf.cpp - benchmark_use_host_ptr_buffer.cpp) + benchmark_use_host_ptr_buffer.cpp + benchmark_copy_image_1d.cpp) SET(CMAKE_CXX_FLAGS "-DBUILD_BENCHMARK ${CMAKE_CXX_FLAGS}") diff --git a/benchmark/benchmark_copy_image_1d.cpp b/benchmark/benchmark_copy_image_1d.cpp new file mode 100644 index 0000000..07cf1b8 --- /dev/null +++ b/benchmark/benchmark_copy_image_1d.cpp @@ -0,0 +1,55 @@ +#include <sys/time.h> +#include <string.h> +#include "utests/utest_helper.hpp" + +static int benchmark_copy_image_1d(void) { + size_t i = 0; + const size_t w = 512; + cl_image_format format; + cl_image_desc desc; + cl_sampler sampler; + struct timeval start; + struct timeval stop; + + memset(&desc, 0x0, sizeof(cl_image_desc)); memset(&format, 0x0, + sizeof(cl_image_format)); + + // Setup kernel and images + OCL_CREATE_KERNEL("runtime_benchmark_copy_image_1d"); + buf_data[0] = (uint32_t*) malloc(sizeof(uint32_t) * w); for + (uint32_t i = 0; i < w; i++) + ((uint32_t*)buf_data[0])[i] = i; + + format.image_channel_order = CL_RGBA; format.image_channel_data_type + = CL_UNSIGNED_INT8; desc.image_type = CL_MEM_OBJECT_IMAGE1D; + desc.image_width = w; desc.image_row_pitch = w * sizeof(uint32_t); + OCL_CREATE_IMAGE(buf[0], CL_MEM_COPY_HOST_PTR, &format, &desc, + buf_data[0]); + + desc.image_row_pitch = 0; + OCL_CREATE_SAMPLER(sampler, CL_ADDRESS_REPEAT, CL_FILTER_NEAREST); + + // Run the kernel + OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]); OCL_SET_ARG(1, + sizeof(sampler), &sampler); globals[0] = w; locals[0] = 16; + + gettimeofday(&start,0); + for (i = 0; i < 10000; i++) + { + OCL_NDRANGE(1); + OCL_FINISH(); + } + gettimeofday(&stop,0); + + clReleaseMemObject(buf[0]); + free(buf_data[0]); + buf_data[0] = NULL; + + return time_subtract(&stop, &start, 0); } + +MAKE_BENCHMARK_FROM_FUNCTION(benchmark_copy_image_1d); diff --git a/kernels/runtime_benchmark_copy_image_1d.cl b/kernels/runtime_benchmark_copy_image_1d.cl new file mode 100644 index 0000000..3aa0429 --- /dev/null +++ b/kernels/runtime_benchmark_copy_image_1d.cl @@ -0,0 +1,8 @@ +__kernel void +runtime_benchmark_copy_image_1d(__read_only image1d_t src, sampler_t +sampler) { + int coord; + int4 color; + coord = (int)get_global_id(0); + color = read_imagei(src, sampler, coord); } -- 1.9.3 _______________________________________________ Beignet mailing list [email protected] http://lists.freedesktop.org/mailman/listinfo/beignet _______________________________________________ Beignet mailing list [email protected] http://lists.freedesktop.org/mailman/listinfo/beignet
