Finished the small test framework

10 files changed, 112 insertions, 99 deletions

diff --git a/utests/compiler_copy_buffer.cpp b/utests/compiler_copy_buffer.cpp
index 9380ace0..ee1dbf1b 100644
--- a/utests/compiler_copy_buffer.cpp
+++ b/utests/compiler_copy_buffer.cpp
@@ -22,9 +22,9 @@
 static void compiler_copy_buffer(void)
 {
   const size_t n = 8192 * 4;
-  int status = 0;
 
-  CALL (cl_test_init, "test_copy_buffer.cl", "test_copy_buffer", SOURCE);
+  // Setup kernel and buffers
+  OCL_CREATE_KERNEL("test_copy_buffer");
   buf_data[0] = (uint32_t*) malloc(sizeof(uint32_t) * n);
   for (uint32_t i = 0; i < n; ++i) ((uint32_t*)buf_data[0])[i] = i;
   OCL_CREATE_BUFFER(buf[0], CL_MEM_COPY_HOST_PTR, n * sizeof(uint32_t), buf_data[0]);
@@ -32,25 +32,19 @@ static void compiler_copy_buffer(void)
   free(buf_data[0]);
   buf_data[0] = NULL;
 
+  // Run the kernel
   OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]);
   OCL_SET_ARG(1, sizeof(cl_mem), &buf[1]);
-
   globals[0] = n;
   locals[0] = 16;
   OCL_NDRANGE(1);
 
+  // Check result
   OCL_MAP_BUFFER(0);
   OCL_MAP_BUFFER(1);
   for (uint32_t i = 0; i < n; ++i)
-    assert(((uint32_t*)buf_data[0])[i] == ((uint32_t*)buf_data[1])[i]);
-  OCL_UNMAP_BUFFER(0);
-  OCL_UNMAP_BUFFER(1);
-
-error:
-  cl_release_buffers();
-  cl_report_error(status);
-  cl_test_destroy();
+    OCL_ASSERT(((uint32_t*)buf_data[0])[i] == ((uint32_t*)buf_data[1])[i]);
 }
 
-UTEST_REGISTER(compiler_copy_buffer)
+MAKE_UTEST_FROM_FUNCTION(compiler_copy_buffer);
 
diff --git a/utests/compiler_copy_buffer_row.cpp b/utests/compiler_copy_buffer_row.cpp
index 308e6e3e..cfaea0e9 100644
--- a/utests/compiler_copy_buffer_row.cpp
+++ b/utests/compiler_copy_buffer_row.cpp
@@ -26,25 +26,21 @@ static void compiler_copy_buffer_row(void)
   const int row = 8192;
   const int row_n = 2;
   const int n =  row * row_n;
-  int status = 0;
 
-  CALL (cl_test_init, "test_copy_buffer_row.cl", "test_copy_buffer_row", SOURCE);
-
-  // Create the input data
+  // Setup kernel and buffers
+  OCL_CREATE_KERNEL("test_copy_buffer_row");
   src_buffer = (uint32_t *) malloc(sizeof(uint32_t) * n);
   for (int32_t i = 0; i < n; ++i) src_buffer[i] = i;
   data_buffer = (int *) malloc(sizeof(int) * 2);
   data_buffer[0] = row;
   data_buffer[1] = n;
-
-  // Allocate all buffers
   OCL_CREATE_BUFFER(buf[0], CL_MEM_COPY_HOST_PTR, n * sizeof(uint32_t), src_buffer);
   OCL_CREATE_BUFFER(buf[1], 0, n * sizeof(uint32_t), NULL);
   OCL_CREATE_BUFFER(buf[2], CL_MEM_COPY_HOST_PTR, 2 * sizeof(uint32_t), data_buffer);
   free(src_buffer);
   free(data_buffer);
 
-  // Run the code
+  // Run the kernel
   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]);
@@ -56,15 +52,8 @@ static void compiler_copy_buffer_row(void)
   OCL_MAP_BUFFER(0);
   OCL_MAP_BUFFER(1);
   for (int32_t i = 0; i < n; ++i)
-    assert(((uint32_t*)buf_data[0])[i] == ((uint32_t*)buf_data[1])[i]);
-  OCL_UNMAP_BUFFER(0);
-  OCL_UNMAP_BUFFER(1);
-error:
-  cl_release_buffers();
-  cl_report_error(status);
-  cl_test_destroy();
-
+    OCL_ASSERT(((uint32_t*)buf_data[0])[i] == ((uint32_t*)buf_data[1])[i]);
 }
 
