[Beignet] [PATCH 2/2] test previous patch

[Zhigang Gong]

Hi Homer,

Could you give this patch a more meaningful subject? Thanks.

On Wed, Jul 24, 2013 at 10:25:40AM +0800, Homer Hsing wrote:
> ---
>  kernels/compiler_double_4.cl |  5 +++++
>  utests/CMakeLists.txt        |  1 +
>  utests/compiler_double_4.cpp | 40 ++++++++++++++++++++++++++++++++++++++++
>  3 files changed, 46 insertions(+)
>  create mode 100644 kernels/compiler_double_4.cl
>  create mode 100644 utests/compiler_double_4.cpp
> 
> diff --git a/kernels/compiler_double_4.cl b/kernels/compiler_double_4.cl
> new file mode 100644
> index 0000000..e5e46f9
> --- /dev/null
> +++ b/kernels/compiler_double_4.cl
> @@ -0,0 +1,5 @@
> +#pragma OPENCL EXTENSION cl_khr_fp64 : enable
> +kernel void compiler_double_4(global double *src1, global double *src2, global double *dst) {
> +  int i = get_global_id(0);
> +  dst[i] = src1[i] + src2[i];
> +}
> diff --git a/utests/CMakeLists.txt b/utests/CMakeLists.txt
> index 95f3d9b..b2670bf 100644
> --- a/utests/CMakeLists.txt
> +++ b/utests/CMakeLists.txt
> @@ -112,6 +112,7 @@ set (utests_sources
>    compiler_double.cpp
>    compiler_double_2.cpp
>    compiler_double_3.cpp
> +  compiler_double_4.cpp
>    utest_assert.cpp
>    utest.cpp
>    utest_file_map.cpp
> diff --git a/utests/compiler_double_4.cpp b/utests/compiler_double_4.cpp
> new file mode 100644
> index 0000000..cb25bd4
> --- /dev/null
> +++ b/utests/compiler_double_4.cpp
> @@ -0,0 +1,40 @@
> +#include <cmath>
> +#include "utest_helper.hpp"
> +
> +void compiler_double_4(void)
> +{
> +  const size_t n = 16;
> +  double cpu_src1[n], cpu_src2[n];
> +
> +  // Setup kernel and buffers
> +  OCL_CREATE_KERNEL("compiler_double_4");
> +  OCL_CREATE_BUFFER(buf[0], 0, n * sizeof(double), NULL);
> +  OCL_CREATE_BUFFER(buf[1], 0, n * sizeof(double), NULL);
> +  OCL_CREATE_BUFFER(buf[2], 0, n * sizeof(double), NULL);
> +  OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]);
> +  OCL_SET_ARG(1, sizeof(cl_mem), &buf[1]);
> +  OCL_SET_ARG(2, sizeof(cl_mem), &buf[2]);
> +  globals[0] = n;
> +  locals[0] = 16;
> +
> +  // Run random tests
> +  OCL_MAP_BUFFER(0);
> +  OCL_MAP_BUFFER(1);
> +  for (int32_t i = 0; i < (int32_t) n; ++i) {
> +    cpu_src1[i] = ((double*)buf_data[0])[i] = rand() * 1e-2;
> +    cpu_src2[i] = ((double*)buf_data[1])[i] = rand() * 1e-2;
> +  }
> +  OCL_UNMAP_BUFFER(0);
> +  OCL_UNMAP_BUFFER(1);
> +
> +  // Run the kernel on GPU
> +  OCL_NDRANGE(1);
> +
> +  // Compare
> +  OCL_MAP_BUFFER(2);
> +  for (int32_t i = 0; i < (int32_t) n; ++i)
> +    OCL_ASSERT(fabs(((double*)buf_data[2])[i] - cpu_src1[i] - cpu_src2[i]) < 1e-4);
> +  OCL_UNMAP_BUFFER(2);
> +}
> +
> +MAKE_UTEST_FROM_FUNCTION(compiler_double_4);
> -- 
> 1.8.1.2
> 
> _______________________________________________
> Beignet mailing list
> Beignet at lists.freedesktop.org
> http://lists.freedesktop.org/mailman/listinfo/beignet