V Computing Language (VCL) 🖥️

VCL is a high-level, V-native interface for GPU computing with OpenCL. It provides an opinionated, simplified approach to GPU programming that emphasizes ease of use while maintaining high performance for scientific computing applications.

🚀 Features

• Device Abstraction: Simplified device management hiding OpenCL complexity
• Memory Management: Automatic buffer creation and data transfer
• Kernel Compilation: Dynamic OpenCL C compilation with detailed error reporting
• Cross-Platform: Supports NVIDIA, AMD, Intel GPUs and multi-core CPUs
• V Integration: Native V syntax with error handling and memory safety

🎯 Quick Start Example

import vsl.vcl

// Initialize VCL and get the first available device
mut device := vcl.get_default_device()!

// OpenCL kernel source code
kernel_source := '
__kernel void vector_add(__global float* a, __global float* b, __global float* c, int n) {
    int id = get_global_id(0);
    if (id < n) {
        c[id] = a[id] + b[id];
    }
}'

// Compile the kernel
device.add_program(kernel_source)!

// Create and use buffers for computation
// (See examples directory for complete implementations)

📊 Visual Gallery

VCL enables creation of stunning GPU-accelerated visualizations:

🔧 Installation & Configuration

Prerequisites

1. OpenCL Runtime: Install GPU vendor's OpenCL runtime

• NVIDIA: CUDA Toolkit or GPU drivers
• AMD: AMD APP SDK or Radeon drivers
• Intel: Intel OpenCL Runtime or integrated graphics drivers

2. OpenCL Headers: Development headers for compilation

Platform-Specific Installation

Ubuntu/Debian:

##sudo apt-get install nvidia-opencl-dev

##sudo apt-get install amd-opencl-dev

##sudo apt-get install intel-opencl-icd

macOS: OpenCL is built into the system (no additional installation required)

Windows: Install your GPU vendor's SDK (NVIDIA CUDA, AMD APP, Intel OpenCL)

Verification

Test your OpenCL installation:

##clinfo  ##
##cd examples/vcl_opencl_basic
v run main.v

By default VCL uses the OpenCL headers from the system path and all the known locations for OpenCL headers (like /usr/include and /usr/local/include) and load the first header it finds. If you want to use a specific OpenCL header, you can add the -I flag into your V program with the path to the headers directory.

#

or at compile time:

v -I/custom/path/to/opencl/headers my_program.v

You can also link or move the headers directory into VCL's source directory. For example:

##ln -s /custom/path/to/opencl/headers ~/.vmodules/vcl/OpenCL

##ln -s /custom/path/to/opencl/headers ~/.vmodules/vcl/CL

or, you can copy the headers directory into VCL's source directory. For example you can clone the OpenCL-Headers repository and copy the headers as follows:

git clone https://github.com/KhronosGroup/OpenCL-Headers /tmp/OpenCL-Headers

##cp -r /tmp/OpenCL-Headers/CL ~/.vmodules/vcl/OpenCL

##cp -r /tmp/OpenCL-Headers/CL ~/.vmodules/vcl/CL

Loading OpenCL dynamically

By default VCL uses OpenCL loading the library statically. If you want to use OpenCL dynamically, you can use the -d vsl_vcl_dlopencl flag.

By default it will look for the OpenCL library in the system path and all the known locations for OpenCL libraries (like /usr/lib and /usr/local/lib) and load the first library it finds. If you want to use a specific OpenCL library, you can declare the environment variable VCL_LIBOPENCL_PATH with the path to the library. Multiple paths can be separated by :.

For example, if you want to use the OpenCL library from the NVIDIA CUDA Toolkit, you can do the following:

export VCL_LIBOPENCL_PATH=/usr/local/cuda/lib64/libOpenCL.so

🎓 Learning Path

Beginner: Start Here

1. Basic Device Setup: vcl_opencl_basic - Device detection and simple kernels
2. Memory Management: Learn buffer creation and data transfer patterns
3. Simple Algorithms: Vector addition, element-wise operations

Intermediate: Explore More

1. Image Processing: vcl_opencl_image_example - 2D data manipulation
2. Mathematical Algorithms: Matrix operations, FFT implementations
3. Performance Optimization: Memory access patterns, work-group sizing

Advanced: Push Limits

1. Fractal Generation: vcl_opencl_fractals_one_argument - Complex mathematical visualization
2. Multi-Device Programming: Utilizing multiple GPUs simultaneously
3. Custom Memory Patterns: Local memory optimization, async operations

