diff --git a/test_conformance/pipes/test_pipe_info.cpp b/test_conformance/pipes/test_pipe_info.cpp
index 5d3e3a4..7543c6c 100644
--- a/test_conformance/pipes/test_pipe_info.cpp
+++ b/test_conformance/pipes/test_pipe_info.cpp
@@ -21,29 +21,25 @@
 
 int test_pipe_info( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements )
 {
-    cl_mem pipe;
+    clMemWrapper pipe;
     cl_int err;
     cl_uint pipe_width = 512;
     cl_uint pipe_depth = 1024;
     cl_uint returnVal;
-    cl_program  program;
-    cl_kernel   kernel;
+    clProgramWrapper program;
+    clKernelWrapper kernel;
 
-    pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, pipe_width, pipe_depth, NULL, &err);
+    pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, pipe_width, pipe_depth,
+                        NULL, &err);
     test_error(err, "clCreatePipe failed.");
 
-    err = clGetPipeInfo(pipe, CL_PIPE_PACKET_SIZE, sizeof(pipe_width), (void *)&returnVal, NULL);
-    if ( err )
-    {
-        log_error( "Error calling clGetPipeInfo(): %d\n", err );
-        clReleaseMemObject(pipe);
-        return -1;
-    }
+    err = clGetPipeInfo(pipe, CL_PIPE_PACKET_SIZE, sizeof(pipe_width),
+                        (void *)&returnVal, NULL);
+    test_error(err, "clGetPipeInfo failed.");
 
-    if(pipe_width != returnVal)
+    if (pipe_width != returnVal)
     {
-        log_error( "Error in clGetPipeInfo() check of pipe packet size\n" );
-        clReleaseMemObject(pipe);
+        log_error("Error in clGetPipeInfo() check of pipe packet size\n");
         return -1;
     }
     else
@@ -52,17 +48,11 @@
     }
 
     err = clGetPipeInfo(pipe, CL_PIPE_MAX_PACKETS, sizeof(pipe_depth), (void *)&returnVal, NULL);
-    if ( err )
-    {
-        log_error( "Error calling clGetPipeInfo(): %d\n", err );
-        clReleaseMemObject(pipe);
-        return -1;
-    }
+    test_error(err, "clGetPipeInfo failed.");
 
     if(pipe_depth != returnVal)
     {
         log_error( "Error in clGetPipeInfo() check of pipe max packets\n" );
-        clReleaseMemObject(pipe);
         return -1;
     }
     else
@@ -71,39 +61,20 @@
     }
 
     err = create_single_kernel_helper_with_build_options(context, &program, &kernel, 1, (const char**)&pipe_kernel_code, "pipe_kernel", "-cl-std=CL2.0 -cl-kernel-arg-info");
-    if(err)
-    {
-        clReleaseMemObject(pipe);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating program", -1);
 
     cl_kernel_arg_type_qualifier arg_type_qualifier = 0;
     cl_kernel_arg_type_qualifier expected_type_qualifier = CL_KERNEL_ARG_TYPE_PIPE;
     err = clGetKernelArgInfo( kernel, 0, CL_KERNEL_ARG_TYPE_QUALIFIER, sizeof(arg_type_qualifier), &arg_type_qualifier, NULL );
-    if(err)
-    {
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel);
-        clReleaseProgram(program);
-        print_error(err, "clSetKernelArg failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArgInfo failed", -1);
     err = (arg_type_qualifier != expected_type_qualifier);
+
     if(err)
     {
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel);
-        clReleaseProgram(program);
         print_error(err, "ERROR: Bad type qualifier\n");
         return -1;
     }
 
-    // cleanup
-    clReleaseMemObject(pipe);
-    clReleaseKernel(kernel);
-    clReleaseProgram(program);
-
     return err;
 
 }
diff --git a/test_conformance/pipes/test_pipe_limits.cpp b/test_conformance/pipes/test_pipe_limits.cpp
index 29199df..85247f8 100644
--- a/test_conformance/pipes/test_pipe_limits.cpp
+++ b/test_conformance/pipes/test_pipe_limits.cpp
@@ -15,63 +15,103 @@
 //
 #include "harness/compat.h"
 
+#include <assert.h>
+#include <iomanip>
+#include <iostream>
+#include <sstream>
 #include <stdio.h>
 #include <string.h>
-#include <assert.h>
-#include <sys/types.h>
+#include <string>
 #include <sys/stat.h>
+#include <sys/types.h>
 
 #include "procs.h"
 #include "harness/errorHelpers.h"
 
 #define STRING_LENGTH  1024
 
-void createKernelSourceCode(char *source, int num_pipes)
+void createKernelSourceCode(std::stringstream &stream, int num_pipes)
 {
     int i;
-    char str[256];
-    int str_length;
 
-    strcpy(source, "__kernel void test_multiple_pipe_write(__global int *src, ");
+    stream << "__kernel void test_multiple_pipe_write(__global int *src, ";
+    for (i = 0; i < num_pipes; i++)
+    {
+        stream << "__write_only pipe int pipe" << i << ", ";
+    }
+    stream << R"(int num_pipes )
+    {
+          int gid = get_global_id(0);
+          reserve_id_t res_id;
 
-    for(i = 0; i < num_pipes; i++) {
-        sprintf(str, "__write_only pipe int pipe%d, ", i);
-        strcat(source, str);
-    }
-    sprintf(str, "int num_pipes ) \n{\n  int gid = get_global_id(0);\n  reserve_id_t res_id;\n\n");
-    strcat(source, str);
-    sprintf(str, "  if(gid < (get_global_size(0))/num_pipes)\n  {\n    res_id = reserve_write_pipe(pipe0, 1);\n    if(is_valid_reserve_id(res_id))\n    {\n");
-    strcat(source, str);
-    sprintf(str, "      write_pipe(pipe0, res_id, 0, &src[gid]);\n      commit_write_pipe(pipe0, res_id);\n    }\n  }\n");
-    strcat(source, str);
-    for(i = 1; i < num_pipes; i++){
-        sprintf(str, "  else if(gid < (%d*get_global_size(0))/num_pipes)\n  {\n    res_id = reserve_write_pipe(pipe%d, 1);\n    if(is_valid_reserve_id(res_id))\n    {\n", i+1, i);
-        strcat(source, str);
-        sprintf(str, "      write_pipe(pipe%d, res_id, 0, &src[gid]);\n      commit_write_pipe(pipe%d, res_id);\n    }\n  }\n", i, i);
-        strcat(source, str);
-    }
-    strcat(source, "}\n\n__kernel void test_multiple_pipe_read(__global int *dst, ");
 
-    for(i = 0; i < num_pipes; i++) {
-        sprintf(str, "__read_only pipe int pipe%d, ", i);
-        strcat(source, str);
-    }
-    sprintf(str, "int num_pipes ) \n{\n  int gid = get_global_id(0);\n  reserve_id_t res_id;\n\n");
-    strcat(source, str);
-    sprintf(str, "  if(gid < (get_global_size(0))/num_pipes)\n  {\n    res_id = reserve_read_pipe(pipe0, 1);\n    if(is_valid_reserve_id(res_id))\n    {\n");
-    strcat(source, str);
-    sprintf(str, "      read_pipe(pipe0, res_id, 0, &dst[gid]);\n      commit_read_pipe(pipe0, res_id);\n    }\n  }\n");
-    strcat(source, str);
-    for(i = 1; i < num_pipes; i++){
-        sprintf(str, "  else if(gid < (%d*get_global_size(0))/num_pipes)\n  {\n    res_id = reserve_read_pipe(pipe%d, 1);\n    if(is_valid_reserve_id(res_id))\n    {\n", i+1, i);
-        strcat(source, str);
-        sprintf(str, "      read_pipe(pipe%d, res_id, 0, &dst[gid]);\n      commit_read_pipe(pipe%d, res_id);\n    }\n  }\n", i, i);
-        strcat(source, str);
-    }
-    strcat(source, "}");
+          if(gid < (get_global_size(0))/num_pipes)
+          {
+                 res_id = reserve_write_pipe(pipe0, 1);
+                 if(is_valid_reserve_id(res_id))
+                 {
+                     write_pipe(pipe0, res_id, 0, &src[gid]);
+                     commit_write_pipe(pipe0, res_id);
+                 }
+          })";
 
-    str_length = strlen(source);
-    assert(str_length <= STRING_LENGTH*num_pipes);
+    for (i = 1; i < num_pipes; i++)
+    {
+        // clang-format off
+        stream << R"(
+          else if(gid < ()" << (i + 1) << R"(*get_global_size(0))/num_pipes)
+          {
+                 res_id = reserve_write_pipe(pipe)" << i << R"(, 1);
+                 if(is_valid_reserve_id(res_id))
+                 {
+                     write_pipe(pipe)" << i << R"(, res_id, 0, &src[gid]);
+                     commit_write_pipe(pipe)" << i << R"(, res_id);
+                  }
+          }
+          )";
+          // clang-format om
+    }
+    stream << R"(
+    }
+
+    __kernel void test_multiple_pipe_read(__global int *dst, )";
+
+    for (i = 0; i < num_pipes; i++)
+    {
+        stream << "__read_only pipe int pipe" << i << ", ";
+    }
+    stream << R"(int num_pipes )
+    {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
+
+
+            if(gid < (get_global_size(0))/num_pipes)
+            {
+                res_id = reserve_read_pipe(pipe0, 1);
+                if(is_valid_reserve_id(res_id))
+                {
+                    read_pipe(pipe0, res_id, 0, &dst[gid]);
+                    commit_read_pipe(pipe0, res_id);
+                }
+            })";
+
+    for (i = 1; i < num_pipes; i++)
+    {
+        // clang-format off
+        stream << R"(
+            else if(gid < ()"    << (i + 1) << R"(*get_global_size(0))/num_pipes)
+            {
+                res_id = reserve_read_pipe(pipe)" << i << R"(, 1);
+                if(is_valid_reserve_id(res_id))
+                {
+                    read_pipe(pipe)" << i << R"(, res_id, 0, &dst[gid]);
+                    commit_read_pipe(pipe)" << i << R"(, res_id);
+                }
+            })";
+        // clang-format on
+    }
+    stream << "}";
 }
 
 static int verify_result(void *ptr1, void *ptr2, int n)