-UTEST_REGISTER(compiler_copy_buffer_row)
+MAKE_UTEST_FROM_FUNCTION(compiler_copy_buffer_row);
 
diff --git a/utests/compiler_write_only.cpp b/utests/compiler_write_only.cpp
index f7382532..b42bfd22 100644
--- a/utests/compiler_write_only.cpp
+++ b/utests/compiler_write_only.cpp
@@ -22,23 +22,22 @@
 void compiler_write_only(void)
 {
   const size_t n = 2048;
-  int status = 0;
 
-  CALL (cl_test_init, "test_write_only.cl", "test_write_only", SOURCE);
+  // Setup kernel and buffers
+  OCL_CREATE_KERNEL("test_write_only");
   OCL_CREATE_BUFFER(buf[0], 0, n * sizeof(uint32_t), NULL);
   OCL_SET_ARG(0, sizeof(cl_mem), &buf[0]);
+
+  // Run the kernel
   globals[0] = n;
   locals[0] = 16;
   OCL_NDRANGE(1);
   OCL_MAP_BUFFER(0);
-  for (uint32_t i = 0; i < n; ++i) assert(((uint32_t*)buf_data[0])[i] == i);
-  OCL_UNMAP_BUFFER(0);
 
-error:
-  cl_release_buffers();
-  cl_report_error(status);
-  cl_test_destroy();
+  // Check results
+  for (uint32_t i = 0; i < n; ++i)
+    OCL_ASSERT(((uint32_t*)buf_data[0])[i] == i);
 }
 
-UTEST_REGISTER(compiler_write_only);
+MAKE_UTEST_FROM_FUNCTION(compiler_write_only);
 
diff --git a/utests/utest.cpp b/utests/utest.cpp
index 421c3b68..fc3467ed 100644
--- a/utests/utest.cpp
+++ b/utests/utest.cpp
@@ -22,6 +22,7 @@
  * \author Benjamin Segovia <benjamin.segovia@intel.com>
  */
 #include "utest.hpp"
+#include "utest_helper.hpp"
 #include <vector>
 #include <string>
 #include <iostream>
@@ -56,6 +57,8 @@ void UTest::run(const char *name) {
       std::cout << utest.name << ":" << std::endl;
       (utest.fn)();
       std::cout << std::endl;
+      cl_kernel_destroy();
+      cl_buffer_destroy();
     }
   }
 }
@@ -68,6 +71,8 @@ void UTest::runAll(void) {
     std::cout << utest.name << ":" << std::endl;
     (utest.fn)();
     std::cout << std::endl;
+    cl_kernel_destroy();
+    cl_buffer_destroy();
   }
 }
 
diff --git a/utests/utest.hpp b/utests/utest.hpp
index e444c721..338a4dca 100644
--- a/utests/utest.hpp
+++ b/utests/utest.hpp
@@ -27,8 +27,9 @@
 #ifndef __UTEST_UTEST_HPP__
 #define __UTEST_UTEST_HPP__
 
+#include "utest_exception.hpp"
 #include <vector>
-#include <exception>
+#include <iostream>
 
 /*! Quick and dirty unit test system with registration */
 struct UTest
@@ -51,9 +52,14 @@ struct UTest
   static void runAll(void);
 };
 
-/*! RegisterData a new unit test */
+/*! Register a new unit test */
 #define UTEST_REGISTER(FN) static const UTest __##FN##__(FN, #FN);
 
+/*! Turn a function into a unit test */
+#define MAKE_UTEST_FROM_FUNCTION(FN) \
+  static void __ANON__##FN##__(void) { UTEST_EXPECT_SUCCESS(FN()); } \
+  static const UTest __##FN##__(__ANON__##FN##__, #FN);
+
 /*! No assert is expected */
 #define UTEST_EXPECT_SUCCESS(EXPR) \
  do { \
@@ -61,7 +67,7 @@ struct UTest
       EXPR; \
       std::cout << "  " << #EXPR << "    [SUCCESS]" << std::endl; \
     } \
