Example OpenCL program with C++ and CMake

README.md

Example OpenCL program with C++ and CMake

Minimal working example of how to add two vectors on a computation device of choice using OpenCL and C++ API. Tested on OSX. Based mainly on exercise 3 from a wonderful Hands On OpenCL course.

Pre-requisites

• OpenCL 1.1 (or greater)
• C++11 compiler
• CMake 3.7

On never OSX versions some work needs to be done to make C++ header for OpenCl available. 2 solved the issue for me, but I had to put cl.hpp in /Library/Developer/CommandLineTools/SDKs/MacOSX10.14.sdk/System/Library/Frameworks/OpenCL.framework/Headers instead of /System/Library/Frameworks/OpenCL.framework/Headers/.

Build & run

To build run the following commands from the project directory.

$ cmake -H. -B_build      # -DCMAKE_EXPORT_COMPILE_COMMANDS=YES if you need compile_commands.json
$ ( cd _build/ && make )
$ ./_build/vadd           # needs to be run from the same directory as vadd.cl

References

• [1] - http://handsonopencl.github.io/
• [2] - https://rageandqq.github.io/blog/2018/03/09/opencl-mac-cpp.html
• [3] - https://streamhpc.com/blog/2015-09-25/handling-opencl-with-cmake-3-1-and-higher/
• [4] - https://cgold.readthedocs.io/en/latest/

CMakeLists.txt

cmake_minimum_required(VERSION 3.7)

project(opencl_cmake VERSION 0.0.1 LANGUAGES CXX)

add_executable(vadd main)
target_compile_features(vadd PRIVATE cxx_auto_type)

find_package(OpenCL REQUIRED)
target_link_libraries(vadd OpenCL::OpenCL)

main.cpp

#ifdef __APPLE__
#include <OpenCL/cl.hpp>
#else
#include <CL/cl.hpp>
#endif

#include <algorithm>
#include <cstdlib>
#include <fstream>
#include <iostream>
#include <vector>

#define LENGTH (1024)
#define TOL (0.001)

// pick up device type from compiler command line or from the default type
#ifndef DEVICE
#define DEVICE CL_DEVICE_TYPE_DEFAULT
#endif

float gen_random() { return rand() / (float)RAND_MAX; }

std::string load_program(std::string input) {
  std::ifstream stream(input.c_str());
  if (!stream.is_open()) {
    std::cout << "Cannot open file: " << input << std::endl;
    exit(1);
  }
  return std::string(std::istreambuf_iterator<char>(stream),
                     (std::istreambuf_iterator<char>()));
}

int main() {
  // declare host containers
  std::vector<float> h_a(LENGTH);
  std::vector<float> h_b(LENGTH);
  std::vector<float> h_c(LENGTH);

  // declare device containers
  cl::Buffer d_a;
  cl::Buffer d_b;
  cl::Buffer d_c;

  // fill host containers with random numbers
  std::generate(h_a.begin(), h_a.end(), gen_random);
  std::generate(h_b.begin(), h_b.end(), gen_random);

  // create a context
  cl::Context context(DEVICE);

  // load in kernel source, creating a program object for the context
  cl::Program program(context, load_program("vadd.cl"), true);

  // get the command queue
  cl::CommandQueue queue(context);

  // create the kernel functor
  auto vadd =
      cl::make_kernel<cl::Buffer, cl::Buffer, cl::Buffer, int>(program, "vadd");

  // copy data to device
  d_a = cl::Buffer(context, begin(h_a), end(h_a), true);
  d_b = cl::Buffer(context, begin(h_b), end(h_b), true);

  // allocate results container
  d_c = cl::Buffer(context, CL_MEM_WRITE_ONLY, sizeof(float) * LENGTH);

  // run calculations
  vadd(cl::EnqueueArgs(queue, cl::NDRange(LENGTH)), d_a, d_b, d_c, LENGTH);

  queue.finish();

  cl::copy(queue, d_c, begin(h_c), end(h_c));

  // test the results
  int correct = 0;
  float tmp;
  for (int i = 0; i < LENGTH; i++) {
    tmp = h_a[i] + h_b[i];       // expected value for d_c[i]
    tmp -= h_c[i];               // compute errors
    if (tmp * tmp < TOL * TOL) { // correct if square deviation is less
      correct++;                 //  than tolerance squared
    } else {
      printf(" tmp %f h_a %f h_b %f  h_c %f \n", tmp, h_a[i], h_b[i], h_c[i]);
    }
  }

  printf("vector add to find C = A+B:  %d out of %d results were correct.\n",
         correct, LENGTH);
}

vadd.cl

__kernel void vadd(                             
   __global float* a,                      
   __global float* b,                      
   __global float* c,                      
   const unsigned int count)               
{                                          
   int i = get_global_id(0);               
   if(i < count)  {
       c[i] = a[i] + b[i];                 
   }
}