@@ -113,28 +153,34 @@
 int test_pipe_max_args(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
 {
 
-    cl_mem        pipes[1024];
-    cl_mem      buffers[2];
-    void        *outptr;
-    cl_int        *inptr;
-    cl_program  program;
-    cl_kernel   kernel[2];
-    size_t      global_work_size[3];
-    cl_int      err;
-    cl_int        size;
-    int            num_pipe_elements = 1024;
-    int         i, j;
-    int            max_pipe_args;
-    char        *source;
-    cl_event    producer_sync_event = NULL;
-    cl_event    consumer_sync_event = NULL;
-    MTdata      d = init_genrand( gRandomSeed );
-    const char*    kernelName[] = {"test_multiple_pipe_write", "test_multiple_pipe_read"};
+    clMemWrapper pipes[1024];
+    clMemWrapper buffers[2];
+    void *outptr;
+    cl_int *inptr;
+    clProgramWrapper program;
+    clKernelWrapper kernel[2];
+    size_t global_work_size[3];
+    cl_int err;
+    cl_int size;
+    int num_pipe_elements = 1024;
+    int i, j;
+    int max_pipe_args;
+    std::stringstream source;
+    clEventWrapper producer_sync_event = NULL;
+    clEventWrapper consumer_sync_event = NULL;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<cl_int> BufferOutPtr;
 
-    size_t      min_alignment = get_min_alignment(context);
+    MTdataHolder d(gRandomSeed);
+    const char *kernelName[] = { "test_multiple_pipe_write",
+                                 "test_multiple_pipe_read" };
 
-    err = clGetDeviceInfo(deviceID, CL_DEVICE_MAX_PIPE_ARGS, sizeof(max_pipe_args), (void*)&max_pipe_args, NULL);
-    if(err){
+    size_t min_alignment = get_min_alignment(context);
+
+    err = clGetDeviceInfo(deviceID, CL_DEVICE_MAX_PIPE_ARGS,
+                          sizeof(max_pipe_args), (void *)&max_pipe_args, NULL);
+    if (err)
+    {
         print_error(err, " clGetDeviceInfo failed\n");
         return -1;
     }
@@ -145,76 +191,41 @@
 
     global_work_size[0] = (cl_uint)num_pipe_elements * max_pipe_args;
     size = sizeof(int) * num_pipe_elements * max_pipe_args;
-    source = (char *)malloc(STRING_LENGTH * sizeof(char) * max_pipe_args);
 
     inptr = (cl_int *)align_malloc(size, min_alignment);
 
     for(i = 0; i < num_pipe_elements * max_pipe_args; i++){
         inptr[i] = (int)genrand_int32(d);
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
 
     buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        free(source);
-        print_error(err, " clCreateBuffer failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     outptr = align_malloc(size, min_alignment);
+    BufferOutPtr.reset(outptr, nullptr, 0, size, true);
     buffers[1] = clCreateBuffer(context, CL_MEM_USE_HOST_PTR,  size, outptr, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        free(source);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     for(i = 0; i < max_pipe_args; i++){
         pipes[i] = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, sizeof(int), num_pipe_elements, NULL, &err);
-        if(err){
-            clReleaseMemObject(buffers[0]);
-            clReleaseMemObject(buffers[1]);
-            align_free( outptr );
-            free(source);
-            for(j = 0; j < i; j++) {
-                clReleaseMemObject(pipes[j]);
-            }
-            print_error(err, " clCreatePipe failed\n");
-            return -1;
-        }
+        test_error_ret(err, " clCreatePipe failed", -1);
     }
 
     createKernelSourceCode(source, max_pipe_args);
 
+    std::string kernel_source = source.str();
+    const char *sources[] = { kernel_source.c_str() };
+
     // Create producer kernel
-    err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, (const char**)&source, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        align_free(outptr);
-        free(source);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    err = create_single_kernel_helper_with_build_options(
+        context, &program, &kernel[0], 1, sources, kernelName[0],
+        "-cl-std=CL2.0");
+    test_error_ret(err, " Error creating program", -1);
+
     //Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        align_free(outptr);
-        free(source);
-        print_error(err, " Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     for( i = 0; i < max_pipe_args; i++){
@@ -226,127 +237,28 @@
         err |= clSetKernelArg(kernel[1], i+1, sizeof(cl_mem), (void*)&pipes[i]);
     }
     err |= clSetKernelArg(kernel[1], max_pipe_args + 1, sizeof(int), (void*)&max_pipe_args);
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, NULL, 0, NULL, &producer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     // Launch Consumer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, NULL, 1, &producer_sync_event, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size, outptr, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clWaitForEvents(1, &consumer_sync_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clWaitForEvents failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clWaitForEvents failed", -1);
 
     if( verify_result( inptr, outptr, num_pipe_elements*sizeof(cl_int))){
         log_error("test_pipe_max_args failed\n");
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        for(j = 0; j < max_pipe_args; j++) {
-            clReleaseMemObject(pipes[j]);
-        }
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
     }
     else {
         log_info("test_pipe_max_args passed\n");
     }
-    //cleanup
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    for(j = 0; j < max_pipe_args; j++) {
-        clReleaseMemObject(pipes[j]);
-    }
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseEvent(producer_sync_event);
-    clReleaseEvent(consumer_sync_event);
-    clReleaseProgram(program);
-    align_free(outptr);
-    free(source);
 
     return 0;
 }
@@ -354,39 +266,41 @@
 
 int test_pipe_max_packet_size(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
 {
-    cl_mem        pipe;
-    cl_mem      buffers[2];
-    void        *outptr;
-    cl_char        *inptr;
-    cl_program  program;
-    cl_kernel   kernel[2];
-    size_t      global_work_size[3];
-    cl_int      err;
-    size_t        size;
-    int            num_pipe_elements = 1024;
-    int         i;
-    cl_uint        max_pipe_packet_size;
-    char        *source;
-    char        str[256];
-    int            str_length;
-    cl_event    producer_sync_event = NULL;
-    cl_event    consumer_sync_event = NULL;
-    MTdata      d = init_genrand( gRandomSeed );
-    const char*    kernelName[] = {"test_pipe_max_packet_size_write", "test_pipe_max_packet_size_read"};
+    clMemWrapper pipe;
+    clMemWrapper buffers[2];
+    void *outptr;
+    cl_char *inptr;
+    clProgramWrapper program;
+    clKernelWrapper kernel[2];
+    size_t global_work_size[3];
+    cl_int err;
+    size_t size;
+    int num_pipe_elements = 1024;
+    int i;
+    cl_uint max_pipe_packet_size;
+    clEventWrapper producer_sync_event = NULL;
+    clEventWrapper consumer_sync_event = NULL;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<cl_int> BufferOutPtr;
+    MTdataHolder d(gRandomSeed);
+    const char *kernelName[] = { "test_pipe_max_packet_size_write",
+                                 "test_pipe_max_packet_size_read" };
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
     global_work_size[0] = (cl_uint)num_pipe_elements;
 
-    source = (char*)malloc(STRING_LENGTH*sizeof(char));
+    std::stringstream source;
 
-    err = clGetDeviceInfo(deviceID, CL_DEVICE_PIPE_MAX_PACKET_SIZE, sizeof(max_pipe_packet_size), (void*)&max_pipe_packet_size, NULL);
-    if(err){
-        print_error(err, " clGetDeviceInfo failed\n");
-        return -1;
-    }
-    if(max_pipe_packet_size < 1024){
-        log_error("The device should support minimum packet size of 1024 bytes");
+    err = clGetDeviceInfo(deviceID, CL_DEVICE_PIPE_MAX_PACKET_SIZE,
+                          sizeof(max_pipe_packet_size),
+                          (void *)&max_pipe_packet_size, NULL);
+    test_error_ret(err, " clCreatePipe failed", -1);
+
+    if (max_pipe_packet_size < 1024)
+    {
+        log_error(
+            "The device should support minimum packet size of 1024 bytes");
         return -1;
     }
 
@@ -402,354 +316,248 @@
     for(i = 0; i < size; i++){
         inptr[i] = (char)genrand_int32(d);
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
 
     buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        free(source);
-        print_error(err, " clCreateBuffer failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     outptr = align_malloc(size, min_alignment);
+    BufferOutPtr.reset(outptr, nullptr, 0, size, true);
+
     buffers[1] = clCreateBuffer(context, CL_MEM_USE_HOST_PTR,  size, outptr, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        free(source);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, max_pipe_packet_size, num_pipe_elements, NULL, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        free(source);
-        clReleaseMemObject(pipe);
-        print_error(err, " clCreatePipe failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreatePipe failed", -1);
 
-    sprintf(str, "typedef struct{\n  char a[%d];\n}TestStruct;\n\n__kernel void test_pipe_max_packet_size_write(__global TestStruct *src, __write_only pipe TestStruct out_pipe)\n{\n", max_pipe_packet_size);
-    strcpy(source,str);
-    strcat(source, "  int gid = get_global_id(0);\n  reserve_id_t res_id;\n\n");
-    sprintf(str, "  res_id = reserve_write_pipe(out_pipe, 1);\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    write_pipe(out_pipe, res_id, 0, &src[gid]);\n    commit_write_pipe(out_pipe, res_id);\n  }\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void test_pipe_max_packet_size_read(__read_only pipe TestStruct in_pipe, __global TestStruct *dst)\n{\n");
-    strcat(source, str);
-    strcat(source, "  int gid = get_global_id(0);\n  reserve_id_t res_id;\n\n");
-    sprintf(str, "  res_id = reserve_read_pipe(in_pipe, 1);\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    read_pipe(in_pipe, res_id, 0, &dst[gid]);\n    commit_read_pipe(in_pipe, res_id);\n  }\n}\n\n");
-    strcat(source, str);
+    // clang-format off
+    source << R"(
+        typedef struct{
+            char a[)" << max_pipe_packet_size << R"(];
+        }TestStruct;
 
-    str_length = strlen(source);
-    assert(str_length <= STRING_LENGTH);
+        __kernel void test_pipe_max_packet_size_write(__global TestStruct *src, __write_only pipe TestStruct out_pipe)
+        {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
+
+            res_id = reserve_write_pipe(out_pipe, 1);
+            if(is_valid_reserve_id(res_id))
+            {
+                write_pipe(out_pipe, res_id, 0, &src[gid]);
+                commit_write_pipe(out_pipe, res_id);
+            }
+        }
+
+        __kernel void test_pipe_max_packet_size_read(__read_only pipe TestStruct in_pipe, __global TestStruct *dst)
+        {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
+
+            res_id = reserve_read_pipe(in_pipe, 1);
+            if(is_valid_reserve_id(res_id))
+            {
+                read_pipe(in_pipe, res_id, 0, &dst[gid]);
+                commit_read_pipe(in_pipe, res_id);
+            }
+        }
+        )";
+    // clang-format on
+
+    std::string kernel_source = source.str();
+    const char *sources[] = { kernel_source.c_str() };
 
     // Create producer kernel
-    err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, (const char**)&source, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        free(source);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    err = create_single_kernel_helper_with_build_options(
+        context, &program, &kernel[0], 1, sources, kernelName[0],
+        "-cl-std=CL2.0");
+    test_error_ret(err, " Error creating program", -1);
+
     //Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        free(source);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     err |= clSetKernelArg(kernel[0], 1, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 0, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 1, sizeof(cl_mem), (void*)&buffers[1]);
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, NULL, 0, NULL, &producer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     // Launch Consumer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, NULL, 1, &producer_sync_event, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size, outptr, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if( verify_result( inptr, outptr, size)){
         log_error("test_pipe_max_packet_size failed\n");
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        free(source);
-        return -1;
     }
     else {
         log_info("test_pipe_max_packet_size passed\n");
     }
-    //cleanup
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    clReleaseMemObject(pipe);
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseEvent(producer_sync_event);
-    clReleaseEvent(consumer_sync_event);
-    clReleaseProgram(program);
-    align_free(outptr);
-    free(source);
 
     return 0;
 }
 
 int test_pipe_max_active_reservations(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
 {
-    cl_mem        pipe;
-    cl_mem      buffers[2];
-    cl_mem        buf_reservations;
-    cl_mem        buf_status;
-    cl_mem        buf_reserve_id_t_size;
-    cl_mem        buf_reserve_id_t_size_aligned;
-    cl_int      *inptr;
-    void        *outptr;
-    int            size, i;
-    cl_program  program;
-    cl_kernel   kernel[3];
-    size_t      global_work_size[3];
-    cl_int      err;
-    int            status = 0;
-    cl_uint        max_active_reservations = 0;
-    cl_ulong    max_global_size = 0;
-    int            reserve_id_t_size;
-    int            temp;
-    char        *source;
-    char        str[256];
-    int            str_length;
-    cl_event    sync_event = NULL;
-    cl_event    read_event = NULL;
-    MTdata      d = init_genrand( gRandomSeed );
-    const char*    kernelName[3] = {"test_pipe_max_active_reservations_write", "test_pipe_max_active_reservations_read", "pipe_get_reserve_id_t_size"};
+    clMemWrapper pipe;
+    clMemWrapper buffers[2];
+    clMemWrapper buf_reservations;
+    clMemWrapper buf_status;
+    clMemWrapper buf_reserve_id_t_size;
+    clMemWrapper buf_reserve_id_t_size_aligned;
+    cl_int *inptr;
+    void *outptr;
+    int size, i;
+    clProgramWrapper program;
+    clKernelWrapper kernel[3];
+    size_t global_work_size[3];
+    cl_int err;
+    int status = 0;
+    cl_uint max_active_reservations = 0;
+    cl_ulong max_global_size = 0;
+    int reserve_id_t_size;
+    int temp;
+    clEventWrapper sync_event = NULL;
+    clEventWrapper read_event = NULL;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<cl_int> BufferOutPtr;
+    MTdataHolder d(gRandomSeed);
+    const char *kernelName[3] = { "test_pipe_max_active_reservations_write",
+                                  "test_pipe_max_active_reservations_read",
+                                  "pipe_get_reserve_id_t_size" };
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
-    source = (char*)malloc(2*STRING_LENGTH*sizeof(char));
+    std::stringstream source;
 
     global_work_size[0] = 1;
 
-    err = clGetDeviceInfo(deviceID, CL_DEVICE_PIPE_MAX_ACTIVE_RESERVATIONS, sizeof(max_active_reservations), (void*)&max_active_reservations, NULL);
-    if(err){
-        print_error(err, " clGetDeviceInfo failed\n");
-        return -1;
-    }
+    err = clGetDeviceInfo(deviceID, CL_DEVICE_PIPE_MAX_ACTIVE_RESERVATIONS,
+                          sizeof(max_active_reservations),
+                          (void *)&max_active_reservations, NULL);
+    test_error_ret(err, " clGetDeviceInfo failed", -1);
 
-    err = clGetDeviceInfo(deviceID, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(max_global_size), (void*)&max_global_size, NULL);
-    if(err){
-        print_error(err, " clGetDeviceInfo failed\n");
-        return -1;
-    }
+    err = clGetDeviceInfo(deviceID, CL_DEVICE_GLOBAL_MEM_SIZE,
+                          sizeof(max_global_size), (void *)&max_global_size,
+                          NULL);
+    test_error_ret(err, " clGetDeviceInfo failed", -1);
 
-    max_active_reservations = (max_active_reservations > max_global_size) ? 1<<16 : max_active_reservations;
+    max_active_reservations = (max_active_reservations > max_global_size)
+        ? 1 << 16
+        : max_active_reservations;
 
-    if(max_active_reservations < 1){
+    if (max_active_reservations < 1)
+    {
         log_error("The device should support minimum active reservations of 1");
         return -1;
     }
 
     // To get reserve_id_t size
     buf_reserve_id_t_size = clCreateBuffer(context, CL_MEM_HOST_READ_ONLY, sizeof(reserve_id_t_size), NULL, &err);
-    if ( err ){
-        clReleaseMemObject(buf_reserve_id_t_size);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
-    sprintf(str, "__kernel void test_pipe_max_active_reservations_write(__global int *src, __write_only pipe int out_pipe, __global char *reserve_id, __global int *reserve_id_t_size_aligned, __global int *status)\n{\n");
-    strcpy(source,str);
-    sprintf(str, "  __global reserve_id_t *res_id_ptr;\n  int reserve_idx;\n  int commit_idx;\n");
-    strcat(source, str);
-    sprintf(str, "  for(reserve_idx = 0; reserve_idx < %d; reserve_idx++)\n  {\n", max_active_reservations);
-    strcat(source, str);
-    sprintf(str, "    res_id_ptr = (__global reserve_id_t*)(reserve_id + reserve_idx*reserve_id_t_size_aligned[0]);\n");
-    strcat(source, str);
-    sprintf(str, "    *res_id_ptr = reserve_write_pipe(out_pipe, 1);\n");
-    strcat(source, str);
-    sprintf(str, "    if(is_valid_reserve_id(res_id_ptr[0]))\n    {\n      write_pipe(out_pipe, res_id_ptr[0], 0, &src[reserve_idx]);\n    }\n");
-    strcat(source, str);
-    sprintf(str, "    else\n    {\n      *status = -1;\n      return;\n    }\n  }\n");
-    strcat(source, str);
-    sprintf(str, "  for(commit_idx = 0; commit_idx < %d; commit_idx++)\n  {\n", max_active_reservations);
-    strcat(source, str);
-    sprintf(str, "    res_id_ptr = (__global reserve_id_t*)(reserve_id + commit_idx*reserve_id_t_size_aligned[0]);\n");
-    strcat(source, str);
-    sprintf(str, "    commit_write_pipe(out_pipe, res_id_ptr[0]);\n  }\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void test_pipe_max_active_reservations_read(__read_only pipe int in_pipe, __global int *dst, __global char *reserve_id, __global int *reserve_id_t_size_aligned, __global int *status)\n{\n");
-    strcat(source, str);
-    sprintf(str, "  __global reserve_id_t *res_id_ptr;\n  int reserve_idx;\n  int commit_idx;\n");
-    strcat(source, str);
-    sprintf(str, "  for(reserve_idx = 0; reserve_idx < %d; reserve_idx++)\n  {\n", max_active_reservations);
-    strcat(source, str);
-    sprintf(str, "    res_id_ptr = (__global reserve_id_t*)(reserve_id + reserve_idx*reserve_id_t_size_aligned[0]);\n");
-    strcat(source, str);
-    sprintf(str, "    *res_id_ptr = reserve_read_pipe(in_pipe, 1);\n");
-    strcat(source, str);
-    sprintf(str, "    if(is_valid_reserve_id(res_id_ptr[0]))\n    {\n      read_pipe(in_pipe, res_id_ptr[0], 0, &dst[reserve_idx]);\n    }\n");
-    strcat(source, str);
-    sprintf(str, "    else\n    {\n      *status = -1;\n      return;\n    }\n  }\n");
-    strcat(source, str);
-    sprintf(str, "  for(commit_idx = 0; commit_idx < %d; commit_idx++)\n  {\n", max_active_reservations);
-    strcat(source, str);
-    sprintf(str, "    res_id_ptr = (__global reserve_id_t*)(reserve_id + commit_idx*reserve_id_t_size_aligned[0]);\n");
-    strcat(source, str);
-    sprintf(str, "    commit_read_pipe(in_pipe, res_id_ptr[0]);\n  }\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void pipe_get_reserve_id_t_size(__global int *reserve_id_t_size) \n");
-    strcat(source, str);
-    sprintf(str, "{\n  *reserve_id_t_size = sizeof(reserve_id_t);\n}\n");
-    strcat(source, str);
+    // clang-format off
+    source << R"(
+        __kernel void test_pipe_max_active_reservations_write(__global int *src, __write_only pipe int out_pipe, __global char *reserve_id, __global int *reserve_id_t_size_aligned, __global int *status)
+        {
+            __global reserve_id_t *res_id_ptr;
+            int reserve_idx;
+            int commit_idx;
 
-    str_length = strlen(source);
-    assert(str_length <= 2*STRING_LENGTH);
+            for(reserve_idx = 0; reserve_idx < )" << max_active_reservations << R"(; reserve_idx++)
+            {
+                res_id_ptr = (__global reserve_id_t*)(reserve_id + reserve_idx*reserve_id_t_size_aligned[0]);
+                *res_id_ptr = reserve_write_pipe(out_pipe, 1);
+                if(is_valid_reserve_id(res_id_ptr[0]))
+                {
+                    write_pipe(out_pipe, res_id_ptr[0], 0, &src[reserve_idx]);
+                }
+                else
+                {
+                    *status = -1;
+                    return;
+                }
+            }
+
+            for(commit_idx = 0; commit_idx < )" << max_active_reservations << R"(; commit_idx++)
+            {
+                res_id_ptr = (__global reserve_id_t*)(reserve_id + commit_idx*reserve_id_t_size_aligned[0]);
+                commit_write_pipe(out_pipe, res_id_ptr[0]);
+            }
+        }
+
+        __kernel void test_pipe_max_active_reservations_read(__read_only pipe int in_pipe, __global int *dst, __global char *reserve_id, __global int *reserve_id_t_size_aligned, __global int *status)
+        {
+            __global reserve_id_t *res_id_ptr;
+            int reserve_idx;
+            int commit_idx;
+
+            for(reserve_idx = 0; reserve_idx < )" << max_active_reservations << R"(; reserve_idx++)
+            {
+                res_id_ptr = (__global reserve_id_t*)(reserve_id + reserve_idx*reserve_id_t_size_aligned[0]);
+                *res_id_ptr = reserve_read_pipe(in_pipe, 1);
+
+                if(is_valid_reserve_id(res_id_ptr[0]))
+                {
+                    read_pipe(in_pipe, res_id_ptr[0], 0, &dst[reserve_idx]);
+                }
+                else
+                {
+                    *status = -1;
+                    return;
+                }
+            }
+
+            for(commit_idx = 0; commit_idx < )" << max_active_reservations << R"(; commit_idx++)
+            {
+                res_id_ptr = (__global reserve_id_t*)(reserve_id + commit_idx*reserve_id_t_size_aligned[0]);
+                commit_read_pipe(in_pipe, res_id_ptr[0]);
+            }
+        }
+
+        __kernel void pipe_get_reserve_id_t_size(__global int *reserve_id_t_size)
+        {
+            *reserve_id_t_size = sizeof(reserve_id_t);
+        }
+        )";
+    // clang-format on
+
+    std::string kernel_source = source.str();
+    const char *sources[] = { kernel_source.c_str() };
 
     // Create producer kernel
-    err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, (const char**)&source, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buf_reserve_id_t_size);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    err = create_single_kernel_helper_with_build_options(
+        context, &program, &kernel[0], 1, sources, kernelName[0],
+        "-cl-std=CL2.0");
+    test_error_ret(err, " Error creating program", -1);
 
     // Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buf_reserve_id_t_size);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     // Create size query kernel for reserve_id_t
     kernel[2] = clCreateKernel(program, kernelName[2], &err);
-    if( kernel[2] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buf_reserve_id_t_size);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
+
     err = clSetKernelArg(kernel[2], 0, sizeof(cl_mem), (void*)&buf_reserve_id_t_size);
-    if(err){
-        clReleaseMemObject(buf_reserve_id_t_size);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
+
     //Launch size query kernel for reserve_id_t
     err = clEnqueueNDRangeKernel( queue, kernel[2], 1, NULL, global_work_size, NULL, 0, NULL, &sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buf_reserve_id_t_size, true, 0, sizeof(reserve_id_t_size), &reserve_id_t_size, 1, &sync_event, &read_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     err = clWaitForEvents(1, &read_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clWaitForEvents failed" );
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clWaitForEvents failed", -1);
 
     // Round reserve_id_t_size to the nearest power of 2
     temp = 1;
@@ -763,323 +571,84 @@
     for(i = 0; i < max_active_reservations; i++){
         inptr[i] = (int)genrand_int32(d);
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
 
     buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
-    if ( err ){
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buffers[0]);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     outptr = align_malloc(size, min_alignment);
+    BufferOutPtr.reset(outptr, nullptr, 0, size, true);
+
     buffers[1] = clCreateBuffer(context, CL_MEM_HOST_READ_ONLY, size, NULL, &err);
-    if ( err ){
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     buf_reserve_id_t_size_aligned = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, sizeof(reserve_id_t_size), &reserve_id_t_size, &err);
-    if ( err ){
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     //For error status
     buf_status = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR,  sizeof(int), &status, &err);
-    if ( err ){
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, sizeof(int), max_active_reservations, NULL, &err);
-    if(err){
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        print_error(err, " clCreatePipe failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreatePipe failed", -1);
 
     // Global buffer to hold all active reservation ids
     buf_reservations = clCreateBuffer(context, CL_MEM_HOST_NO_ACCESS, reserve_id_t_size*max_active_reservations, NULL, &err);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clCreateBuffer failed" );
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseEvent(read_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     err |= clSetKernelArg(kernel[0], 1, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[0], 2, sizeof(cl_mem), (void*)&buf_reservations);
     err |= clSetKernelArg(kernel[0], 3, sizeof(cl_mem), (void*)&buf_reserve_id_t_size_aligned);
     err |= clSetKernelArg(kernel[0], 4, sizeof(cl_mem), (void*)&buf_status);
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseProgram(program);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     err = clSetKernelArg(kernel[1], 0, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 1, sizeof(cl_mem), (void*)&buffers[1]);
     err |= clSetKernelArg(kernel[1], 2, sizeof(cl_mem), (void*)&buf_reservations);
     err |= clSetKernelArg(kernel[1], 3, sizeof(cl_mem), (void*)&buf_reserve_id_t_size_aligned);
     err |= clSetKernelArg(kernel[1], 4, sizeof(cl_mem), (void*)&buf_status);
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseProgram(program);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
+
+    clReleaseEvent(sync_event);
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel(queue, kernel[0], 1, NULL, global_work_size, NULL, 0, NULL, &sync_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buf_status, true, 0, sizeof(int), &status, 1, &sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if(status != 0)
     {
         log_error("test_pipe_max_active_reservations failed\n");
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
         return -1;
     }
 
+    clReleaseEvent(sync_event);
     // Launch Consumer kernel
     err = clEnqueueNDRangeKernel(queue, kernel[1], 1, NULL, global_work_size, NULL, 0, NULL, &sync_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buf_status, true, 0, sizeof(int), &status, 1, &sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if(status != 0)
     {
         log_error("test_pipe_max_active_reservations failed\n");
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
         return -1;
     }
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size, outptr, 1, &sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if( verify_result_int( inptr, outptr, max_active_reservations)){
         log_error("test_pipe_max_active_reservations failed\n");
-        clReleaseMemObject(buf_status);
-        clReleaseMemObject(buf_reserve_id_t_size);
-        clReleaseMemObject(buf_reserve_id_t_size_aligned);
-        clReleaseMemObject(buf_reservations);
-        clReleaseMemObject(pipe);
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free(outptr);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(sync_event);
-        clReleaseProgram(program);
         return -1;
     }
     else {
         log_info("test_pipe_max_active_reservations passed\n");
     }
 
-    //cleanup
-    clReleaseMemObject(buf_status);
-    clReleaseMemObject(buf_reserve_id_t_size);
-    clReleaseMemObject(buf_reserve_id_t_size_aligned);
-    clReleaseMemObject(buf_reservations);
-    clReleaseMemObject(pipe);
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    align_free(outptr);
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseEvent(sync_event);
-    clReleaseProgram(program);
     return 0;
 }
\ No newline at end of file
diff --git a/test_conformance/pipes/test_pipe_query_functions.cpp b/test_conformance/pipes/test_pipe_query_functions.cpp
index c1b4d92..f9c93aa 100644
--- a/test_conformance/pipes/test_pipe_query_functions.cpp
+++ b/test_conformance/pipes/test_pipe_query_functions.cpp
@@ -79,28 +79,32 @@
 
 int test_pipe_query_functions(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
 {
-    cl_mem        pipe;
-    cl_mem      buffers[4];
-    void        *outptr1;
-    void        *outptr2;
-    cl_int        *inptr;
-    cl_program  program;
-    cl_kernel   kernel[3];
-    size_t      global_work_size[3];
-    size_t        half_global_work_size[3];
-    size_t        global_work_size_pipe_query[3];
-    cl_int        pipe_max_packets, pipe_num_packets;
-    cl_int      err;
-    cl_int        size;
-    cl_int      i;
-    cl_event    producer_sync_event = NULL;
-    cl_event    consumer_sync_event = NULL;
-    cl_event    pipe_query_sync_event = NULL;
-    cl_event    pipe_read_sync_event = NULL;
-    MTdata      d = init_genrand( gRandomSeed );
-    const char*    kernelName[] = {"test_pipe_write", "test_pipe_read", "test_pipe_query_functions"};
+    clMemWrapper pipe;
+    clMemWrapper buffers[4];
+    void *outptr1;
+    void *outptr2;
+    cl_int *inptr;
+    clProgramWrapper program;
+    clKernelWrapper kernel[3];
+    size_t global_work_size[3];
+    size_t half_global_work_size[3];
+    size_t global_work_size_pipe_query[3];
+    cl_int pipe_max_packets, pipe_num_packets;
+    cl_int err;
+    cl_int size;
+    cl_int i;
+    clEventWrapper producer_sync_event = NULL;
+    clEventWrapper consumer_sync_event = NULL;
+    clEventWrapper pipe_query_sync_event = NULL;
+    clEventWrapper pipe_read_sync_event = NULL;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<cl_int> BufferOutPtr1;
+    BufferOwningPtr<cl_int> BufferOutPtr2;
+    MTdataHolder d(gRandomSeed);
+    const char *kernelName[] = { "test_pipe_write", "test_pipe_read",
+                                 "test_pipe_query_functions" };
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
     size = sizeof(int) * num_elements;
     global_work_size[0] = (cl_uint)num_elements;
@@ -109,98 +113,43 @@
 
     inptr = (int *)align_malloc(size, min_alignment);
 
-    for(i = 0; i < num_elements; i++){
+    for (i = 0; i < num_elements; i++)
+    {
         inptr[i] = TEST_PRIME_INT;
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
 
     buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        print_error(err, " clCreateBuffer failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     outptr1 = align_malloc(size/2, min_alignment);
     outptr2 = align_malloc(size, min_alignment);
+    BufferOutPtr1.reset(outptr1, nullptr, 0, size, true);
+    BufferOutPtr2.reset(outptr2, nullptr, 0, size, true);
+
     buffers[1] = clCreateBuffer(context, CL_MEM_HOST_READ_ONLY,  size, NULL, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr1 );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     buffers[2] = clCreateBuffer(context, CL_MEM_READ_WRITE,  sizeof(int), NULL, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        align_free( outptr1 );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     buffers[3] = clCreateBuffer(context, CL_MEM_READ_WRITE,  sizeof(int), NULL, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        align_free( outptr1 );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, sizeof(int), num_elements, NULL, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        align_free( outptr1 );
-        clReleaseMemObject(pipe);
-        print_error(err, " clCreatePipe failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreatePipe failed", -1);
 
     // Create producer kernel
     err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, (const char**)&pipe_query_functions_kernel_code, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        align_free(outptr1);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating program", -1);
+
     //Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        align_free(outptr1);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
+
     //Create pipe query functions kernel
     kernel[2] = clCreateKernel(program, kernelName[2], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        align_free(outptr1);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     err |= clSetKernelArg(kernel[0], 1, sizeof(cl_mem), (void*)&pipe);
@@ -209,104 +158,21 @@
     err |= clSetKernelArg(kernel[2], 0, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[2], 1, sizeof(cl_mem), (void*)&buffers[2]);
     err |= clSetKernelArg(kernel[2], 2, sizeof(cl_mem), (void*)&buffers[3]);
-
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        print_error(err, " clSetKernelArg failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, NULL, 0, NULL, &producer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     // Launch Pipe query kernel
     err = clEnqueueNDRangeKernel( queue, kernel[2], 1, NULL, global_work_size_pipe_query, NULL, 1, &producer_sync_event, &pipe_query_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, sizeof(cl_int), &pipe_num_packets, 1, &pipe_query_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
-
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[3], true, 0, sizeof(cl_int), &pipe_max_packets, 1, &pipe_query_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if(pipe_num_packets != num_elements || pipe_max_packets != num_elements)
     {
@@ -316,85 +182,20 @@
 
     // Launch Consumer kernel with half the previous global size
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, half_global_work_size, NULL, 1, &producer_sync_event, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size / 2, outptr1, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
+
+    // We will reuse this variable so release the previous referred event.
+    clReleaseEvent(pipe_query_sync_event);
 
     // Launch Pipe query kernel
     err = clEnqueueNDRangeKernel( queue, kernel[2], 1, NULL, global_work_size_pipe_query, NULL, 1, &consumer_sync_event, &pipe_query_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, sizeof(cl_int), &pipe_num_packets, 1, &pipe_query_sync_event, &pipe_read_sync_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     // After consumer kernel consumes num_elements/2 from the pipe,
     // there are (num_elements - num_elements/2) remaining package in the pipe.
@@ -404,68 +205,24 @@
         return -1;
     }
 
+    // We will reuse this variable so release the previous referred event.
+    clReleaseEvent(producer_sync_event);
+
     // Launch Producer kernel to fill the pipe again
     global_work_size[0] = pipe_num_packets;
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, NULL, 1, &pipe_read_sync_event, &producer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
+    // We will reuse this variable so release the previous referred event.
+    clReleaseEvent(pipe_query_sync_event);
     // Launch Pipe query kernel
     err = clEnqueueNDRangeKernel( queue, kernel[2], 1, NULL, global_work_size_pipe_query, NULL, 1, &producer_sync_event, &pipe_query_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
+    // We will reuse this variable so release the previous referred event.
+    clReleaseEvent(pipe_read_sync_event);
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, sizeof(cl_int), &pipe_num_packets, 1, &pipe_query_sync_event, &pipe_read_sync_event);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if(pipe_num_packets != num_elements)
     {
@@ -473,48 +230,16 @@
         return -1;
     }
 
+    // We will reuse this variable so release the previous referred event.
+    clReleaseEvent(consumer_sync_event);
+
     // Launch Consumer kernel to consume all packets from pipe
     global_work_size[0] = pipe_num_packets;
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, NULL, 1, &pipe_read_sync_event, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size, outptr2, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed\n" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(buffers[3]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseKernel(kernel[2]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseEvent(pipe_query_sync_event);
-        clReleaseEvent(pipe_read_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr1);
-        return -1;
-    }
-
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if( verify_result(outptr1, outptr2, num_elements )){
         log_error("test_pipe_query_functions failed\n");
@@ -523,22 +248,6 @@
     else {
         log_info("test_pipe_query_functions passed\n");
     }
-    //cleanup
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    clReleaseMemObject(buffers[2]);
-    clReleaseMemObject(buffers[3]);
-    clReleaseMemObject(pipe);
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseKernel(kernel[2]);
-    clReleaseEvent(producer_sync_event);
-    clReleaseEvent(consumer_sync_event);
-    clReleaseEvent(pipe_query_sync_event);
-    clReleaseEvent(pipe_read_sync_event);
-    clReleaseProgram(program);
-    align_free(outptr1);
-
     return 0;
 }
 
diff --git a/test_conformance/pipes/test_pipe_read_write.cpp b/test_conformance/pipes/test_pipe_read_write.cpp
index 91a4990..4bb4468 100644
--- a/test_conformance/pipes/test_pipe_read_write.cpp
+++ b/test_conformance/pipes/test_pipe_read_write.cpp
@@ -15,11 +15,15 @@
 //
 #include "harness/compat.h"
 
+#include <assert.h>
+#include <iomanip>
+#include <iostream>
+#include <sstream>
 #include <stdio.h>
 #include <string.h>
-#include <sys/types.h>
+#include <string>
 #include <sys/stat.h>
-#include <assert.h>
+#include <sys/types.h>
 
 #include "procs.h"
 #include "kernels.h"
@@ -89,113 +93,139 @@
 static const char* convenience_half_kernel_name[] = { "test_pipe_convenience_write_half", "test_pipe_convenience_read_half", "test_pipe_convenience_write_half2", "test_pipe_convenience_read_half2", "test_pipe_convenience_write_half4", "test_pipe_convenience_read_half4", "test_pipe_convenience_write_half8", "test_pipe_convenience_read_half8", "test_pipe_convenience_write_half16", "test_pipe_convenience_read_half16" };
 static const char* convenience_double_kernel_name[] = { "test_pipe_convenience_write_double", "test_pipe_convenience_read_double", "test_pipe_convenience_write_double2", "test_pipe_convenience_read_double2", "test_pipe_convenience_write_double4", "test_pipe_convenience_read_double4", "test_pipe_convenience_write_double8", "test_pipe_convenience_read_double8", "test_pipe_convenience_write_double16", "test_pipe_convenience_read_double16" };
 
-static void insertPragmaForHalfType(char *source, char *type)
+static void insertPragmaForHalfType(std::stringstream &stream, char *type)
 {
-    source[0] = 0;
-    if(strncmp(type, "half",4) == 0)
+    if (strncmp(type, "half", 4) == 0)
     {
-        strcat(source, "#pragma OPENCL EXTENSION cl_khr_fp16 : enable\n");
+        stream << "#pragma OPENCL EXTENSION cl_khr_fp16 : enable\n";
     }
 }
 
-void createKernelSource(char *source, char *type)
+void createKernelSource(std::stringstream &stream, char *type)
 {
-    char str[512];
-    int     str_length;
+    insertPragmaForHalfType(stream, type);
 
-    insertPragmaForHalfType(source, type);
+    // clang-format off
+    stream << R"(
+        __kernel void test_pipe_write_)" << type << "(__global " << type << " *src, __write_only pipe " << type << R"( out_pipe)
+        {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
 
-    sprintf(str, "__kernel void test_pipe_write_%s(__global %s *src, __write_only pipe %s out_pipe)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  reserve_id_t res_id; \n\n");
-    strcat(source, str);
-    sprintf(str, "  res_id = reserve_write_pipe(out_pipe, 1);\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    write_pipe(out_pipe, res_id, 0, &src[gid]);\n    commit_write_pipe(out_pipe, res_id);\n  }\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void test_pipe_read_%s(__read_only pipe %s in_pipe, __global %s *dst)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  reserve_id_t res_id; \n\n");
-    strcat(source, str);
-    sprintf(str, "  res_id = reserve_read_pipe(in_pipe, 1);\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    read_pipe(in_pipe, res_id, 0, &dst[gid]);\n    commit_read_pipe(in_pipe, res_id);\n  }\n}\n");
-    strcat(source, str);
-    str_length = strlen(source);
-    assert(str_length <= STRING_LENGTH);
+            res_id = reserve_write_pipe(out_pipe, 1);
+            if(is_valid_reserve_id(res_id))
+            {
+                write_pipe(out_pipe, res_id, 0, &src[gid]);
+                commit_write_pipe(out_pipe, res_id);
+            }
+        }
+
+        __kernel void test_pipe_read_)" << type << "(__read_only pipe " << type << " in_pipe, __global " << type << R"( *dst)
+        {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
+
+            res_id = reserve_read_pipe(in_pipe, 1);
+            if(is_valid_reserve_id(res_id))
+            {
+                read_pipe(in_pipe, res_id, 0, &dst[gid]);
+                commit_read_pipe(in_pipe, res_id);
+            }
+        }
+        )";
+    // clang-format on
 }
 
-void createKernelSourceWorkGroup(char *source, char *type)
+void createKernelSourceWorkGroup(std::stringstream &stream, char *type)
 {
-    char str[512];
-    int     str_length;
+    insertPragmaForHalfType(stream, type);
 
-    insertPragmaForHalfType(source, type);
+    // clang-format off
+    stream << R"(
+        __kernel void test_pipe_workgroup_write_)" << type << "(__global " << type << " *src, __write_only pipe " << type << R"( out_pipe)
+        {
+            int gid = get_global_id(0);
+            __local reserve_id_t res_id;
 
-    sprintf(str, "__kernel void test_pipe_workgroup_write_%s(__global %s *src, __write_only pipe %s out_pipe)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  __local reserve_id_t res_id; \n\n");
-    strcat(source, str);
-    sprintf(str, "  res_id = work_group_reserve_write_pipe(out_pipe, get_local_size(0));\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    write_pipe(out_pipe, res_id, get_local_id(0), &src[gid]);\n    work_group_commit_write_pipe(out_pipe, res_id);\n  }\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void test_pipe_workgroup_read_%s(__read_only pipe %s in_pipe, __global %s *dst)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  __local reserve_id_t res_id; \n\n");
-    strcat(source, str);
-    sprintf(str, "  res_id = work_group_reserve_read_pipe(in_pipe, get_local_size(0));\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    read_pipe(in_pipe, res_id, get_local_id(0), &dst[gid]);\n    work_group_commit_read_pipe(in_pipe, res_id);\n  }\n}\n");
-    strcat(source, str);
-    str_length = strlen(source);
-    assert(str_length <= STRING_LENGTH);
+            res_id = work_group_reserve_write_pipe(out_pipe, get_local_size(0));
+            if(is_valid_reserve_id(res_id))
+            {
+                write_pipe(out_pipe, res_id, get_local_id(0), &src[gid]);
+                work_group_commit_write_pipe(out_pipe, res_id);
+            }
+        }
+
+        __kernel void test_pipe_workgroup_read_)" << type << "(__read_only pipe " << type << " in_pipe, __global " << type << R"( *dst)
+        {
+            int gid = get_global_id(0);
+            __local reserve_id_t res_id;
+
+            res_id = work_group_reserve_read_pipe(in_pipe, get_local_size(0));
+            if(is_valid_reserve_id(res_id))
+            {
+                read_pipe(in_pipe, res_id, get_local_id(0), &dst[gid]);
+                work_group_commit_read_pipe(in_pipe, res_id);
+            }
+        }
+        )";
+    // clang-format on
 }
 
-void createKernelSourceSubGroup(char *source, char *type)
+void createKernelSourceSubGroup(std::stringstream &stream, char *type)
 {
-    char str[512];
-    int  str_length;
+    insertPragmaForHalfType(stream, type);
 
-    insertPragmaForHalfType(source, type);
+    // clang-format off
+    stream << R"(
+        #pragma OPENCL EXTENSION cl_khr_subgroups : enable
+        __kernel void test_pipe_subgroup_write_)" << type << "(__global " << type << " *src, __write_only pipe " << type << R"( out_pipe)
+        {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
 
-    sprintf(str, "#pragma OPENCL EXTENSION cl_khr_subgroups : enable\n__kernel void test_pipe_subgroup_write_%s(__global %s *src, __write_only pipe %s out_pipe)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  reserve_id_t res_id; \n\n");
-    strcat(source, str);
-    sprintf(str, "  res_id = sub_group_reserve_write_pipe(out_pipe, get_sub_group_size());\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    write_pipe(out_pipe, res_id, get_sub_group_local_id(), &src[gid]);\n    sub_group_commit_write_pipe(out_pipe, res_id);\n  }\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void test_pipe_subgroup_read_%s(__read_only pipe %s in_pipe, __global %s *dst)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  reserve_id_t res_id; \n\n");
-    strcat(source, str);
-    sprintf(str, "  res_id = sub_group_reserve_read_pipe(in_pipe, get_sub_group_size());\n  if(is_valid_reserve_id(res_id))\n  {\n");
-    strcat(source, str);
-    sprintf(str, "    read_pipe(in_pipe, res_id, get_sub_group_local_id(), &dst[gid]);\n    sub_group_commit_read_pipe(in_pipe, res_id);\n  }\n}\n");
-    strcat(source, str);
-    str_length = strlen(source);
-    assert(str_length <= STRING_LENGTH);
+            res_id = sub_group_reserve_write_pipe(out_pipe, get_sub_group_size());
+            if(is_valid_reserve_id(res_id))
+            {
+                write_pipe(out_pipe, res_id, get_sub_group_local_id(), &src[gid]);
+                sub_group_commit_write_pipe(out_pipe, res_id);
+            }
+        }
+
+        __kernel void test_pipe_subgroup_read_)" << type << "(__read_only pipe " << type << " in_pipe, __global " << type << R"( *dst)
+        {
+            int gid = get_global_id(0);
+            reserve_id_t res_id;
+
+            res_id = sub_group_reserve_read_pipe(in_pipe, get_sub_group_size());
+            if(is_valid_reserve_id(res_id))
+            {
+                read_pipe(in_pipe, res_id, get_sub_group_local_id(), &dst[gid]);
+                sub_group_commit_read_pipe(in_pipe, res_id);
+            }
+        }
+        )";
+    // clang-format on
 }
 
-void createKernelSourceConvenience(char *source, char *type)
+void createKernelSourceConvenience(std::stringstream &stream, char *type)
 {
-    char str[512];
-    int     str_length;
+    insertPragmaForHalfType(stream, type);
 
-    insertPragmaForHalfType(source, type);
+    // clang-format off
+    stream << R"(
+        __kernel void test_pipe_convenience_write_)" << type << "(__global " << type << " *src, __write_only pipe " << type << R"( out_pipe)
+        {
+            int gid = get_global_id(0);
+            write_pipe(out_pipe, &src[gid]);
+        }
 
-    sprintf(str, "__kernel void test_pipe_convenience_write_%s(__global %s *src, __write_only pipe %s out_pipe)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  write_pipe(out_pipe, &src[gid]);\n}\n\n");
-    strcat(source, str);
-    sprintf(str, "__kernel void test_pipe_convenience_read_%s(__read_only pipe %s in_pipe, __global %s *dst)\n", type, type, type);
-    strcat(source, str);
-    sprintf(str, "{\n  int gid = get_global_id(0);\n  read_pipe(in_pipe, &dst[gid]);\n}\n");
-    strcat(source, str);
-    str_length = strlen(source);
-    assert(str_length <= STRING_LENGTH);
+        __kernel void test_pipe_convenience_read_)" << type << "(__read_only pipe " << type << " in_pipe, __global " << type << R"( *dst)
+        {
+            int gid = get_global_id(0);
+            read_pipe(in_pipe, &dst[gid]);
+        }
+        )";
+    // clang-format on
 }
 
 // verify functions
@@ -424,23 +454,24 @@
 int test_pipe_readwrite( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements, size_t size, char *type, int loops,
                          void *inptr[5], const char *kernelName[], int (*fn)(void *, void *, int) )
 {
-    cl_mem        pipes[5];
-    cl_mem      buffers[10];
-    void        *outptr[5];
-    cl_program  program[5];
-    cl_kernel   kernel[10];
-    size_t      global_work_size[3];
-    size_t        local_work_size[3];
-    cl_int      err;
-    int         i, ii;
-    size_t      ptrSizes[5];
-    int         total_errors = 0;
-    cl_event    producer_sync_event[5];
-    cl_event    consumer_sync_event[5];
-    char        *sourceCode[5];
-    char        vector_type[10];
+    clMemWrapper pipes[5];
+    clMemWrapper buffers[10];
+    void *outptr[5];
+    BufferOwningPtr<cl_int> BufferOutPtr[5];
+    clProgramWrapper program[5];
+    clKernelWrapper kernel[10];
+    size_t global_work_size[3];
+    size_t local_work_size[3];
+    cl_int err;
+    int i, ii;
+    size_t ptrSizes[5];
+    int total_errors = 0;
+    clEventWrapper producer_sync_event[5];
+    clEventWrapper consumer_sync_event[5];
+    std::stringstream sourceCode[5];
+    char vector_type[10];
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
     global_work_size[0] = (cl_uint)num_elements;
 
@@ -450,217 +481,133 @@
     ptrSizes[3] = ptrSizes[2] << 1;
     ptrSizes[4] = ptrSizes[3] << 1;
 
-    for( i = 0; i < loops; i++)
+    for (i = 0; i < loops; i++)
     {
         ii = i << 1;
-        sourceCode[i] = (char*) malloc(STRING_LENGTH * sizeof(char));
-        buffers[ii] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, ptrSizes[i] * num_elements, inptr[i], &err);
-        if(err){
-            clReleaseMemObject(buffers[ii]);
-            align_free( outptr[i] );
-            print_error(err, " clCreateBuffer failed\n");
-            return -1;
-        }
-        outptr[i] = align_malloc( ptrSizes[i] * num_elements, min_alignment);
-        buffers[ii+1] = clCreateBuffer(context, CL_MEM_USE_HOST_PTR,  ptrSizes[i] * num_elements, outptr[i], &err);
 
-        if ( err ){
-            clReleaseMemObject(buffers[ii]);
-            align_free( outptr[i] );
-            print_error(err, " clCreateBuffer failed\n" );
-            return -1;
-        }
+        buffers[ii] =
+            clCreateBuffer(context, CL_MEM_COPY_HOST_PTR,
+                           ptrSizes[i] * num_elements, inptr[i], &err);
+        test_error_ret(err, " clCreateBuffer failed", -1);
+
+        outptr[i] = align_malloc(ptrSizes[i] * num_elements, min_alignment);
+        BufferOutPtr[i].reset(outptr[i], nullptr, 0, size, true);
+        buffers[ii + 1] =
+            clCreateBuffer(context, CL_MEM_USE_HOST_PTR,
+                           ptrSizes[i] * num_elements, outptr[i], &err);
+        test_error_ret(err, " clCreateBuffer failed", -1);
+
         // Creating pipe with non-power of 2 size
-        pipes[i] = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, ptrSizes[i], num_elements+3, NULL, &err);
-        if(err){
-            clReleaseMemObject(pipes[i]);
-            print_error(err, " clCreatePipe failed\n");
-            return -1;
-        }
+        pipes[i] = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, ptrSizes[i],
+                                num_elements + 3, NULL, &err);
+        test_error_ret(err, " clCreatePipe failed", -1);
 
-        switch(i)
+        switch (i)
         {
-        case 0:
-            sprintf(vector_type, "%s", type);
-            break;
-        case 1:
-            sprintf(vector_type, "%s%d", type, 2);
-            break;
-        case 2:
-            sprintf(vector_type, "%s%d", type, 4);
-            break;
-        case 3:
-            sprintf(vector_type, "%s%d", type, 8);
-            break;
-        case 4:
-            sprintf(vector_type, "%s%d", type, 16);
-            break;
+            case 0: sprintf(vector_type, "%s", type); break;
+            case 1: sprintf(vector_type, "%s%d", type, 2); break;
+            case 2: sprintf(vector_type, "%s%d", type, 4); break;
+            case 3: sprintf(vector_type, "%s%d", type, 8); break;
+            case 4: sprintf(vector_type, "%s%d", type, 16); break;
         }
 
-        if(useWorkgroupReserve == 1){
+        if (useWorkgroupReserve == 1)
+        {
             createKernelSourceWorkGroup(sourceCode[i], vector_type);
         }
-        else if(useSubgroupReserve == 1){
+        else if (useSubgroupReserve == 1)
+        {
             createKernelSourceSubGroup(sourceCode[i], vector_type);
         }
-        else if(useConvenienceBuiltIn == 1){
+        else if (useConvenienceBuiltIn == 1)
+        {
             createKernelSourceConvenience(sourceCode[i], vector_type);
         }
-        else{
+        else
+        {
             createKernelSource(sourceCode[i], vector_type);
         }
 
+        std::string kernel_source = sourceCode[i].str();
+        const char *sources[] = { kernel_source.c_str() };
         // Create producer kernel
-        err = create_single_kernel_helper_with_build_options(context, &program[i], &kernel[ii], 1, (const char**)&sourceCode[i], kernelName[ii], "-cl-std=CL2.0");
-        if(err){
-            clReleaseMemObject(buffers[ii]);
-            clReleaseMemObject(buffers[ii+1]);
-            clReleaseMemObject(pipes[i]);
-            align_free( outptr[i] );
-            print_error(err, "Error creating program\n");
-            return -1;
-        }
-        //Create consumer kernel
+        err = create_single_kernel_helper_with_build_options(
+            context, &program[i], &kernel[ii], 1, sources, kernelName[ii],
+            "-cl-std=CL2.0");
+
+        test_error_ret(err, " Error creating program", -1);
+
+        // Create consumer kernel
         kernel[ii + 1] = clCreateKernel(program[i], kernelName[ii + 1], &err);
-        if( kernel[ii + 1] == NULL || err != CL_SUCCESS)
-        {
-            clReleaseMemObject(buffers[ii]);
-            clReleaseMemObject(buffers[ii+1]);
-            clReleaseMemObject(pipes[i]);
-            align_free( outptr[i] );
-            log_error("Creating program for %s\n", type);
-            print_error( err, "Unable to create kernel" );
-            return -1;
-        }
+        test_error_ret(err, " Error creating kernel", -1);
 
-        err = clSetKernelArg(kernel[ii], 0, sizeof(cl_mem), (void*)&buffers[ii]);
-        err |= clSetKernelArg(kernel[ii], 1, sizeof(cl_mem), (void*)&pipes[i]);
-        err |= clSetKernelArg(kernel[ii + 1], 0, sizeof(cl_mem), (void*)&pipes[i]);
-        err |= clSetKernelArg(kernel[ii + 1], 1, sizeof(cl_mem), (void*)&buffers[ii + 1]);
-        if ( err != CL_SUCCESS ){
-            clReleaseMemObject(buffers[ii]);
-            clReleaseMemObject(buffers[ii+1]);
-            clReleaseMemObject(pipes[i]);
-            clReleaseKernel(kernel[ii]);
-            clReleaseKernel(kernel[ii+1]);
-            clReleaseProgram(program[i]);
-            align_free(outptr[i]);
-            print_error(err, " clSetKernelArg failed");
-            return -1;
-        }
+        err =
+            clSetKernelArg(kernel[ii], 0, sizeof(cl_mem), (void *)&buffers[ii]);
+        err |= clSetKernelArg(kernel[ii], 1, sizeof(cl_mem), (void *)&pipes[i]);
+        err |= clSetKernelArg(kernel[ii + 1], 0, sizeof(cl_mem),
+                              (void *)&pipes[i]);
+        err |= clSetKernelArg(kernel[ii + 1], 1, sizeof(cl_mem),
+                              (void *)&buffers[ii + 1]);
+        test_error_ret(err, " clSetKernelArg failed", -1);
 
-        if(useWorkgroupReserve == 1 || useSubgroupReserve == 1)
+        if (useWorkgroupReserve == 1 || useSubgroupReserve == 1)
         {
-            err = get_max_common_work_group_size( context, kernel[ii], global_work_size[0], &local_work_size[0] );
-            test_error( err, "Unable to get work group size to use" );
+            err = get_max_common_work_group_size(
+                context, kernel[ii], global_work_size[0], &local_work_size[0]);
+            test_error(err, "Unable to get work group size to use");
             // Launch Producer kernel
-            err = clEnqueueNDRangeKernel( queue, kernel[ii], 1, NULL, global_work_size, local_work_size, 0, NULL, &producer_sync_event[i] );
-            if ( err != CL_SUCCESS ){
-                print_error( err, " clEnqueueNDRangeKernel failed" );
-                clReleaseMemObject(buffers[ii]);
-                clReleaseMemObject(buffers[ii+1]);
-                clReleaseMemObject(pipes[i]);
-                clReleaseKernel(kernel[ii]);
-                clReleaseKernel(kernel[ii+1]);
-                clReleaseEvent(producer_sync_event[i]);
-                clReleaseEvent(consumer_sync_event[i]);
-                clReleaseProgram(program[i]);
-                align_free(outptr[i]);
-                return -1;
-            }
+            err = clEnqueueNDRangeKernel(queue, kernel[ii], 1, NULL,
+                                         global_work_size, local_work_size, 0,
+                                         NULL, &producer_sync_event[i]);
+            test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
         }
         else
         {
             // Launch Producer kernel
-            err = clEnqueueNDRangeKernel( queue, kernel[ii], 1, NULL, global_work_size, NULL, 0, NULL, &producer_sync_event[i] );
-            if ( err != CL_SUCCESS ){
-                print_error( err, " clEnqueueNDRangeKernel failed" );
-                clReleaseMemObject(buffers[ii]);
-                clReleaseMemObject(buffers[ii+1]);
-                clReleaseMemObject(pipes[i]);
-                clReleaseKernel(kernel[ii]);
-                clReleaseKernel(kernel[ii+1]);
-                clReleaseEvent(producer_sync_event[i]);
-                clReleaseEvent(consumer_sync_event[i]);
-                clReleaseProgram(program[i]);
-                align_free(outptr[i]);
-                return -1;
-            }
+            err = clEnqueueNDRangeKernel(queue, kernel[ii], 1, NULL,
+                                         global_work_size, NULL, 0, NULL,
+                                         &producer_sync_event[i]);
+            test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
         }
 
-        if(useWorkgroupReserve == 1 || useSubgroupReserve == 1)
+        if (useWorkgroupReserve == 1 || useSubgroupReserve == 1)
         {
-            err = get_max_common_work_group_size( context, kernel[ii + 1], global_work_size[0], &local_work_size[0] );
-            test_error( err, "Unable to get work group size to use" );
+            err = get_max_common_work_group_size(context, kernel[ii + 1],
+                                                 global_work_size[0],
+                                                 &local_work_size[0]);
+            test_error(err, "Unable to get work group size to use");
 
             // Launch Consumer kernel
-            err = clEnqueueNDRangeKernel( queue, kernel[ii + 1], 1, NULL, global_work_size, local_work_size, 1, &producer_sync_event[i], &consumer_sync_event[i] );
-            if ( err != CL_SUCCESS ){
-                print_error( err, " clEnqueueNDRangeKernel failed" );
-                clReleaseMemObject(buffers[ii]);
-                clReleaseMemObject(buffers[ii+1]);
-                clReleaseMemObject(pipes[i]);
-                clReleaseKernel(kernel[ii]);
-                clReleaseKernel(kernel[ii+1]);
-                clReleaseEvent(producer_sync_event[i]);
-                clReleaseEvent(consumer_sync_event[i]);
-                clReleaseProgram(program[i]);
-                align_free(outptr[i]);
-                return -1;
-            }
+            err = clEnqueueNDRangeKernel(queue, kernel[ii + 1], 1, NULL,
+                                         global_work_size, local_work_size, 1,
+                                         &producer_sync_event[i],
+                                         &consumer_sync_event[i]);
+            test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
         }
         else
         {
             // Launch Consumer kernel
-            err = clEnqueueNDRangeKernel( queue, kernel[ii + 1], 1, NULL, global_work_size, NULL, 1,  &producer_sync_event[i], &consumer_sync_event[i] );
-            if ( err != CL_SUCCESS ){
-                print_error( err, " clEnqueueNDRangeKernel failed" );
-                clReleaseMemObject(buffers[ii]);
-                clReleaseMemObject(buffers[ii+1]);
-                clReleaseMemObject(pipes[i]);
-                clReleaseKernel(kernel[ii]);
-                clReleaseKernel(kernel[ii+1]);
-                clReleaseEvent(producer_sync_event[i]);
-                clReleaseEvent(consumer_sync_event[i]);
-                clReleaseProgram(program[i]);
-                align_free(outptr[i]);
-                return -1;
-            }
+            err = clEnqueueNDRangeKernel(
+                queue, kernel[ii + 1], 1, NULL, global_work_size, NULL, 1,
+                &producer_sync_event[i], &consumer_sync_event[i]);
+            test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
         }
 
-        err = clEnqueueReadBuffer(queue, buffers[ii+1], true, 0, ptrSizes[i]*num_elements, outptr[i], 1, &consumer_sync_event[i], NULL);
-        if ( err != CL_SUCCESS ){
-            print_error( err, " clEnqueueReadBuffer failed" );
-            clReleaseMemObject(buffers[ii]);
-            clReleaseMemObject(buffers[ii+1]);
-            clReleaseMemObject(pipes[i]);
-            clReleaseKernel(kernel[ii]);
-            clReleaseKernel(kernel[ii+1]);
-            clReleaseEvent(producer_sync_event[i]);
-            clReleaseEvent(consumer_sync_event[i]);
-            clReleaseProgram(program[i]);
-            align_free(outptr[i]);
-            return -1;
-        }
+        err = clEnqueueReadBuffer(queue, buffers[ii + 1], true, 0,
+                                  ptrSizes[i] * num_elements, outptr[i], 1,
+                                  &consumer_sync_event[i], NULL);
+        test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
-        if( fn( inptr[i], outptr[i], (int)(ptrSizes[i] * (size_t)num_elements / ptrSizes[0]))){
-            log_error("%s%d test failed\n", type, 1<<i);
+        if (fn(inptr[i], outptr[i],
+               (int)(ptrSizes[i] * (size_t)num_elements / ptrSizes[0])))
+        {
+            log_error("%s%d test failed\n", type, 1 << i);
             total_errors++;
         }
-        else {
-            log_info("%s%d test passed\n", type, 1<<i);
+        else
+        {
+            log_info("%s%d test passed\n", type, 1 << i);
         }
-        //cleanup
-        clReleaseMemObject(buffers[ii]);
-        clReleaseMemObject(buffers[ii+1]);
-        clReleaseMemObject(pipes[i]);
-        clReleaseKernel(kernel[ii]);
-        clReleaseKernel(kernel[ii+1]);
-        clReleaseEvent(producer_sync_event[i]);
-        clReleaseEvent(consumer_sync_event[i]);
-        clReleaseProgram(program[i]);
-        align_free(outptr[i]);
-
     }
 
     return total_errors;
@@ -669,166 +616,80 @@
 int test_pipe_readwrite_struct_generic( cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements,
                                 const char *kernelCode, const char *kernelName[])
 {
-    cl_mem        buffers[2];
-    cl_mem        pipe;
-    void        *outptr;
-    TestStruct    *inptr;
-    cl_program    program;
-    cl_kernel    kernel[2];
-    size_t        size = sizeof(TestStruct);
-    size_t      global_work_size[3];
-    cl_int      err;
-    int         total_errors = 0;
-    int            i;
-    MTdata      d = init_genrand( gRandomSeed );
-    cl_event    producer_sync_event = NULL;
-    cl_event    consumer_sync_event = NULL;
+    clMemWrapper buffers[2];
+    clMemWrapper pipe;
+    void *outptr;
+    TestStruct *inptr;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<TestStruct> BufferOutPtr;
+    clProgramWrapper program;
+    clKernelWrapper kernel[2];
+    size_t size = sizeof(TestStruct);
+    size_t global_work_size[3];
+    cl_int err;
+    int total_errors = 0;
+    int i;
+    MTdataHolder d(gRandomSeed);
+    clEventWrapper producer_sync_event = NULL;
+    clEventWrapper consumer_sync_event = NULL;
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
     global_work_size[0] = (size_t)num_elements;
 
     inptr = (TestStruct *)align_malloc(size * num_elements, min_alignment);
 
-    for ( i = 0; i < num_elements; i++ ){
+    for (i = 0; i < num_elements; i++)
+    {
         inptr[i].a = (char)genrand_int32(d);
         inptr[i].b = genrand_int32(d);
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
+
     buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size * num_elements, inptr, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        print_error(err, " clCreateBuffer failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     outptr = align_malloc( size * num_elements, min_alignment);
+    BufferOutPtr.reset(outptr, nullptr, 0, size, true);
+
     buffers[1] = clCreateBuffer(context, CL_MEM_USE_HOST_PTR,  size * num_elements, outptr, &err);
-    if (err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, size, num_elements, NULL, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        clReleaseMemObject(pipe);
-        print_error(err, " clCreatePipe failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreatePipe failed", -1);
+
     // Create producer kernel
     err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, &kernelCode, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        log_error(" Error creating program for struct\n");
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating program", -1);
+
     //Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     err |= clSetKernelArg(kernel[0], 1, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 0, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 1, sizeof(cl_mem), (void*)&buffers[1]);
-    if (err != CL_SUCCESS){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, NULL, 0, NULL, &producer_sync_event );
-    if (err != CL_SUCCESS){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     // Launch Consumer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, NULL, 1, &producer_sync_event, &consumer_sync_event );
-    if (err != CL_SUCCESS){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size*num_elements, outptr, 1, &consumer_sync_event, NULL);
-    if (err != CL_SUCCESS){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if( verify_readwrite_struct( inptr, outptr, num_elements)){
         log_error("struct_readwrite test failed\n");
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
         return -1;
     }
     else {
         log_info("struct_readwrite test passed\n");
     }
-    //cleanup
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    clReleaseMemObject(pipe);
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseEvent(producer_sync_event);
-    clReleaseEvent(consumer_sync_event);
-    clReleaseProgram(program);
-    align_free(outptr);
 
     return 0;
 }
diff --git a/test_conformance/pipes/test_pipe_readwrite_errors.cpp b/test_conformance/pipes/test_pipe_readwrite_errors.cpp
index cdfaf5e..1b9fc31 100644
--- a/test_conformance/pipes/test_pipe_readwrite_errors.cpp
+++ b/test_conformance/pipes/test_pipe_readwrite_errors.cpp
@@ -64,23 +64,26 @@
 
 int test_pipe_readwrite_errors(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
 {
-    cl_mem        pipe;
-    cl_mem      buffers[3];
-    void        *outptr;
-    cl_int        *inptr;
-    cl_program  program;
-    cl_kernel   kernel[2];
-    size_t      global_work_size[3];
-    cl_int      err;
-    cl_int        size;
-    cl_int      i;
-    cl_int        status = 0;
-    cl_event    producer_sync_event;
-    cl_event    consumer_sync_event;
-    MTdata      d = init_genrand( gRandomSeed );
-    const char*    kernelName[] = {"test_pipe_write_error", "test_pipe_read_error"};
+    clMemWrapper pipe;
+    clMemWrapper buffers[3];
+    void *outptr;
+    cl_int *inptr;
+    clProgramWrapper program;
+    clKernelWrapper kernel[2];
+    size_t global_work_size[3];
+    cl_int err;
+    cl_int size;
+    cl_int i;
+    cl_int status = 0;
+    clEventWrapper producer_sync_event;
+    clEventWrapper consumer_sync_event;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<cl_int> BufferOutPtr;
+    MTdataHolder d(gRandomSeed);
+    const char *kernelName[] = { "test_pipe_write_error",
+                                 "test_pipe_read_error" };
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
     global_work_size[0] = num_elements;
 
@@ -88,69 +91,36 @@
 
     inptr = (cl_int *)align_malloc(size, min_alignment);
 
-    for(i = 0; i < (cl_int)(size / sizeof(int)); i++){
+    for (i = 0; i < num_elements; i++)
+    {
         inptr[i] = (int)genrand_int32(d);
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
 
-    buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        print_error(err, " clCreateBuffer failed\n");
-        return -1;
-    }
+    buffers[0] =
+        clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     outptr = align_malloc(size, min_alignment);
+    BufferOutPtr.reset(outptr, nullptr, 0, size, true);
+
     buffers[1] = clCreateBuffer(context, CL_MEM_USE_HOST_PTR,  size, outptr, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     buffers[2] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR,  sizeof(int), &status, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        align_free( outptr );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
+
     //Pipe created with max_packets less than global size
     pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, sizeof(int), num_elements - (num_elements/2), NULL, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        align_free( outptr );
-        clReleaseMemObject(pipe);
-        print_error(err, " clCreatePipe failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreatePipe failed", -1);
 
     // Create producer kernel
     err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, (const char**)&pipe_readwrite_errors_kernel_code, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating program", -1);
+
     //Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     err |= clSetKernelArg(kernel[0], 1, sizeof(cl_mem), (void*)&pipe);
@@ -158,49 +128,15 @@
     err |= clSetKernelArg(kernel[1], 0, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 1, sizeof(cl_mem), (void*)&buffers[1]);
     err |= clSetKernelArg(kernel[1], 2, sizeof(cl_mem), (void*)&buffers[2]);
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     // Launch Consumer kernel for empty pipe
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, NULL, 0, NULL, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, sizeof(status), &status, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if(status == 0){
         log_error("test_pipe_readwrite_errors failed\n");
@@ -212,34 +148,13 @@
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, NULL, 0, NULL, &producer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, sizeof(status), &status, 1, &producer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
-    if(status == 0){
+    if (status == 0)
+    {
         log_error("test_pipe_readwrite_errors failed\n");
         return -1;
     }
@@ -247,66 +162,27 @@
         status = 0;
     }
 
+    // We will reuse this variable so release the previous referred event.
+    clReleaseEvent(consumer_sync_event);
+
     // Launch Consumer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, NULL, 1, &producer_sync_event, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, sizeof(status), &status, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
-    if(status == 0)
+    if (status == 0)
     {
         log_error("test_pipe_readwrite_errors failed\n");
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
         return -1;
     }
+    else
+    {
+        status = 0;
+    }
 
     log_info("test_pipe_readwrite_errors passed\n");
-    //cleanup
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    clReleaseMemObject(buffers[2]);
-    clReleaseMemObject(pipe);
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseEvent(producer_sync_event);
-    clReleaseEvent(consumer_sync_event);
-    clReleaseProgram(program);
-    align_free(outptr);
+
     return 0;
 }
diff --git a/test_conformance/pipes/test_pipe_subgroups.cpp b/test_conformance/pipes/test_pipe_subgroups.cpp
index 35519b0..b41170c 100644
--- a/test_conformance/pipes/test_pipe_subgroups.cpp
+++ b/test_conformance/pipes/test_pipe_subgroups.cpp
@@ -88,30 +88,35 @@
 
 int test_pipe_subgroups_divergence(cl_device_id deviceID, cl_context context, cl_command_queue queue, int num_elements)
 {
-    cl_mem        pipe;
-    cl_mem      buffers[3];
-    cl_int        *outptr;
-    cl_int        *inptr;
-    cl_int        *active_work_item_buffer;
-    cl_program  program;
-    cl_kernel   kernel[2];
-    size_t      global_work_size[3];
-    size_t      local_work_size[3];
-    cl_int      err;
-    cl_int        size;
-    int         i;
-    size_t      subgroup_count;
-    cl_event    producer_sync_event = NULL;
-    cl_event    consumer_sync_event = NULL;
-    const char*    kernelName[] = {"test_pipe_subgroups_divergence_write", "test_pipe_subgroups_divergence_read"};
+    clMemWrapper pipe;
+    clMemWrapper buffers[3];
+    cl_int *outptr;
+    cl_int *inptr;
+    cl_int *active_work_item_buffer;
+    clProgramWrapper program;
+    clKernelWrapper kernel[2];
+    size_t global_work_size[3];
+    size_t local_work_size[3];
+    cl_int err;
+    cl_int size;
+    int i;
+    size_t subgroup_count;
+    clEventWrapper producer_sync_event = NULL;
+    clEventWrapper consumer_sync_event = NULL;
+    BufferOwningPtr<cl_int> BufferInPtr;
+    BufferOwningPtr<cl_int> BufferOutPtr;
+    const char *kernelName[] = { "test_pipe_subgroups_divergence_write",
+                                 "test_pipe_subgroups_divergence_read" };
 
-    size_t      min_alignment = get_min_alignment(context);
+    size_t min_alignment = get_min_alignment(context);
 
     global_work_size[0] = (cl_uint)num_elements;
 
-    if(!is_extension_available(deviceID, "cl_khr_subgroups"))
+    if (!is_extension_available(deviceID, "cl_khr_subgroups"))
     {
-        log_info("cl_khr_subgroups is not supported on this platoform. Skipping test.\n");
+        log_info(
+            "cl_khr_subgroups is not supported on this platoform. Skipping "
+            "test.\n");
         return CL_SUCCESS;
     }
 
@@ -125,215 +130,77 @@
         outptr[i] = 0;
         active_work_item_buffer[i] = 0;
     }
+    BufferInPtr.reset(inptr, nullptr, 0, size, true);
+    BufferOutPtr.reset(outptr, nullptr, 0, size, true);
 
     buffers[0] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR, size, inptr, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        print_error(err, " clCreateBuffer failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     buffers[1] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR,  size, outptr, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        align_free( outptr );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     buffers[2] = clCreateBuffer(context, CL_MEM_COPY_HOST_PTR,  size, active_work_item_buffer, &err);
-    if ( err ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        align_free( outptr );
-        print_error(err, " clCreateBuffer failed\n" );
-        return -1;
-    }
+    test_error_ret(err, " clCreateBuffer failed", -1);
 
     pipe = clCreatePipe(context, CL_MEM_HOST_NO_ACCESS, sizeof(int), num_elements, NULL, &err);
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        align_free( outptr );
-        clReleaseMemObject(pipe);
-        print_error(err, " clCreatePipe failed\n");
-        return -1;
-    }
+    test_error_ret(err, " clCreatePipe failed", -1);
 
     // Create producer kernel
     err = create_single_kernel_helper_with_build_options(context, &program, &kernel[0], 1, (const char**)&pipe_subgroups_kernel_code, kernelName[0], "-cl-std=CL2.0");
-    if(err){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        print_error(err, "Error creating program\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating program", -1);
+
     //Create consumer kernel
     kernel[1] = clCreateKernel(program, kernelName[1], &err);
-    if( kernel[1] == NULL || err != CL_SUCCESS)
-    {
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        align_free(outptr);
-        print_error(err, "Error creating kernel\n");
-        return -1;
-    }
+    test_error_ret(err, " Error creating kernel", -1);
 
     err = clSetKernelArg(kernel[0], 0, sizeof(cl_mem), (void*)&buffers[0]);
     err |= clSetKernelArg(kernel[0], 1, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[0], 2, sizeof(cl_mem), (void*)&buffers[2]);
     err |= clSetKernelArg(kernel[1], 0, sizeof(cl_mem), (void*)&pipe);
     err |= clSetKernelArg(kernel[1], 1, sizeof(cl_mem), (void*)&buffers[1]);
-    if ( err != CL_SUCCESS ){
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        print_error(err, " clSetKernelArg failed");
-        return -1;
-    }
+    test_error_ret(err, " clSetKernelArg failed", -1);
 
     err = get_max_common_work_group_size( context, kernel[0], global_work_size[0], &local_work_size[0] );
-    if( err != CL_SUCCESS)
-    {
-        test_error( err, "Unable to get work group size to use" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " Unable to get work group size to use", -1);
 
-        cl_platform_id platform;
-        err = clGetDeviceInfo(deviceID, CL_DEVICE_PLATFORM, sizeof(platform), &platform, NULL);
-        clGetKernelSubGroupInfoKHR_fn clGetKernelSubGroupInfoKHR = (clGetKernelSubGroupInfoKHR_fn) clGetExtensionFunctionAddressForPlatform(platform, "clGetKernelSubGroupInfoKHR");
+    cl_platform_id platform;
+    err = clGetDeviceInfo(deviceID, CL_DEVICE_PLATFORM, sizeof(platform),
+                          &platform, NULL);
+    test_error_ret(err, " clGetDeviceInfo failed", -1);
+
+    clGetKernelSubGroupInfoKHR_fn clGetKernelSubGroupInfoKHR =
+        (clGetKernelSubGroupInfoKHR_fn)clGetExtensionFunctionAddressForPlatform(
+            platform, "clGetKernelSubGroupInfoKHR");
 
     err = clGetKernelSubGroupInfoKHR(kernel[0], deviceID, CL_KERNEL_SUB_GROUP_COUNT_FOR_NDRANGE_KHR, sizeof(local_work_size[0]), &local_work_size[0], sizeof(subgroup_count), &subgroup_count, NULL);
+    test_error_ret(err, " clGetKernelSubGroupInfoKHR failed", -1);
     if(subgroup_count <= 1)
     {
         log_info("Only 1 subgroup per workgroup for the kernel. Hence no divergence among subgroups possible. Skipping test.\n");
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseProgram(program);
-        align_free(outptr);
         return CL_SUCCESS;
     }
 
     // Launch Producer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[0], 1, NULL, global_work_size, local_work_size, 0, NULL, &producer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[2], true, 0, size, active_work_item_buffer, 1, &producer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
-
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     // Launch Consumer kernel
     err = clEnqueueNDRangeKernel( queue, kernel[1], 1, NULL, global_work_size, local_work_size, 1, &producer_sync_event, &consumer_sync_event );
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueNDRangeKernel failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueNDRangeKernel failed", -1);
 
     err = clEnqueueReadBuffer(queue, buffers[1], true, 0, size, outptr, 1, &consumer_sync_event, NULL);
-    if ( err != CL_SUCCESS ){
-        print_error( err, " clEnqueueReadBuffer failed" );
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
-        return -1;
-    }
+    test_error_ret(err, " clEnqueueReadBuffer failed", -1);
 
     if( verify_result( active_work_item_buffer, outptr, num_elements)){
         log_error("test_pipe_subgroups_divergence failed\n");
-        clReleaseMemObject(buffers[0]);
-        clReleaseMemObject(buffers[1]);
-        clReleaseMemObject(buffers[2]);
-        clReleaseMemObject(pipe);
-        clReleaseKernel(kernel[0]);
-        clReleaseKernel(kernel[1]);
-        clReleaseEvent(producer_sync_event);
-        clReleaseEvent(consumer_sync_event);
-        clReleaseProgram(program);
-        align_free(outptr);
         return -1;
     }
     else {
         log_info("test_pipe_subgroups_divergence passed\n");
     }
-    //cleanup
-    clReleaseMemObject(buffers[0]);
-    clReleaseMemObject(buffers[1]);
-    clReleaseMemObject(buffers[2]);
-    clReleaseMemObject(pipe);
-    clReleaseKernel(kernel[0]);
-    clReleaseKernel(kernel[1]);
-    clReleaseEvent(producer_sync_event);
-    clReleaseEvent(consumer_sync_event);
-    clReleaseProgram(program);
-    align_free(outptr);
 
     return 0;
 }