🔧 Custom OpenCL Headers (Advanced)

IMPORTANT: Header version compatibility is crucial. Mismatched OpenCL > headers can cause runtime failures.

By default, VCL uses system OpenCL headers. For custom installations:

Method 1: Compiler Flags

#

Or at compile time:

v -I/custom/path/to/opencl/headers my_program.v

Method 2: Symbolic Links

##ln -s /custom/path/to/opencl/headers ~/.vmodules/vcl/OpenCL

# L#ln -s /custom/path/to/opencl/headers ~/.vmodules/vcl/CL

Method 3: Direct Copy

##git clone https://github.com/KhronosGroup/OpenCL-Headers /tmp/OpenCL-Headers

####cp -r /tmp/OpenCL-Headers/CL ~/.vmodules/vcl/OpenCL

##cp -r /tmp/OpenCL-Headers/CL ~/.vmodules/vcl/CL

⚡ Dynamic OpenCL Loading

For runtime flexibility, use dynamic loading:

##v -d vsl_vcl_dlopencl my_program.v

Custom Library Paths

Set environment variable for specific OpenCL libraries:

##export VCL_LIBOPENCL_PATH=/usr/local/cuda/lib64/libOpenCL.so

##export VCL_LIBOPENCL_PATH=/usr/lib/libOpenCL.so:/opt/intel/opencl/lib64/libOpenCL.so

🐛 Troubleshooting

Common Issues

No OpenCL platforms found- Install GPU drivers and OpenCL runtime

• Check clinfo output for available platforms
• Verify OpenCL library is in system path

Kernel compilation errors- Check OpenCL C syntax (different from standard C)

• Verify data type compatibility (float vs double support)
• Use VCL's detailed error reporting for debugging

Performance issues- Profile memory transfer vs computation time

• Optimize work-group sizes for your hardware
• Consider local memory usage patterns

Build failures- Ensure OpenCL development headers are installed

• Check compiler flag compatibility
• Verify V compiler version compatibility

Platform-Specific Notes

NVIDIA GPUs:- Best performance with latest CUDA toolkit

• Double precision requires compute capability 1.3+
• Use nvidia-smi to check GPU utilization

AMD GPUs:- Strong OpenCL support across all generations

• Consider ROCm for professional workloads
• Use rocm-smi for monitoring

Intel Graphics:- Good for development and light workloads

• Limited by memory bandwidth
• Excellent for heterogeneous CPU+GPU computing

📚 Examples

Explore VCL capabilities through examples:

• vcl_opencl_basic - Device setup and simple kernels
• vcl_opencl_image_example - Image processing operations
• vcl_opencl_fractals_one_argument - Mathematical visualization
• vcl_opencl_kernel_params - Parameter passing techniques

🔗 Resources

• OpenCL Programming Guide
• VSL Documentation
• V Language Documentation
• GPU Architecture Guides

Accelerate your scientific computing with VCL! 🚀

fn error_from_code(code int) IError

fn error_or_default[T](code int, default T) !T

fn get_default_device() !&Device

get_default_device ...

fn get_devices(device_type DeviceType) ![]&Device

get_devices returns all devices of all platforms with specified type

fn panic_on_error(code int)

fn typed_error[T](code int) !T

fn vcl_error(code int) !

interface ArgumentType {}

interface IImage {
	width       int
	height      int
	nr_channels int
	data        &u8
}

IImage holds the fileds and data needed to represent a bitmap/pixel based image in memory.

type ErrVCL = int

ErrVCL converts that OpenCL error code to an V error

fn (e ErrVCL) err() IError

fn (v &Vector[T]) length() int

Length the length of the vector

fn (v &Vector[T]) release() !

Release releases the buffer on the device

fn (mut v Vector[T]) load(data []T) chan IError

load copies the T data from host data to device buffer it's a non-blocking call, channel will return an error or nil if the data transfer is complete

fn (v &Vector[T]) data() ![]T

data gets T data from device, it's a blocking call

fn (v &Vector[T]) map(k &Kernel) chan IError

map applies an map kernel on all elements of the vector

fn (v &Vector[T]) buffer() &Buffer

buffer returns the underlying buffer

enum DeviceType as i64 {
	// device types - bitfield
	default     = (1 << 0)
	cpu         = (1 << 1)
	gpu         = (1 << 2)
	accelerator = (1 << 3)
	custom      = (1 << 4)
	all         = 0xFFFFFFFF
}