-    catch (gbe::Exception e) { \
+    catch (Exception e) { \
       std::cout << "  " << #EXPR << "    [FAILED]" << std::endl; \
       std::cout << "    " << e.what() << std::endl; \
     } \
diff --git a/utests/utest_assert.cpp b/utests/utest_assert.cpp
index da656e8d..f3b9a006 100644
--- a/utests/utest_assert.cpp
+++ b/utests/utest_assert.cpp
@@ -31,12 +31,11 @@ void onFailedAssertion(const char *msg, const char *file, const char *fn, int li
   char lineString[256];
   sprintf(lineString, "%i", line);
   assert(msg != NULL && file != NULL && fn != NULL);
-  const std::string str = "Compiler error: "
+  const std::string str = "Error: "
                         + std::string(msg) + "\n  at file "
                         + std::string(file)
                         + ", function " + std::string(fn)
                         + ", line " + std::string(lineString);
-  // assert(0);
   throw Exception(str);
 }
 
diff --git a/utests/utest_assert.hpp b/utests/utest_assert.hpp
index d124e9fb..f93f9ac7 100644
--- a/utests/utest_assert.hpp
+++ b/utests/utest_assert.hpp
@@ -28,5 +28,17 @@
 /*! To ensure that condition truth. Optional message is supported */
 void onFailedAssertion(const char *msg, const char *file, const char *fn, int line);
 
+#define OCL_ASSERT(EXPR) \
+  do { \
+    if (!(EXPR)) \
+      onFailedAssertion(#EXPR, __FILE__, __FUNCTION__, __LINE__); \
+  } while (0)
+
+#define OCL_ASSERTM(EXPR, MSG) \
+  do { \
+    if (!(EXPR)) \
+      onFailedAssertion(MSG, __FILE__, __FUNCTION__, __LINE__); \
+  } while (0)
+
 #endif /* __OCL_ASSERT_HPP__ */
 
diff --git a/utests/utest_helper.cpp b/utests/utest_helper.cpp
index 3423b685..0083618f 100644
--- a/utests/utest_helper.cpp
+++ b/utests/utest_helper.cpp
@@ -160,7 +160,7 @@ cl_kernel_init(const char *file_name, const char *kernel_name, int format)
   }
 
   /* OCL requires to build the program even if it is created from a binary */
-  CALL (clBuildProgram, program, 1, &device, NULL, NULL, NULL);
+  OCL_CALL (clBuildProgram, program, 1, &device, NULL, NULL, NULL);
 
   /* Create a kernel from the program */
   kernel = clCreateKernel(program, kernel_name, &status);
@@ -186,30 +186,30 @@ cl_ocl_init(void)
   size_t i;
 
   /* Get the platform number */
-  CALL (clGetPlatformIDs, 0, NULL, &platform_n);
+  OCL_CALL (clGetPlatformIDs, 0, NULL, &platform_n);
   printf("platform number %u\n", platform_n);
   assert(platform_n >= 1);
 
   /* Get a valid platform */
-  CALL (clGetPlatformIDs, 1, &platform, &platform_n);
-  CALL (clGetPlatformInfo, platform, CL_PLATFORM_PROFILE, sizeof(name), name, NULL);
+  OCL_CALL (clGetPlatformIDs, 1, &platform, &platform_n);
+  OCL_CALL (clGetPlatformInfo, platform, CL_PLATFORM_PROFILE, sizeof(name), name, NULL);
   printf("platform_profile \"%s\"\n", name);
-  CALL (clGetPlatformInfo, platform, CL_PLATFORM_NAME, sizeof(name), name, NULL);
+  OCL_CALL (clGetPlatformInfo, platform, CL_PLATFORM_NAME, sizeof(name), name, NULL);
   printf("platform_name \"%s\"\n", name);
-  CALL (clGetPlatformInfo, platform, CL_PLATFORM_VENDOR, sizeof(name), name, NULL);
+  OCL_CALL (clGetPlatformInfo, platform, CL_PLATFORM_VENDOR, sizeof(name), name, NULL);
   printf("platform_vendor \"%s\"\n", name);
-  CALL (clGetPlatformInfo, platform, CL_PLATFORM_VERSION, sizeof(name), name, NULL);
+  OCL_CALL (clGetPlatformInfo, platform, CL_PLATFORM_VERSION, sizeof(name), name, NULL);
   printf("platform_version \"%s\"\n", name);
 
   /* Get the device (only GPU device is supported right now) */
-  CALL (clGetDeviceIDs, platform, CL_DEVICE_TYPE_GPU, 1, &device, NULL);
-  CALL (clGetDeviceInfo, device, CL_DEVICE_PROFILE, sizeof(name), name, NULL);
+  OCL_CALL (clGetDeviceIDs, platform, CL_DEVICE_TYPE_GPU, 1, &device, NULL);
+  OCL_CALL (clGetDeviceInfo, device, CL_DEVICE_PROFILE, sizeof(name), name, NULL);
   printf("device_profile \"%s\"\n", name);
-  CALL (clGetDeviceInfo, device, CL_DEVICE_NAME, sizeof(name), name, NULL);
+  OCL_CALL (clGetDeviceInfo, device, CL_DEVICE_NAME, sizeof(name), name, NULL);
   printf("device_name \"%s\"\n", name);
-  CALL (clGetDeviceInfo, device, CL_DEVICE_VENDOR, sizeof(name), name, NULL);
+  OCL_CALL (clGetDeviceInfo, device, CL_DEVICE_VENDOR, sizeof(name), name, NULL);
   printf("device_vendor \"%s\"\n", name);
-  CALL (clGetDeviceInfo, device, CL_DEVICE_VERSION, sizeof(name), name, NULL);
+  OCL_CALL (clGetDeviceInfo, device, CL_DEVICE_VERSION, sizeof(name), name, NULL);
   printf("device_version \"%s\"\n", name);
 
   /* Now create a context */
@@ -260,8 +260,8 @@ error:
 void
 cl_kernel_destroy(void)
 {
-  clReleaseKernel(kernel);
-  clReleaseProgram(program);
+  if (kernel) clReleaseKernel(kernel);
+  if (program) clReleaseProgram(program);
   kernel = NULL;
   program = NULL;
 }
@@ -283,26 +283,19 @@ cl_test_destroy(void)
 }
 
 void
