diff options
author | Ruiling Song <ruiling.song@intel.com> | 2013-06-14 16:32:54 +0800 |
---|---|---|
committer | Zhigang Gong <zhigang.gong@linux.intel.com> | 2013-06-14 17:01:19 +0800 |
commit | 89b5e400e81c8b14dd6c1f5cdf986dc38ca8ef22 (patch) | |
tree | 8973858c93abe8c3f1a980b74501169008802493 | |
parent | b8e88ff296308072986841bbf4a8143e1af6236e (diff) |
add test case for null kernel arg of global/constant buffer
Signed-off-by: Ruiling Song <ruiling.song@intel.com>
Reviewed-by: Zhigang Gong <zhigang.gong@linux.intel.com>
-rw-r--r-- | kernels/null_kernel_arg.cl | 9 | ||||
-rw-r--r-- | utests/CMakeLists.txt | 1 | ||||
-rw-r--r-- | utests/runtime_null_kernel_arg.cpp | 27 |
3 files changed, 37 insertions, 0 deletions
diff --git a/kernels/null_kernel_arg.cl b/kernels/null_kernel_arg.cl new file mode 100644 index 00000000..68a42809 --- /dev/null +++ b/kernels/null_kernel_arg.cl @@ -0,0 +1,9 @@ +__kernel void +null_kernel_arg(__global unsigned int *dst, __global unsigned int * mask_global, __constant unsigned int* mask_const) +{ + if(dst && mask_global==0 && mask_const == NULL) + { + uint idx = (uint)get_global_id(0); + dst[idx] = idx; + } +} diff --git a/utests/CMakeLists.txt b/utests/CMakeLists.txt index 033643cd..e5c03ee8 100644 --- a/utests/CMakeLists.txt +++ b/utests/CMakeLists.txt @@ -83,6 +83,7 @@ set (utests_sources compiler_cl_finish.cpp buildin_work_dim.cpp runtime_createcontext.cpp + runtime_null_kernel_arg.cpp utest_assert.cpp utest.cpp utest_file_map.cpp diff --git a/utests/runtime_null_kernel_arg.cpp b/utests/runtime_null_kernel_arg.cpp new file mode 100644 index 00000000..447e3452 --- /dev/null +++ b/utests/runtime_null_kernel_arg.cpp @@ -0,0 +1,27 @@ +#include "utest_helper.hpp" + +void runtime_null_kernel_arg(void) +{ + const size_t n = 32; + + // Setup kernel and buffers + OCL_CREATE_KERNEL("null_kernel_arg"); + OCL_CREATE_BUFFER(buf[0], 0, n * sizeof(uint32_t), NULL); + OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]); + OCL_SET_ARG(1, sizeof(cl_mem), NULL); + OCL_SET_ARG(2, sizeof(cl_mem), NULL); + + // Run the kernel + globals[0] = n; + locals[0] = 16; + OCL_NDRANGE(1); + OCL_MAP_BUFFER(0); + + // Check results + for (uint32_t i = 0; i < n; ++i) + OCL_ASSERT(((uint32_t*)buf_data[0])[i] == i); + OCL_UNMAP_BUFFER(0); +} + + +MAKE_UTEST_FROM_FUNCTION(runtime_null_kernel_arg); |