clReleaseProgram_segfault.cpp

// =================================================================================================
// Compile and run as: g++ --std=c++11 -g clReleaseProgram_segfault.cpp -o test -lOpenCL && ./test
// =================================================================================================

// Enabling any of these below will make it work again
//#define MAKE_IT_WORK_NO_BUILDING  // disable building the program
//#define MAKE_IT_WORK_NO_CACHE  // disable adding to cache
//#define MAKE_IT_WORK_NO_SHARED_PTR  // change shared_ptr with regular pointer

// =================================================================================================

#include <string>
#include <vector>
#include <memory>
#include <numeric>
#include <cstdio>

#if defined(__APPLE__) || defined(__MACOSX)
  #include <OpenCL/opencl.h>
#else
  #include <CL/opencl.h>
#endif

// =================================================================================================

class Program {
public:

  explicit Program(const cl_context context, const std::string &source) {
    const char *source_ptr = &source[0];
    const auto length = source.length();
    auto status = CL_SUCCESS;
    program_ = clCreateProgramWithSource(context, 1, &source_ptr, &length, &status);
    printf("[Program] Creating: clCreateProgramWithSource %p\n", program_);
  }

  ~Program() {
    printf("[Program] Cleaning-up: clReleaseProgram %p\n", program_);
    clReleaseProgram(program_);
    printf("[Program] All cleaned-up\n", program_);
  }

  void Build(const cl_device_id &device) {
    auto options = std::string{""};
    clBuildProgram(program_, 1, &device, options.c_str(), nullptr, nullptr);
  }

private:
  cl_program program_;
};

// =================================================================================================

#ifdef MAKE_IT_WORK_NO_SHARED_PTR
std::vector<Program*> cache;
#else
std::vector<std::shared_ptr<Program>> cache;
#endif

// =================================================================================================

int main() {
  const auto platform_id = size_t{0};
  const auto device_id = size_t{0};

  printf("[main] Getting platform\n");
  auto num_platforms = cl_uint{0};
  clGetPlatformIDs(0, nullptr, &num_platforms);
  auto platforms = std::vector<cl_platform_id>(num_platforms);
  clGetPlatformIDs(num_platforms, platforms.data(), nullptr);
  auto platform = platforms[platform_id];

  printf("[main] Getting device\n");
  auto num_devices_result = cl_uint{0};
  clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 0, nullptr, &num_devices_result);
  auto num_devices = static_cast<size_t>(num_devices_result);
  auto devices = std::vector<cl_device_id>(num_devices);
  clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, static_cast<cl_uint>(num_devices), devices.data(), nullptr);
  auto device = devices[device_id];

  printf("[main] Getting context\n");
  auto status = CL_SUCCESS;
  auto context = clCreateContext(nullptr, 1, &device, nullptr, nullptr, &status);

  printf("[main] Creating program\n");
  auto program_string = "__kernel void example_kernel() { }";
#ifdef MAKE_IT_WORK_NO_SHARED_PTR
  auto program = new Program(context, program_string);
#else
  auto program = std::make_shared<Program>(context, program_string);
#endif

#ifndef MAKE_IT_WORK_NO_BUILDING
  printf("[main] Building program\n");
  program->Build(device);
#endif

#ifndef MAKE_IT_WORK_NO_CACHE
  printf("[main] Adding to cache\n");
  cache.push_back(program);
#endif

  printf("[main] Cleaning-up: clReleaseContext\n");
  clReleaseContext(context);

  printf("[main] All done\n");
#ifdef MAKE_IT_WORK_NO_SHARED_PTR
  delete program;
#endif
  return 0;
}

// =================================================================================================