-cl_release_buffers(void)
+cl_buffer_destroy(void)
 {
   int i;
-  for (i = 0; i < MAX_BUFFER_N; ++i)
+  for (i = 0; i < MAX_BUFFER_N; ++i) {
+    if (buf_data[i] != NULL) {
+      clIntelUnmapBuffer(buf[i]);
+      buf_data[i] = NULL;
+    }
     if (buf[i] != NULL) {
       clReleaseMemObject(buf[i]);
       buf[i] = NULL;
     }
-}
-
-void
-cl_report_error(cl_int err)
-{
-  if (err > 0)
-    return;
-  if ((uint32_t) -err > err_msg_n)
-    return;
-  if (err == CL_SUCCESS)
-    return;
-  fprintf(stderr, "error %s\n", err_msg[-err]);
+  }
 }
 
 void
diff --git a/utests/utest_helper.hpp b/utests/utest_helper.hpp
index f014b37c..bf0c3ab2 100644
--- a/utests/utest_helper.hpp
+++ b/utests/utest_helper.hpp
@@ -28,47 +28,58 @@
 #include "CL/cl.h"
 #include "CL/cl_intel.h"
 #include "utest.hpp"
+#include "utest_assert.hpp"
+#include "utest_error.h"
 #include <assert.h>
 #include <stdio.h>
 #include <stdlib.h>
 