enum ImageChannelDataType {
	unorm_int8 = C.CL_UNORM_INT8
}

ImageChannelDataType describes the size of the channel data type

enum ImageChannelOrder {
	intensity = C.CL_INTENSITY
	rgba      = C.CL_RGBA
}

ImageChannelOrder represents available image types

struct Bytes {
	buf &Buffer = unsafe { nil }
}

Bytes is a memory buffer on the device that holds []byte

fn (b &Bytes) size() int

size the size of the bytes buffer

fn (b &Bytes) release() !

release releases the buffer on the device

fn (b &Bytes) load(data []byte) chan IError

load copies the data from host data to device buffer it's a non-blocking call, channel will return an error or nil if the data transfer is complete

fn (b &Bytes) data() ![]u8

data gets data from device, it's a blocking call

fn (b &Bytes) map(mut k Kernel) chan IError

map applies an map kernel on all elements of the buffer

fn (b &Bytes) buffer() &Buffer

buffer returns the underlying buffer

struct Device {
mut:
	id       ClDeviceId
	ctx      ClContext
	queue    ClCommandQueue
	programs []ClProgram
}

Device the only needed entrence for the VCL represents the device on which memory can be allocated and kernels run it abstracts away all the complexity of contexts/platforms/queues

fn (mut d Device) add_program(source string) !

add_program compiles program source from OpenCL C code This method takes OpenCL C source code, compiles it, and stores the resulting program for later kernel creation and execution.

Parameters: source: OpenCL C source code as a string

Returns: Error if compilation fails, including detailed build log for debugging

fn (d &Device) bytes(size int) !&Bytes

bytes allocates new memory buffer with specified size on device

fn (d &Device) driver_version() !string

driver_version device info - driver version

fn (d &Device) extensions() !string

extensions device info - extensions

fn (d &Device) from_image(img IImage) !&Image

from_image creates new Image and copies data from Image

fn (d &Device) image(@type ImageChannelOrder, bounds Rect) !&Image

image allocates an image buffer

fn (d &Device) kernel(name string) !&Kernel

kernel returns a kernel if retrieving the kernel didn't complete the function will return an error

fn (d &Device) name() !string

name device info - name

fn (d &Device) open_clc_version() !string

open_clc_version device info - OpenCL C version

fn (d &Device) profile() !string

profile device info - profile

fn (mut d Device) release() !

release releases the device

fn (d &Device) str() string

fn (d &Device) vector[T](length int) !&Vector[T]

vector allocates new vector buffer with specified length

fn (d &Device) vendor() !string

vendor device info - vendor

fn (d &Device) version() !string

version device info - version

struct Image {
	format   ClImageFormat
	desc     &ClImageDesc = unsafe { nil }
	img_data voidptr
mut:
	buf &Buffer
pub:
	@type  ImageChannelOrder
	bounds Rect
}

Image memory buffer on the device with image data

fn (mut img Image) release() !

release releases the buffer on the device

fn (image &Image) data() !IImage

struct Kernel {
	d &Device = unsafe { nil }
	k ClKernel
}

Kernel represent a single kernel

fn (k &Kernel) global(global_work_sizes ...int) KernelWithGlobal

global returns an kernel with global size set

struct KernelCall {
	kernel            &Kernel = unsafe { nil }
	global_work_sizes []int
	local_work_sizes  []int
}

KernelCall is a kernel with global and local work sizes set and it's ready to be run

fn (kc KernelCall) run(args ...ArgumentType) chan IError

run calls the kernel on its device with specified global and local work sizes and arguments it's a non-blocking call, so it returns a channel that will send an error value when the kernel is done or nil if the call was successful

struct KernelWithGlobal {
	kernel            &Kernel = unsafe { nil }
	global_work_sizes []int
}

KernelWithGlobal is a kernel with the global size set to run the kernel it must also set the local size

fn (kg KernelWithGlobal) local(local_work_sizes ...int) KernelCall

local ets the local work sizes and returns an KernelCall which takes kernel arguments and runs the kernel

struct Rect {
pub:
	// pixel need integers
	x      f32
	y      f32
	width  f32
	height f32
}

Rect is a struct that represents a rectangle shape

struct UnsupportedArgumentTypeError {
	Error
pub:
	index int
	value ArgumentType
}

fn (err UnsupportedArgumentTypeError) msg() string

struct Vector[T] {
mut:
	buf &Buffer = unsafe { nil }
}

Vector is a memory buffer on device that holds []T