opencl-wrapper

OpenCL-Wrapper allows embedding OpenCL C code directly into C++ using a stringification macro R(...). This preserves syntax highlighting in most editors.

Important Syntax Rules:

• Stringification: The macro R(...) converts arguments to string literals. It converts ' ' to ' '.
• Length Limits: Because string literals cannot be arbitrarily long, you must interrupt them periodically using the )+R( pattern.
• Unbalanced Brackets: To use unbalanced brackets like ( or ), exit the macro and insert the literal manually: ... ) + "void func(" + R( ....
• Preprocessor Macros: Standard preprocessor replacement macros (e.g., #define VAR 42) do not work inside the R(...) macro. Instead, pass these values directly to the Device constructor.
• Switch Macros: To use #define or #ifdef inside the kernel code, you must manually exit and re-enter the R(...) macro.

OpenCL-Wrapper simplifies OpenCL development by abstracting away the boilerplate of Platforms, Contexts, Commands Queues, and Programs. It focuses on three primary abstractions:

1. Device

Represents the OpenCL hardware. Creating a Device object automatically handles OpenCL C code compilation and includes device-specific workarounds (e.g., fixing VRAM reporting on Intel Arc or buffer limits on AMD).

• Selection: Can be selected in one line (fastest device, most memory, or specific ID).
• Capabilities: Automatically enables FP64, FP16, and INT64 atomics where supported.

2. Memory

A unified object for both host and device memory.

• Simplification: Eliminates the need to maintain separate host and device buffers or manually track buffer lengths/types.
• Features: Supports 1D, 2D, and 3D grid domains, multi-dimensional vectors, and automatic zero-copy on CPUs/iGPUs.
• Tracking: Automatically tracks total global memory usage of the device.

3. Kernel

Represents the OpenCL C function to be executed.

• Parameter Linking: Memory objects and constants are linked to kernel parameters during creation.
• Execution: Kernels can be executed via kernel.run() and parameters can be updated via set_parameters(...) or chained: kernel.set_parameters(args).run().
• Safety: Provides error messages if C++ parameter types mismatch the OpenCL C code.

Before using OpenCL-Wrapper, you must install the appropriate GPU drivers and OpenCL runtimes for your hardware and operating system.

Windows

• GPUs (AMD/Intel/Nvidia): Download and install the official drivers for your specific GPU. These contain the OpenCL Runtime. Reboot after installation.
• CPUs: Download and install the Intel CPU Runtime for OpenCL. Reboot after installation.

Linux

• AMD GPUs: Install drivers and the rocm usecase via amdgpu-install. Ensure you add your user to the render and video groups.
• Intel GPUs: Install intel-opencl-icd and add your user to the render group.
• Nvidia GPUs: Install the appropriate nvidia-driver package.
• CPUs: Use either the oneAPI DPC++ Compiler (requires manual installation of oneTBB and configuration of .icd and .conf files) or PoCL (sudo apt install pocl-opencl-icd).

Android

• Install the Termux .apk.
• Inside Termux, run: apt update && apt upgrade -y && apt install -y clang git make.

# Example for Intel GPU on Linux
sudo apt update && sudo apt upgrade -y
sudo apt install -y g++ git make ocl-icd-libopencl1 ocl-icd-opencl-dev intel-opencl-icd
sudo usermod -a -G render $(whoami)
sudo shutdown -r now

To use OpenCL-Wrapper, you need to define your OpenCL C kernel code in a separate file (or string) and then manage memory and execution in your main application.

1. Define the Kernel: Write your OpenCL C code. Note that when using opencl_c_container(), you must ensure unbalanced round brackets () are avoided and string literals are not arbitrarily long to prevent parsing issues.
2. Initialize Device: Use Device(select_device_with_most_flops()) to automatically select the fastest available hardware.
3. Allocate Memory: Use the Memory<T> class to allocate memory that is accessible on both the host and the device.
4. Create Kernel Object: Instantiate a Kernel object by providing the device, the global work size, the kernel name string, and the required Memory buffers.
5. Data Transfer and Execution:
   • Initialize host-side data using standard array indexing on the Memory object.
   • Call .write_to_device() to upload data to the GPU/accelerator.
   • Call .run() to execute the kernel.
   • Call .read_from_device() to download results back to the host.
6. Synchronize: Call wait() to ensure all asynchronous operations are complete before exiting.

#include "opencl.hpp"

int main() {
	Device device(select_device_with_most_flops());

	const uint N = 1024u;
	Memory<float> A(device, N);
	Memory<float> B(device, N);
	Memory<float> C(device, N);

	Kernel add_kernel(device, N, "add_kernel", A, B, C);

	for(uint n=0u; n<N; n++) {
		A[n] = 3.0f;
		B[n] = 2.0f;
		C[n] = 1.0f;
	}

	A.write_to_device();
	B.write_to_device();
	add_kernel.run();
	C.read_from_device();

	wait();
	return 0;
}

When writing kernels for OpenCL-Wrapper, you typically wrap your OpenCL C code in a function that returns a string. This allows the wrapper to handle the compilation process.

Important Syntax Constraints:

• Avoid unbalanced round brackets () within the string literal.
• Avoid arbitrarily long string literals.
• Use )+R( to periodically interrupt the string if necessary to maintain syntax highlighting or parsing stability.

#include "kernel.hpp"

string opencl_c_container() {
    return R( // ########################## begin of OpenCL C code ####################################################################

kernel void add_kernel(global float* A, global float* B, global float* C) {
	const uint n = get_global_id(0);
	C[n] = A[n]+B[n];
}

);} // ############################################################### end of OpenCL C code #####################################################################