[Beignet] [PATCH 1/2] add test case barrier_list and marker_list.

[xionghu.luo at intel.com]

From: Luo <xionghu.luo at intel.com>


Signed-off-by: Luo <xionghu.luo at intel.com>
---
 utests/CMakeLists.txt   |  2 ++
 utests/barrier_list.cpp | 75 +++++++++++++++++++++++++++++++++++++++++++++++++
 utests/marker_list.cpp  | 75 +++++++++++++++++++++++++++++++++++++++++++++++++
 3 files changed, 152 insertions(+)
 create mode 100644 utests/barrier_list.cpp
 create mode 100644 utests/marker_list.cpp

diff --git a/utests/CMakeLists.txt b/utests/CMakeLists.txt
index 415dcb6..f9cc3f1 100644
--- a/utests/CMakeLists.txt
+++ b/utests/CMakeLists.txt
@@ -153,6 +153,8 @@ set (utests_sources
   runtime_createcontext.cpp
   runtime_null_kernel_arg.cpp
   runtime_event.cpp
+  barrier_list.cpp
+  marker_list.cpp
   compiler_double.cpp
   compiler_double_2.cpp
   compiler_double_3.cpp
diff --git a/utests/barrier_list.cpp b/utests/barrier_list.cpp
new file mode 100644
index 0000000..e672482
--- /dev/null
+++ b/utests/barrier_list.cpp
@@ -0,0 +1,75 @@
+#include "utest_helper.hpp"
+
+#define BUFFERSIZE  32*1024
+void barrier_list(void)
+{
+  const size_t n = BUFFERSIZE;
+  cl_int cpu_src[BUFFERSIZE];
+  cl_int cpu_src_2[BUFFERSIZE];
+  cl_event ev[5];
+  cl_int status = 0;
+  cl_int value = 34;
+
+  // Setup kernel and buffers
+  OCL_CREATE_KERNEL("compiler_event");
+  OCL_CREATE_BUFFER(buf[0], 0, BUFFERSIZE*sizeof(int), NULL);
+  OCL_CREATE_BUFFER(buf[1], 0, BUFFERSIZE*sizeof(int), NULL);
+
+  for(cl_uint i=0; i<BUFFERSIZE; i++)
+  {
+    cpu_src[i] = 3;
+    cpu_src_2[i] = 5;
+  }
+
+  OCL_CREATE_USER_EVENT(ev[0]);
+
+  clEnqueueWriteBuffer(queue, buf[0], CL_TRUE, 0, BUFFERSIZE*sizeof(int), (void *)cpu_src, 1, &ev[0], &ev[1]);
+
+  OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]);
+  OCL_SET_ARG(1, sizeof(int), &value);
+
+  // Run the kernel
+  globals[0] = n;
+  locals[0] = 32;
+
+  clEnqueueNDRangeKernel(queue, kernel, 1, NULL, globals, locals, 2, &ev[0], &ev[2]);
+
+  for (cl_uint i = 0; i != sizeof(ev) / sizeof(cl_event); ++i) {
+    clGetEventInfo(ev[i], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+    OCL_ASSERT(status >= CL_SUBMITTED);
+  }
+
+
+  buf_data[0] = clEnqueueMapBuffer(queue, buf[0], CL_TRUE, 0, 0, BUFFERSIZE*sizeof(int), 1, &ev[2], NULL, NULL);
+
+  clEnqueueBarrierWithWaitList(queue, 0, NULL, &ev[3]);
+
+  clEnqueueWriteBuffer(queue, buf[1], CL_TRUE, 0, BUFFERSIZE*sizeof(int), (void *)cpu_src_2, 0, NULL, &ev[4]);
+
+  OCL_FINISH();
+  clGetEventInfo(ev[4], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+  OCL_ASSERT(status != CL_COMPLETE);
+
+  OCL_SET_USER_EVENT_STATUS(ev[0], CL_COMPLETE);
+
+  clGetEventInfo(ev[0], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+  OCL_ASSERT(status == CL_COMPLETE);
+
+  OCL_FINISH();
+
+  for (cl_uint i = 0; i != sizeof(ev) / sizeof(cl_event); ++i) {
+    clGetEventInfo(ev[i], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+    OCL_ASSERT(status <= CL_COMPLETE);
+  }
+
+  for (uint32_t i = 0; i < n; ++i) {
+    OCL_ASSERT(((int*)buf_data[0])[i] == (int)value + 0x3);
+  }
+  clEnqueueUnmapMemObject(queue, buf[0], buf_data[0], 0, NULL, NULL);
+
+  for (cl_uint i = 0; i != sizeof(ev) / sizeof(cl_event); ++i) {
+    clReleaseEvent(ev[i]);
+  }
+}
+
+MAKE_UTEST_FROM_FUNCTION(barrier_list);
diff --git a/utests/marker_list.cpp b/utests/marker_list.cpp
new file mode 100644
index 0000000..cb4e749
--- /dev/null
+++ b/utests/marker_list.cpp
@@ -0,0 +1,75 @@
+#include "utest_helper.hpp"
+
+#define BUFFERSIZE  32*1024
+void marker_list(void)
+{
+  const size_t n = BUFFERSIZE;
+  cl_int cpu_src[BUFFERSIZE];
+  cl_int cpu_src_2[BUFFERSIZE];
+  cl_event ev[5];
+  cl_int status = 0;
+  cl_int value = 34;
+
+  // Setup kernel and buffers
+  OCL_CREATE_KERNEL("compiler_event");
+  OCL_CREATE_BUFFER(buf[0], 0, BUFFERSIZE*sizeof(int), NULL);
+  OCL_CREATE_BUFFER(buf[1], 0, BUFFERSIZE*sizeof(int), NULL);
+
+  for(cl_uint i=0; i<BUFFERSIZE; i++)
+  {
+    cpu_src[i] = 3;
+    cpu_src_2[i] = 5;
+  }
+
+  OCL_CREATE_USER_EVENT(ev[0]);
+
+  clEnqueueWriteBuffer(queue, buf[0], CL_TRUE, 0, BUFFERSIZE*sizeof(int), (void *)cpu_src, 1, &ev[0], &ev[1]);
+
+  OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]);
+  OCL_SET_ARG(1, sizeof(int), &value);
+
+  // Run the kernel
+  globals[0] = n;
+  locals[0] = 32;
+
+  clEnqueueNDRangeKernel(queue, kernel, 1, NULL, globals, locals, 2, &ev[0], &ev[2]);
+
+  for (cl_uint i = 0; i != sizeof(ev) / sizeof(cl_event); ++i) {
+    clGetEventInfo(ev[i], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+    OCL_ASSERT(status >= CL_SUBMITTED);
+  }
+
+
+  buf_data[0] = clEnqueueMapBuffer(queue, buf[0], CL_TRUE, 0, 0, BUFFERSIZE*sizeof(int), 1, &ev[2], NULL, NULL);
+
+  clEnqueueMarkerWithWaitList(queue, 0, NULL, &ev[3]);
+
+  clEnqueueWriteBuffer(queue, buf[1], CL_TRUE, 0, BUFFERSIZE*sizeof(int), (void *)cpu_src_2, 0, NULL, &ev[4]);
+
+  OCL_FINISH();
+  clGetEventInfo(ev[4], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+  OCL_ASSERT(status == CL_COMPLETE);
+
+  OCL_SET_USER_EVENT_STATUS(ev[0], CL_COMPLETE);
+
+  clGetEventInfo(ev[0], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+  OCL_ASSERT(status == CL_COMPLETE);
+
+  OCL_FINISH();
+
+  for (cl_uint i = 0; i != sizeof(ev) / sizeof(cl_event); ++i) {
+    clGetEventInfo(ev[i], CL_EVENT_COMMAND_EXECUTION_STATUS, sizeof(status), &status, NULL);
+    OCL_ASSERT(status <= CL_COMPLETE);
+  }
+
+  for (uint32_t i = 0; i < n; ++i) {
+    OCL_ASSERT(((int*)buf_data[0])[i] == (int)value + 0x3);
+  }
+  clEnqueueUnmapMemObject(queue, buf[0], buf_data[0], 0, NULL, NULL);
+
+  for (cl_uint i = 0; i != sizeof(ev) / sizeof(cl_event); ++i) {
+    clReleaseEvent(ev[i]);
+  }
+}
+
+MAKE_UTEST_FROM_FUNCTION(marker_list);
-- 
1.8.1.2