-#define CALL(FN, ...) \
+#define OCL_THROW_ERROR(FN, STATUS) \
   do { \
-    status = FN(__VA_ARGS__); \
-    if (status != CL_SUCCESS) { \
-      fprintf(stderr, "error calling %s \n", #FN); \
-      goto error; \
-    } \
+    char msg[2048]; \
+    sprintf(msg, "error calling %s with error%s \n", #FN, err_msg[-STATUS]); \
+    OCL_ASSERTM(false, msg); \
+  } while (0)
+
+#define OCL_CALL(FN, ...) \
+  do { \
+    int status = FN(__VA_ARGS__); \
+    if (status != CL_SUCCESS) OCL_THROW_ERROR(FN, status); \
+  } while (0)
+
+#define OCL_CREATE_KERNEL(NAME) \
+  do { \
+    OCL_CALL (cl_kernel_init, NAME".cl", NAME, SOURCE); \
   } while (0)
 
 #define OCL_CREATE_BUFFER(BUFFER, FLAGS, SIZE, DATA) \
   do { \
     cl_int status; \
     BUFFER = clCreateBuffer(ctx, FLAGS, SIZE, DATA, &status); \
-    if (status != CL_SUCCESS) goto error; \
+    if (status != CL_SUCCESS) OCL_THROW_ERROR(FN, status); \
   } while (0)
 
 #define OCL_MAP_BUFFER(ID) \
   do { \
     cl_int status; \
     buf_data[ID] = (int *) clIntelMapBuffer(buf[ID], &status); \
-    if (status != CL_SUCCESS) goto error; \
+    if (status != CL_SUCCESS) OCL_THROW_ERROR(FN, status); \
   } while (0)
 
 #define OCL_UNMAP_BUFFER(ID) \
- do { \
-   CALL (clIntelUnmapBuffer, buf[ID]); \
-   buf_data[ID] = NULL; \
- } while (0)
+  do { \
+    OCL_CALL (clIntelUnmapBuffer, buf[ID]); \
+    buf_data[ID] = NULL; \
+  } while (0)
 
 #define OCL_NDRANGE(DIM_N) \
   do { \
-    CALL (clEnqueueNDRangeKernel, queue, kernel, DIM_N, NULL, globals, locals, 0, NULL, NULL); \
+    OCL_CALL (clEnqueueNDRangeKernel, queue, kernel, DIM_N, NULL, globals, locals, 0, NULL, NULL); \
   } while (0)
 
 #define OCL_SET_ARG(ID, SIZE, ARG) \
   do { \
-    CALL (clSetKernelArg, kernel, ID, SIZE, ARG); \
+    OCL_CALL (clSetKernelArg, kernel, ID, SIZE, ARG); \
   } while (0)
 
 enum { MAX_BUFFER_N = 16 };
@@ -78,10 +89,10 @@ extern cl_context ctx;
 extern cl_program program;
 extern cl_kernel kernel;
 extern cl_command_queue queue;
-extern cl_mem buf[MAX_BUFFER_N];     // initialized at NULL
-extern void* buf_data[MAX_BUFFER_N]; // initialized at NULL
-extern size_t globals[3];            // initialized at zero
-extern size_t locals[3];             // initialized at zero
+extern cl_mem buf[MAX_BUFFER_N];
+extern void* buf_data[MAX_BUFFER_N];
+extern size_t globals[3];
+extern size_t locals[3];
 
 enum {
   SOURCE = 0,
@@ -98,8 +109,8 @@ extern int cl_kernel_init(const char *file_name, const char *kernel_name, int fo
 /* init the bunch of global varaibles here */
 extern int cl_test_init(const char *file_name, const char *kernel_name, int format);
 
-/* Release all the created buffers */
-extern void cl_release_buffers(void);
+/* Unmap and release all the created buffers */
+extern void cl_buffer_destroy(void);
 
 /* Release OCL queue, context and device */
 extern void cl_ocl_destroy(void);
@@ -110,9 +121,6 @@ extern void cl_kernel_destroy(void);
 /* Release everything allocated in cl_test_init */
 extern void cl_test_destroy(void);
 
-/* Properly report the error in stderr */
-extern void cl_report_error(cl_int err);
-
 /* Nicely output the performance counters */
 extern void cl_report_perf_counters(cl_mem perf);
 
diff --git a/utests/utest_run.cpp b/utests/utest_run.cpp
index 9721c75c..e577b7be 100644
--- a/utests/utest_run.cpp
+++ b/utests/utest_run.cpp
@@ -23,14 +23,22 @@
  *
  * Just run the unit tests. The user can possibly provides the subset of it
  */
-#include "utest.hpp"
+#include "utest_helper.hpp"
+#include "utest_exception.hpp"
+#include <iostream>
 
 int main(int argc, char *argv[])
 {
-  if (argc >= 2)
-    for (int i = 1; i < argc; ++i)
-      UTest::run(argv[i]);
-  else
-    UTest::runAll();
+  try {
+    cl_ocl_init();
+    if (argc >= 2)
+      for (int i = 1; i < argc; ++i)
+        UTest::run(argv[i]);
+    else
+      UTest::runAll();
+    cl_ocl_destroy();
+  } catch (Exception e) {
+      std::cout << "  " << e.what() << "    [SUCCESS]" << std::endl;
+  }
 }