example.cl

Go to the documentation of this file.

/* -*- coding: latin-1 -*- */
/** \file examples/kernel/example.cl
 * \brief
 * Simple example of use of assert*() and PRINT*() macros.
 *
 * Piece of assertOpenCL
 * --- GPLv3 --- Copyright (C) 2018 Olivier Pirson
 * --- http://www.opimedia.be/
 * --- September 19, 2018
 */

#include "assert.cl"  // include assert* macros
#include "print.cl"   // include PRINT* macros


/* ***********
 * Functions *
 *************/

/**
 * Example of function without return value
 * that contains assertions.
 */
void
foo(ASSERT_DECLARE_UNIQUE_PARAMS) {  // in debug mode add extra parameters
  assert_n_x(false, -1, 0.5);                        // check assertion and set if failed
  // This is the first assertion that it fails in this example.

  assert_n_x_r(get_global_id(0) == -1, 123, 124.0);  // check assertion and set and return if failed
  // This assertion fails, but only printing,
  // doesn't set values because there has already a previous assertion that has failed.

  assert_n_x(false, 123, 124.0);                     // do nothing or never executed
}


/**
 * Example of function without return value
 * that contains assertions.
 *
 * @param n
 */
void
foo2(unsigned int n  // be careful, do not add comma
     ASSERT_DECLARE_PARAMS) {  // in debug mode add extra parameters
  assert(n == 0);    // check assertion
  assert_r(n == 0);  // check assertion and return if failed
  assert(n == 0);    // do nothing or never executed
}


/**
 * Example of function with a return value
 * that contains assertions.
 *
 * @param n
 *
 * @return n*2
 */
unsigned int
bar(unsigned int n  // be careful, do not add comma
    ASSERT_DECLARE_PARAMS) {  // in debug mode add extra parameters
  assert_r0(n == 0);  // check assertion and return 0 if failed
  assert_r0(n == 0);  // do nothing or never executed

  return n * 2;
}



/* *****************
 * Kernel function *
 *******************/

/**
 * Example of kernel function that contains assertions.
 *
 * @param in
 * @param outs
 */
__kernel
void
example(const unsigned int in,
        __global unsigned int* const outs  // be careful, do not add comma
        ASSERT_DECLARE_PARAMS) {  // in debug mode add extra parameters
  const unsigned int g_id = get_global_id(0);
  const unsigned int l_id = get_local_id(0);


  // Printing example (probably mixed between all work items)
  PRINT2("global, local id: %u, %u\n", g_id, l_id);

  for (unsigned int i = 0; i < 1000000; ++i) {  // artificial wait
    barrier(CLK_GLOBAL_MEM_FENCE | CLK_LOCAL_MEM_FENCE);
  }


  // Some assertion examples that succeed
  assert(in == 666);
  assert_n(in == 666, 42);
  assert_x(in == 666, 3.14);
  assert_n_x(in == 666, 42, 3.14);

  assert_r(in == 666);
  assert_n_r(in == 666, 42);
  assert_x_r(in == 666, 3.14);
  assert_n_x_r(in == 666, 42, 3.14);

  assert_n(sizeof(unsigned int) == 4, sizeof(unsigned int));
  assert_n(sizeof(float) == 4, sizeof(float));

  if (g_id == 0) {  // only the first work item, to avoid mixed printing
    outs[0] = in;
    outs[1] = 42;


    // Some printing examples
    PRINT1("in: %u\n", in);

    PRINT("PRINT message\n");
    PRINT1("PRINT1 %u\n", 42);
    PRINT2("PRINT2 %u %i\n", 42, -1);
    PRINT3("PRINT2 %u %i %f\n", 42, -1, 3.14);

#define NPRINT
#include "print.cl"   // include again PRINT* macros, but with NPRINT defined


    // The following printing examples are disabled
    PRINT("PRINT message\n");
    PRINT1("PRINT1 %u\n", 42);
    PRINT2("PRINT2 %u %i\n", 42, -1);
    PRINT3("PRINT2 %u %i %f\n", 42, -1, 3.14);


    // Call functions that contain assertions
    foo(ASSERT_UNIQUE_PARAMS);  // in debug mode transmits extra parameters

    foo2(42  // be careful, do not add comma
         ASSERT_PARAMS);  // in debug mode transmits extra parameters

    bar(42  // be careful, do not add comma
        ASSERT_PARAMS);  // in debug mode transmits extra parameters


    // Some assertions
    assert(in == 0);                 // assertion failed
    assert_n(in == 0, 42);           // assertion failed
    assert_n(in == 0, -42);          // assertion failed
    assert_x(in == 0, 3.14);         // assertion failed
    assert_n_x(in == 0, 666, 3.14);  // assertion failed


#ifndef NASSERT
# define NASSERT_NOT_DEFINED
#endif

#define NASSERT
#include "assert.cl"   // include again assert* macros, but with NASSERT defined
    assert(in == 0);                 // assertion disabled
    assert_n(in == 0, 42);           // assertion disabled
    assert_n(in == 0, -42);          // assertion disabled
    assert_x(in == 0, 3.14);         // assertion disabled
    assert_n_x(in == 0, 666, 3.14);  // assertion disabled

    assert_r(in == 0);                 // assertion disabled
    assert_n_r(in == 0, 42);           // assertion disabled
    assert_n_r(in == 0, -42);          // assertion disabled
    assert_x_r(in == 0, 3.14);         // assertion disabled
    assert_n_x_r(in == 0, 666, 3.14);  // assertion disabled


#ifdef NASSERT_NOT_DEFINED
# undef NASSERT
#endif
#include "assert.cl"   // include again assert* macros, but with NASSERT not defined

    assert(in == 0);                 // assertion failed
    assert_n(in == 0, 42);           // assertion failed
    assert_n(in == 0, -42);          // assertion failed
    assert_x(in == 0, 3.14);         // assertion failed
    assert_n_x(in == 0, 111, 1.23);  // assertion failed

    last_assert_n_x(in == 0, -222, 2.34);  // assertion failed and set new values

    assert_r(in == 0);                 // assertion failed and return
    assert_n_r(in == 0, 42);           // do nothing or never executed
    assert_n_r(in == 0, -42);          // do nothing or never executed
    assert_x_r(in == 0, 3.14);         // do nothing or never executed
    assert_n_x_r(in == 0, 666, 3.14);  // do nothing or never executed
  }
}