sschaetz

// Choose one of:
#define AURA_BASE_CUDA
#define AURA_BASE_OPENCL
#define AURA_BASE_METAL
#define AURA_BASE_VULKAN

sschaetz

boost::aura::environment e;
boost::aura::device d(AURA_UNIT_TEST_DEVICE, e);
boost::aura::feed f(d);

boost::aura::library l(
    boost::aura::path(
        get_kernel_dir() + "/kernels.al"
    ), 
d);

sschaetz

#if defined AURA_BASE_OPENCL

    #define AURA_KERNEL __kernel
    #define AURA_DEVMEM __global

    inline float alang_fast_cos(float x)
    {
        return native_cos(x);
    }

sschaetz

AURA_KERNEL void add(
    AURA_DEVMEM float *a,
    AURA_DEVMEM float *b, 
    AURA_DEVMEM float *c
#if defined AURA_BASE_OPENCL
)
{
    const uint tid = get_global_id(0);
#else
    ,

sschaetz

class pthreadbarrier
{
  private:
    pthread_barrier_t barr;

  public:
    pthreadbarrier(unsigned int count)
    {
      BOOST_VERIFY(!pthread_barrier_init(&barr, NULL, count));
    }