render with raylib

10 files changed, 426 insertions, 411 deletions

diff --git a/.clang-tidy b/.clang-tidy
new file mode 100644
index 0000000..20b4476
--- /dev/null
+++ b/.clang-tidy
@@ -0,0 +1,5 @@
+# clangd doesn't like sizeof expressions for pointers to pointers, eg:
+#     int **x = malloc(sizeof(*x));
+# disables annoying warnings (be careful!)
+
+Checks: '-bugprone-sizeof-expression'
diff --git a/Makefile b/Makefile
index 0a49d95..feb5551 100644
--- a/Makefile
+++ b/Makefile
@@ -19,9 +19,9 @@
 
 include config.mk
 
-SRC   = src/spirt.c src/util.c src/farbfeld.c
+SRC   = src/spirt.c src/util.c src/util_cl.c
 CLSRC = src/cl/spirt.cl
-HDR   = src/util.h src/farbfeld.h
+HDR   = src/util.h src/util_cl.h
 
 LL  = ${CLSRC:.cl=.ll}
 SPV = ${CLSRC:.cl=.spv}
@@ -30,16 +30,16 @@ OBJ = ${SRC:.c=.o} ${CLSRC:.cl=.o}
 
 all: spirt
 
-src/spirt.o: config.mk src/util.h src/farbfeld.h
 src/util.o: config.mk
-src/farbfeld.o: config.mk src/util.h
+src/util_cl.o: config.mk src/util.h
+src/spirt.o: config.mk src/util.h src/util_cl.h
 src/cl/spirt.o: config.mk
 
 spirt: ${OBJ}
-	${CC} -o $@ $^ ${LDFLAGS}
+	${CC} $(LDFLAGS) $^ $(LDLIBS) -o $@
 
 %.o: %.c
-	${CC} -c ${CFLAGS} $< -o $@
+	${CC} $(CFLAGS) -c -o $@ $<
 
 %.o: %.spv
 	objcopy ${OBJFLAGS} $< $@
diff --git a/config.mk b/config.mk
index 69656f9..b278d8d 100644
--- a/config.mk
+++ b/config.mk
@@ -4,12 +4,12 @@ PREFIX = /usr/local
 
 CC = gcc
 
-LIBS = -lOpenCL
+LDLIBS = -lOpenCL -lraylib -lm
 
 CFLAGS = -std=c89 -pedantic -Wall -Wextra -O2 -pipe \
          -DCL_TARGET_OPENCL_VERSION=210 -D_POSIX_C_SOURCE=200809L
 
-LDFLAGS = -Wl,-O1,--sort-common,--as-needed ${LIBS}
+LDFLAGS = -Wl,-O1,--sort-common,--as-needed
 
 CLFLAGS = -target spir64 -emit-llvm -Xclang -finclude-default-header -O3 \
           -cl-std=CL1.2 -Wall -Wextra
diff --git a/src/cl/spirt.cl b/src/cl/spirt.cl
index 300f552..12022bb 100644
--- a/src/cl/spirt.cl
+++ b/src/cl/spirt.cl
@@ -14,7 +14,7 @@ You should have received a copy of the GNU General Public License
 along with this program.  If not, see <http://www.gnu.org/licenses/>. */
 
 __kernel void
-hello(__global ushort *const canvas)
+hello(__global unsigned char *const canvas)
 {
 	size_t idx, w, h, i, j;
 
@@ -26,8 +26,8 @@ hello(__global ushort *const canvas)
 
 	idx = (w * j + i) * 4;
 
-	canvas[idx] = (ushort)((65535 * i) / (w - 1));
-	canvas[idx + 1] = (ushort)((65535 * j) / (h - 1));
-	canvas[idx + 2] = 16383;
-	canvas[idx + 3] = 65535;
+	canvas[idx] = (unsigned char)((255 * i) / (w - 1));
+	canvas[idx + 1] = (unsigned char)((255 * j) / (h - 1));
+	canvas[idx + 2] = 63;
+	canvas[idx + 3] = 255;
 }
diff --git a/src/farbfeld.c b/src/farbfeld.c
deleted file mode 100644
index 71903b1..0000000
--- a/src/farbfeld.c
+++ /dev/null
@@ -1,40 +0,0 @@
-/* Copyright (C) 2024 Aiden Gall
-
-This program is free software: you can redistribute it and/or modify
-it under the terms of the GNU General Public License as published by
-the Free Software Foundation, either version 3 of the License, or
-(at your option) any later version.
-
-This program is distributed in the hope that it will be useful,
-but WITHOUT ANY WARRANTY; without even the implied warranty of
-MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
-GNU General Public License for more details.
-
-You should have received a copy of the GNU General Public License
-along with this program.  If not, see <http://www.gnu.org/licenses/>. */
-
-#include "util.h"
-
-#include <arpa/inet.h>
-#include <stdint.h>
-#include <stdio.h>
-
-/* write farbfeld image to fd
- * - canvas will be mutated unless host architecture is big endian */
-void
-farbfeld_write(FILE *const fd, uint16_t *const canvas, const size_t canvas_size,
-               uint32_t width, uint32_t height)
-{
-	size_t i;
-
-	width = htonl(width);
-	height = htonl(height);
-
-	for (i = 0; i < canvas_size; i++)
-		canvas[i] = htons(canvas[i]);
-
-	efwrite("farbfeld", sizeof(char), 8, fd);
-	efwrite(&width, sizeof(uint32_t), 1, fd);
-	efwrite(&height, sizeof(uint32_t), 1, fd);
-	efwrite(canvas, sizeof(uint16_t), canvas_size, fd);
-}
diff --git a/src/spirt.c b/src/spirt.c
index d1e8399..fd6457b 100644
--- a/src/spirt.c
+++ b/src/spirt.c
@@ -13,27 +13,21 @@ GNU General Public License for more details.
 You should have received a copy of the GNU General Public License
 along with this program.  If not, see <http://www.gnu.org/licenses/>. */
 
-#include "farbfeld.h"
 #include "util.h"
+#include "util_cl.h"
 
 #include <CL/cl.h>
 #include <CL/cl_platform.h>
-#include <errno.h>
 #include <limits.h>
-#include <stdint.h>
+#include <raylib.h>
 #include <stdio.h>
 #include <stdlib.h>
-#include <string.h>
-
-#ifndef SPIRV_START
-#	define SPIRV_START _binary_src_cl_spirt_spv_start
-#endif
-#ifndef SPIRV_END
-#	define SPIRV_END _binary_src_cl_spirt_spv_end
-#endif
+
+#define SPIRV_START _binary_src_cl_spirt_spv_start
+#define SPIRV_END _binary_src_cl_spirt_spv_end
 #define SPIRV_SIZE ((size_t)SPIRV_END - (size_t)SPIRV_START)
 
-struct ClRuntime {
+struct kernel_context {
 	cl_platform_id platform;
 	cl_device_id device;
 	cl_context context;
@@ -42,119 +36,72 @@ struct ClRuntime {
 	cl_kernel k_hello;
 };
 
-static cl_platform_id *get_platforms(cl_uint *num_platforms);
-static cl_device_id *get_devices(cl_platform_id platform, cl_uint *num_devices);
-static cl_program compile_spirv_program(cl_context context, cl_device_id device,
-                                        const void *spirv_start,
-                                        size_t spirv_size);
+struct opts {
+	long width;
+	long height;
+	long platidx;
+	long devidx;
+};
+
+static unsigned char *kernel_hello(struct kernel_context *runtime,
+                                   size_t image_width, size_t image_height);
 
-static struct ClRuntime init(size_t platidx, size_t devidx);
-static uint16_t *run_k_hello(struct ClRuntime *runtime, size_t *canvas_size_ret,
-                             size_t image_width, size_t image_height);
-static void clean(struct ClRuntime *runtime);
+static struct kernel_context kernel_context_init(size_t platidx, size_t devidx);
+static void kernel_context_clean(struct kernel_context *runtime);
 
-static long argtonum(const char *arg, long minval, long maxval);
+static struct opts get_args(int argc, char *argv[]);
 
 extern const char *SPIRV_START[];
 extern const char *SPIRV_END[];
 
-static cl_platform_id *
-get_platforms(cl_uint *const num_platforms)
+static unsigned char *
+kernel_hello(struct kernel_context *const runtime, const size_t image_width,
+             const size_t image_height)
 {
 	cl_int err;
 
-	cl_platform_id *platforms;
-	cl_uint len;
-
-	err = clGetPlatformIDs(0, NULL, &len);
-	if (err != CL_SUCCESS)
-		die("clGetPlatformIDs: %s\n", cl_strerror(err));
-	warn("number of platforms = %d\n", len);
-	if (len < 1)
-		die("clGetPlatformIDs: No OpenCL platforms\n");
-
-	platforms = calloc(len, sizeof(*platforms));
-	if (!platforms)
-		die("calloc: Out of memory\n");
-	err = clGetPlatformIDs(len, platforms, NULL);
-	if (err != CL_SUCCESS)
-		die("clGetPlatformIDs: %s\n", cl_strerror(err));
+	size_t global_item_size[2];
 
-	*num_platforms = len;
-	return platforms;
-}
+	unsigned char *h_canvas;
+	cl_mem d_canvas;
+	size_t canvas_size;
 
-static cl_device_id *
-get_devices(const cl_platform_id platform, cl_uint *const num_devices)
-{
-	cl_int err;
+	canvas_size = (sizeof("RGBA") - 1) * image_width * image_height;
 
-	cl_device_id *devices;
-	cl_uint len;
+	h_canvas = calloc(canvas_size, sizeof(*h_canvas));
+	if (!h_canvas)
+		die("calloc: Out of memory\n");
 
-	err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 0, NULL, &len);
+	d_canvas = clCreateBuffer(runtime->context, CL_MEM_WRITE_ONLY,
+	                          sizeof(*h_canvas) * canvas_size, NULL, &err);
 	if (err != CL_SUCCESS)
-		die("clGetDeviceIDs: %s\n", cl_strerror(err));
-	warn("number of devices in platform = %d\n", len);
-	if (len < 1)
-		die("clGetDeviceIDs: No OpenCL devices in platform\n");
+		die("clCreateBuffer: %s\n", cl_strerror(err));
 
-	devices = calloc(len, sizeof(*devices));
-	if (!devices)
-		die("calloc: Out of memory\n");
-	err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, len, devices, NULL);
+	err = clSetKernelArg(runtime->k_hello, 0, sizeof(d_canvas), &d_canvas);
 	if (err != CL_SUCCESS)
-		die("clGetDeviceIDs: %s\n", cl_strerror(err));
-
-	*num_devices = len;
-	return devices;
-}
+		die("clSetKernelArg: %s\n", cl_strerror(err));
 
-static cl_program
-compile_spirv_program(const cl_context context, const cl_device_id device,
-                      const void *const spirv_start, const size_t spirv_size)
-{
-	cl_int err, build_err;
-	cl_program program;
+	global_item_size[0] = image_width;
+	global_item_size[1] = image_height;
+	err = clEnqueueNDRangeKernel(runtime->command_queue, runtime->k_hello,
+	                             2, NULL, global_item_size, NULL, 0, NULL,
+	                             NULL);
+	if (err != CL_SUCCESS)
+		die("clEnqueueNDRangeKernel: %s\n", cl_strerror(err));
 
-	program = clCreateProgramWithIL(context, spirv_start, spirv_size, &err);
+	err = clEnqueueReadBuffer(runtime->command_queue, d_canvas, CL_TRUE, 0,
+	                          sizeof(*h_canvas) * canvas_size, h_canvas, 0,
+	                          NULL, NULL);
 	if (err != CL_SUCCESS)
-		die("clCreateProgramWithIL: %s\n", cl_strerror(err));
-
-	build_err = clBuildProgram(program, 1, &device, NULL, NULL, NULL);
-	if (build_err != CL_SUCCESS) {
-		cl_char *log;
-		size_t buffer_size;
-
-		err = clGetProgramBuildInfo(program, device,
-		                            CL_PROGRAM_BUILD_LOG, 0, NULL,
-		                            &buffer_size);
-		if (err != CL_SUCCESS)
-			die("clGetProgramBuildInfo: %s\n", cl_strerror(err));
-		if (buffer_size < 1)
-			die("clGetProgramBuildInfo: "
-			    "Build log buffer is empty\n");
-
-		log = calloc(buffer_size, sizeof(*log));
-		if (!log)
-			die("calloc: Out of memory\n");
-		err = clGetProgramBuildInfo(
-		        program, device, CL_PROGRAM_BUILD_LOG,
-		        sizeof(*log) * buffer_size, log, NULL);
-		if (err != CL_SUCCESS)
-			die("clGetProgramBuildInfo: %s\n", cl_strerror(err));
-
-		efwrite(log, sizeof(*log), buffer_size, stderr);
-
-		free(log);
-		die("\nclBuildProgram: %s\n", cl_strerror(build_err));
-	}
+		die("clEnqueueReadBuffer: %s\n", cl_strerror(err));
+
+	clReleaseMemObject(d_canvas);
 
-	return program;
+	return h_canvas;
 }
 
-static struct ClRuntime
-init(const size_t platidx, const size_t devidx)
+static struct kernel_context
+kernel_context_init(const size_t platidx, const size_t devidx)
 {
 	cl_int err;
 
@@ -163,7 +110,7 @@ init(const size_t platidx, const size_t devidx)
 	cl_uint num_platforms;
 	cl_uint num_devices;
 
-	struct ClRuntime runtime;
+	struct kernel_context runtime;
 
 	platforms = get_platforms(&num_platforms);
 	if (platidx >= num_platforms)
@@ -198,55 +145,8 @@ init(const size_t platidx, const size_t devidx)
 	return runtime;
 }
 
-static uint16_t *
-run_k_hello(struct ClRuntime *const runtime, size_t *const canvas_size_ret,
-            const size_t image_width, const size_t image_height)
-{
-	cl_int err;
-
-	size_t global_item_size[2];
-
-	uint16_t *h_canvas;
-	cl_mem d_canvas;
-	size_t canvas_size;
-
-	canvas_size = (sizeof("RGBA") - 1) * image_width * image_height;
-
-	h_canvas = calloc(canvas_size, sizeof(*h_canvas));
-	if (!h_canvas)
-		die("calloc: Out of memory\n");
-
-	d_canvas = clCreateBuffer(runtime->context, CL_MEM_WRITE_ONLY,
-	                          sizeof(*h_canvas) * canvas_size, NULL, &err);
-	if (err != CL_SUCCESS)
-		die("clCreateBuffer: %s\n", cl_strerror(err));
-
-	err = clSetKernelArg(runtime->k_hello, 0, sizeof(d_canvas), &d_canvas);
-	if (err != CL_SUCCESS)
-		die("clSetKernelArg: %s\n", cl_strerror(err));
-
-	global_item_size[0] = image_width;
-	global_item_size[1] = image_height;
-	err = clEnqueueNDRangeKernel(runtime->command_queue, runtime->k_hello,
-	                             2, NULL, global_item_size, NULL, 0, NULL,
-	                             NULL);
-	if (err != CL_SUCCESS)
-		die("clEnqueueNDRangeKernel: %s\n", cl_strerror(err));
-
-	err = clEnqueueReadBuffer(runtime->command_queue, d_canvas, CL_TRUE, 0,
-	                          sizeof(*h_canvas) * canvas_size, h_canvas, 0,
-	                          NULL, NULL);
-	if (err != CL_SUCCESS)
-		die("clEnqueueReadBuffer: %s\n", cl_strerror(err));
-
-	clReleaseMemObject(d_canvas);
-
-	*canvas_size_ret = canvas_size;
-	return h_canvas;
-}
-
 static void
-clean(struct ClRuntime *const runtime)
+kernel_context_clean(struct kernel_context *const runtime)
 {
 	clFlush(runtime->command_queue);
 	clFinish(runtime->command_queue);
@@ -257,45 +157,15 @@ clean(struct ClRuntime *const runtime)
 	clReleaseContext(runtime->context);
 }
 
-static long
-argtonum(const char *const arg, const long minval, const long maxval)
+static struct opts
+get_args(int argc, char *argv[])
 {
-	long ret;
-	char *endptr;
+	struct opts opts;
 
-	if (minval > maxval)
-		die("argtonum: Invalid range\n");
-	if (!arg)
-		die("argtonum: Missing argument\n");
-
-	errno = 0;
-	ret = strtol(arg, &endptr, 10);
-	if (errno)
-		die("strtol: %s\n", strerror(errno));
-	if (arg == endptr || *endptr)
-		die("strtol: Invalid integer\n");
-
-	if (ret < minval || ret > maxval)
-		die("argtonum: Argument out of range\n");
-
-	return ret;
-}
-
-int
-main(int argc, char *argv[])
-{
-	long opt_platidx, opt_devidx, opt_width, opt_height;
-	FILE *fd;
-
-	struct ClRuntime runtime;
-
-	uint16_t *h_canvas;
-	size_t canvas_size;
-
-	opt_platidx = 0;
-	opt_devidx = 0;
-	opt_width = 256;
-	opt_height = 256;
+	opts.platidx = 0;
+	opts.devidx = 0;
+	opts.width = 256;
+	opts.height = 256;
 
 	if (argv[0])
 		argv++, argc--;
@@ -311,16 +181,16 @@ main(int argc, char *argv[])
 
 		switch (flag) {
 		case 'p':
-			opt_platidx = argtonum(argv[0], 0, LONG_MAX);
+			opts.platidx = argtonum(argv[0], 0, LONG_MAX);
 			break;
 		case 'd':
-			opt_devidx = argtonum(argv[0], 0, LONG_MAX);
+			opts.devidx = argtonum(argv[0], 0, LONG_MAX);
 			break;
 		case 'w':
-			opt_width = argtonum(argv[0], 1, LONG_MAX);
+			opts.width = argtonum(argv[0], 1, LONG_MAX);
 			break;
 		case 'h':
-			opt_height = argtonum(argv[0], 1, LONG_MAX);
+			opts.height = argtonum(argv[0], 1, LONG_MAX);
 			break;
 		case '-':
 			goto end_flags;
@@ -330,28 +200,55 @@ main(int argc, char *argv[])
 	}
 
 end_flags:
-	if (argc > 1)
+	if (argc > 0)
 		die("arg_parse: Unexpected argument\n");
 
-	fd = stdout;
-	if (argv[0]) {
-		errno = 0;
-		fd = fopen(argv[0], "w");
-		if (errno)
-			die("fopen: %s\n", strerror(errno));
-		if (!fd)
-			die("fopen: Returned NULL\n");
-	}
+	return opts;
+}
 
-	runtime = init(opt_platidx, opt_devidx);
+int
+main(int argc, char *argv[])
+{
+	struct opts opts;
+
+	struct kernel_context runtime;
+	unsigned char *h_canvas;
+
+	Image img;
+	Texture2D texture;
+
+	opts = get_args(argc, argv);
+
+	runtime = kernel_context_init(opts.platidx, opts.devidx);
 
-	h_canvas = run_k_hello(&runtime, &canvas_size, opt_width, opt_height);
+	h_canvas = kernel_hello(&runtime, opts.width, opts.height);
+
+	img.data = h_canvas;
+	img.width = opts.width;
+	img.height = opts.height;
+	img.format = PIXELFORMAT_UNCOMPRESSED_R8G8B8A8;
+	img.mipmaps = 1;
+
+	clFlush(runtime.command_queue);
+
+	InitWindow(opts.width, opts.height, "spirt");
+
+	SetTargetFPS(60);
+
+	texture = LoadTextureFromImage(img);
+
+	while (!WindowShouldClose()) {
+		BeginDrawing();
+		ClearBackground(BLACK);
+		DrawTexture(texture, 0, 0, WHITE);
+		EndDrawing();
+	}
 
-	farbfeld_write(fd, h_canvas, canvas_size, opt_width, opt_height);
-	fclose(fd);
+	UnloadTexture(texture);
+	CloseWindow();
 
 	free(h_canvas);
-	clean(&runtime);
+	kernel_context_clean(&runtime);
 
 	return 0;
 }
diff --git a/src/util.c b/src/util.c
index bf06bca..dcf3212 100644
--- a/src/util.c
+++ b/src/util.c
@@ -13,19 +13,18 @@ GNU General Public License for more details.
 You should have received a copy of the GNU General Public License
 along with this program.  If not, see <http://www.gnu.org/licenses/>. */
 
-#include <CL/cl.h>
-#include <CL/cl_platform.h>
-#include <CL/cl_version.h>
+#include <errno.h>
 #include <stdarg.h>
 #include <stdio.h>
 #include <stdlib.h>
+#include <string.h>
 
 void warn(const char *fmt, ...);
 void die(const char *fmt, ...);
 
 void efwrite(const void *ptr, size_t size, size_t nmemb, FILE *stream);
 
-const char *cl_strerror(cl_int err);
+long argtonum(const char *arg, long minval, long maxval);
 
 void
 warn(const char *const fmt, ...)
@@ -57,145 +56,26 @@ efwrite(const void *const ptr, const size_t size, const size_t nmemb,
 		die("fwrite: Write failed\n");
 }
 
-const char *
-cl_strerror(const cl_int err)
+long
+argtonum(const char *const arg, const long minval, const long maxval)
 {
-	switch (err) {
-	case CL_SUCCESS:
-		return "CL_SUCCESS";
-	case CL_DEVICE_NOT_FOUND:
-		return "CL_DEVICE_NOT_FOUND";
-	case CL_DEVICE_NOT_AVAILABLE:
-		return "CL_DEVICE_NOT_AVAILABLE";
-	case CL_COMPILER_NOT_AVAILABLE:
-		return "CL_COMPILER_NOT_AVAILABLE";
-	case CL_MEM_OBJECT_ALLOCATION_FAILURE:
-		return "CL_MEM_OBJECT_ALLOCATION_FAILURE";
-	case CL_OUT_OF_RESOURCES:
-		return "CL_OUT_OF_RESOURCES";
-	case CL_OUT_OF_HOST_MEMORY:
-		return "CL_OUT_OF_HOST_MEMORY";
-	case CL_PROFILING_INFO_NOT_AVAILABLE:
-		return "CL_PROFILING_INFO_NOT_AVAILABLE";
-	case CL_MEM_COPY_OVERLAP:
-		return "CL_MEM_COPY_OVERLAP";
-	case CL_IMAGE_FORMAT_MISMATCH:
-		return "CL_IMAGE_FORMAT_MISMATCH";
-	case CL_IMAGE_FORMAT_NOT_SUPPORTED:
-		return "CL_IMAGE_FORMAT_NOT_SUPPORTED";
-	case CL_BUILD_PROGRAM_FAILURE:
-		return "CL_BUILD_PROGRAM_FAILURE";
-	case CL_MAP_FAILURE:
-		return "CL_MAP_FAILURE";
-	case CL_INVALID_VALUE:
-		return "CL_INVALID_VALUE";
-	case CL_INVALID_DEVICE_TYPE:
-		return "CL_INVALID_DEVICE_TYPE";
-	case CL_INVALID_PLATFORM:
-		return "CL_INVALID_PLATFORM";
-	case CL_INVALID_DEVICE:
-		return "CL_INVALID_DEVICE";
-	case CL_INVALID_CONTEXT:
-		return "CL_INVALID_CONTEXT";
-	case CL_INVALID_QUEUE_PROPERTIES:
-		return "CL_INVALID_QUEUE_PROPERTIES";
-	case CL_INVALID_COMMAND_QUEUE:
-		return "CL_INVALID_COMMAND_QUEUE";
-	case CL_INVALID_HOST_PTR:
-		return "CL_INVALID_HOST_PTR";
-	case CL_INVALID_MEM_OBJECT:
-		return "CL_INVALID_MEM_OBJECT";
-	case CL_INVALID_IMAGE_FORMAT_DESCRIPTOR:
-		return "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR";
-	case CL_INVALID_IMAGE_SIZE:
-		return "CL_INVALID_IMAGE_SIZE";
-	case CL_INVALID_SAMPLER:
-		return "CL_INVALID_SAMPLER";
-	case CL_INVALID_BINARY:
-		return "CL_INVALID_BINARY";
-	case CL_INVALID_BUILD_OPTIONS:
-		return "CL_INVALID_BUILD_OPTIONS";
-	case CL_INVALID_PROGRAM:
-		return "CL_INVALID_PROGRAM";
-	case CL_INVALID_PROGRAM_EXECUTABLE:
-		return "CL_INVALID_PROGRAM_EXECUTABLE";
-	case CL_INVALID_KERNEL_NAME:
-		return "CL_INVALID_KERNEL_NAME";
-	case CL_INVALID_KERNEL_DEFINITION:
-		return "CL_INVALID_KERNEL_DEFINITION";
-	case CL_INVALID_KERNEL:
-		return "CL_INVALID_KERNEL";
-	case CL_INVALID_ARG_INDEX:
-		return "CL_INVALID_ARG_INDEX";
-	case CL_INVALID_ARG_VALUE:
-		return "CL_INVALID_ARG_VALUE";
-	case CL_INVALID_ARG_SIZE:
-		return "CL_INVALID_ARG_SIZE";
-	case CL_INVALID_KERNEL_ARGS:
-		return "CL_INVALID_KERNEL_ARGS";
-	case CL_INVALID_WORK_DIMENSION:
-		return "CL_INVALID_WORK_DIMENSION";
-	case CL_INVALID_WORK_GROUP_SIZE:
-		return "CL_INVALID_WORK_GROUP_SIZE";
-	case CL_INVALID_WORK_ITEM_SIZE:
-		return "CL_INVALID_WORK_ITEM_SIZE";
-	case CL_INVALID_GLOBAL_OFFSET:
-		return "CL_INVALID_GLOBAL_OFFSET";
-	case CL_INVALID_EVENT_WAIT_LIST:
-		return "CL_INVALID_EVENT_WAIT_LIST";
-	case CL_INVALID_EVENT:
-		return "CL_INVALID_EVENT";
-	case CL_INVALID_OPERATION:
-		return "CL_INVALID_OPERATION";
-	case CL_INVALID_GL_OBJECT:
-		return "CL_INVALID_GL_OBJECT";
-	case CL_INVALID_BUFFER_SIZE:
-		return "CL_INVALID_BUFFER_SIZE";
-	case CL_INVALID_MIP_LEVEL:
-		return "CL_INVALID_MIP_LEVEL";
-	case CL_INVALID_GLOBAL_WORK_SIZE:
-		return "CL_INVALID_GLOBAL_WORK_SIZE";
-#ifdef CL_VERSION_1_1
-	case CL_MISALIGNED_SUB_BUFFER_OFFSET:
-		return "CL_MISALIGNED_SUB_BUFFER_OFFSET";
-	case CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST:
-		return "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST";
-	case CL_INVALID_PROPERTY:
-		return "CL_INVALID_PROPERTY";
-#endif
-#ifdef CL_VERSION_1_2
-	case CL_COMPILE_PROGRAM_FAILURE:
-		return "CL_COMPILE_PROGRAM_FAILURE";
-	case CL_LINKER_NOT_AVAILABLE:
-		return "CL_LINKER_NOT_AVAILABLE";
-	case CL_LINK_PROGRAM_FAILURE:
-		return "CL_LINK_PROGRAM_FAILURE";
-	case CL_DEVICE_PARTITION_FAILED:
-		return "CL_DEVICE_PARTITION_FAILED";
-	case CL_KERNEL_ARG_INFO_NOT_AVAILABLE:
-		return "CL_KERNEL_ARG_INFO_NOT_AVAILABLE";
-	case CL_INVALID_IMAGE_DESCRIPTOR:
-		return "CL_INVALID_IMAGE_DESCRIPTOR";
-	case CL_INVALID_COMPILER_OPTIONS:
-		return "CL_INVALID_COMPILER_OPTIONS";
-	case CL_INVALID_LINKER_OPTIONS:
-		return "CL_INVALID_LINKER_OPTIONS";
-	case CL_INVALID_DEVICE_PARTITION_COUNT:
-		return "CL_INVALID_DEVICE_PARTITION_COUNT";
-#endif
-#ifdef CL_VERSION_2_0
-	case CL_INVALID_PIPE_SIZE:
-		return "CL_INVALID_PIPE_SIZE";
-	case CL_INVALID_DEVICE_QUEUE:
-		return "CL_INVALID_DEVICE_QUEUE";
-#endif
-#ifdef CL_VERSION_2_2
-	case CL_INVALID_SPEC_ID:
-		return "CL_INVALID_SPEC_ID";
-	case CL_MAX_SIZE_RESTRICTION_EXCEEDED:
-		return "CL_MAX_SIZE_RESTRICTION_EXCEEDED";
-#endif
-	default:
-		return "OpenCL unknown error";
-	}
+	long ret;
+	char *endptr;
+
+	if (minval > maxval)
+		die("argtonum: Invalid range\n");
+	if (!arg)
+		die("argtonum: Missing argument\n");
+
+	errno = 0;
+	ret = strtol(arg, &endptr, 10);
+	if (errno)
+		die("strtol: %s\n", strerror(errno));
+	if (arg == endptr || *endptr)
+		die("strtol: Invalid integer\n");
+
+	if (ret < minval || ret > maxval)
+		die("argtonum: Argument out of range\n");
+
+	return ret;
 }
diff --git a/src/util.h b/src/util.h
index d36746c..cf17bb3 100644
--- a/src/util.h
+++ b/src/util.h
@@ -13,7 +13,6 @@ GNU General Public License for more details.
 You should have received a copy of the GNU General Public License
 along with this program.  If not, see <http://www.gnu.org/licenses/>. */
 
-#include <CL/cl_platform.h>
 #include <stdio.h>
 
 void warn(const char *fmt, ...);
@@ -21,4 +20,4 @@ void die(const char *fmt, ...);
 
 void efwrite(const void *ptr, size_t size, size_t nmemb, FILE *stream);
 
-const char *cl_strerror(cl_int err);
+long argtonum(const char *arg, long minval, long maxval);
diff --git a/src/util_cl.c b/src/util_cl.c
new file mode 100644
index 0000000..5b571f5
--- /dev/null
+++ b/src/util_cl.c
@@ -0,0 +1,268 @@
+/* Copyright (C) 2024 Aiden Gall
+
+This program is free software: you can redistribute it and/or modify
+it under the terms of the GNU General Public License as published by
+the Free Software Foundation, either version 3 of the License, or
+(at your option) any later version.
+
+This program is distributed in the hope that it will be useful,
+but WITHOUT ANY WARRANTY; without even the implied warranty of
+MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
+GNU General Public License for more details.
+
+You should have received a copy of the GNU General Public License
+along with this program.  If not, see <http://www.gnu.org/licenses/>. */
+
+#include "util.h"
+
+#include <CL/cl.h>
+#include <CL/cl_platform.h>
+#include <CL/cl_version.h>
+#include <stdio.h>
+#include <stdlib.h>
+
+cl_platform_id *get_platforms(cl_uint *num_platforms);
+cl_device_id *get_devices(cl_platform_id platform, cl_uint *num_devices);
+
+cl_program compile_spirv_program(cl_context context, cl_device_id device,
+                                 const void *spirv_start, size_t spirv_size);
+
+const char *cl_strerror(cl_int err);
+
+cl_platform_id *
+get_platforms(cl_uint *const num_platforms)
+{
+	cl_int err;
+
+	cl_platform_id *platforms;
+	cl_uint len;
+
+	err = clGetPlatformIDs(0, NULL, &len);
+	if (err != CL_SUCCESS)
+		die("clGetPlatformIDs: %s\n", cl_strerror(err));
+	warn("number of platforms = %d\n", len);
+	if (len < 1)
+		die("clGetPlatformIDs: No OpenCL platforms\n");
+
+	platforms = calloc(len, sizeof(*platforms));
+	if (!platforms)
+		die("calloc: Out of memory\n");
+	err = clGetPlatformIDs(len, platforms, NULL);
+	if (err != CL_SUCCESS)
+		die("clGetPlatformIDs: %s\n", cl_strerror(err));
+
+	*num_platforms = len;
+	return platforms;
+}
+
+cl_device_id *
+get_devices(const cl_platform_id platform, cl_uint *const num_devices)
+{
+	cl_int err;
+
+	cl_device_id *devices;
+	cl_uint len;
+
+	err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 0, NULL, &len);
+	if (err != CL_SUCCESS)
+		die("clGetDeviceIDs: %s\n", cl_strerror(err));
+	warn("number of devices in platform = %d\n", len);
+	if (len < 1)
+		die("clGetDeviceIDs: No OpenCL devices in platform\n");
+
+	devices = calloc(len, sizeof(*devices));
+	if (!devices)
+		die("calloc: Out of memory\n");
+	err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, len, devices, NULL);
+	if (err != CL_SUCCESS)
+		die("clGetDeviceIDs: %s\n", cl_strerror(err));
+
+	*num_devices = len;
+	return devices;
+}
+
+cl_program
+compile_spirv_program(const cl_context context, const cl_device_id device,
+                      const void *const spirv_start, const size_t spirv_size)
+{
+	cl_int err, build_err;
+	cl_program program;
+
+	program = clCreateProgramWithIL(context, spirv_start, spirv_size, &err);
+	if (err != CL_SUCCESS)
+		die("clCreateProgramWithIL: %s\n", cl_strerror(err));
+
+	build_err = clBuildProgram(program, 1, &device, NULL, NULL, NULL);
+	if (build_err != CL_SUCCESS) {
+		cl_char *log;
+		size_t buffer_size;
+
+		err = clGetProgramBuildInfo(program, device,
+		                            CL_PROGRAM_BUILD_LOG, 0, NULL,
+		                            &buffer_size);
+		if (err != CL_SUCCESS)
+			die("clGetProgramBuildInfo: %s\n", cl_strerror(err));
+		if (buffer_size < 1)
+			die("clGetProgramBuildInfo: "
+			    "Build log buffer is empty\n");
+
+		log = calloc(buffer_size, sizeof(*log));
+		if (!log)
+			die("calloc: Out of memory\n");
+		err = clGetProgramBuildInfo(
+		        program, device, CL_PROGRAM_BUILD_LOG,
+		        sizeof(*log) * buffer_size, log, NULL);
+		if (err != CL_SUCCESS)
+			die("clGetProgramBuildInfo: %s\n", cl_strerror(err));
+
+		efwrite(log, sizeof(*log), buffer_size, stderr);
+
+		free(log);
+		die("\nclBuildProgram: %s\n", cl_strerror(build_err));
+	}
+
+	return program;
+}
+
+const char *
+cl_strerror(const cl_int err)
+{
+	switch (err) {
+	case CL_SUCCESS:
+		return "CL_SUCCESS";
+	case CL_DEVICE_NOT_FOUND:
+		return "CL_DEVICE_NOT_FOUND";
+	case CL_DEVICE_NOT_AVAILABLE:
+		return "CL_DEVICE_NOT_AVAILABLE";
+	case CL_COMPILER_NOT_AVAILABLE:
+		return "CL_COMPILER_NOT_AVAILABLE";
+	case CL_MEM_OBJECT_ALLOCATION_FAILURE:
+		return "CL_MEM_OBJECT_ALLOCATION_FAILURE";
+	case CL_OUT_OF_RESOURCES:
+		return "CL_OUT_OF_RESOURCES";
+	case CL_OUT_OF_HOST_MEMORY:
+		return "CL_OUT_OF_HOST_MEMORY";
+	case CL_PROFILING_INFO_NOT_AVAILABLE:
+		return "CL_PROFILING_INFO_NOT_AVAILABLE";
+	case CL_MEM_COPY_OVERLAP:
+		return "CL_MEM_COPY_OVERLAP";
+	case CL_IMAGE_FORMAT_MISMATCH:
+		return "CL_IMAGE_FORMAT_MISMATCH";
+	case CL_IMAGE_FORMAT_NOT_SUPPORTED:
+		return "CL_IMAGE_FORMAT_NOT_SUPPORTED";
+	case CL_BUILD_PROGRAM_FAILURE:
+		return "CL_BUILD_PROGRAM_FAILURE";
+	case CL_MAP_FAILURE:
+		return "CL_MAP_FAILURE";
+	case CL_INVALID_VALUE:
+		return "CL_INVALID_VALUE";
+	case CL_INVALID_DEVICE_TYPE:
+		return "CL_INVALID_DEVICE_TYPE";
+	case CL_INVALID_PLATFORM:
+		return "CL_INVALID_PLATFORM";
+	case CL_INVALID_DEVICE:
+		return "CL_INVALID_DEVICE";
+	case CL_INVALID_CONTEXT:
+		return "CL_INVALID_CONTEXT";
+	case CL_INVALID_QUEUE_PROPERTIES:
+		return "CL_INVALID_QUEUE_PROPERTIES";
+	case CL_INVALID_COMMAND_QUEUE:
+		return "CL_INVALID_COMMAND_QUEUE";
+	case CL_INVALID_HOST_PTR:
+		return "CL_INVALID_HOST_PTR";
+	case CL_INVALID_MEM_OBJECT:
+		return "CL_INVALID_MEM_OBJECT";
+	case CL_INVALID_IMAGE_FORMAT_DESCRIPTOR:
+		return "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR";
+	case CL_INVALID_IMAGE_SIZE:
+		return "CL_INVALID_IMAGE_SIZE";
+	case CL_INVALID_SAMPLER:
+		return "CL_INVALID_SAMPLER";
+	case CL_INVALID_BINARY:
+		return "CL_INVALID_BINARY";
+	case CL_INVALID_BUILD_OPTIONS:
+		return "CL_INVALID_BUILD_OPTIONS";
+	case CL_INVALID_PROGRAM:
+		return "CL_INVALID_PROGRAM";
+	case CL_INVALID_PROGRAM_EXECUTABLE:
+		return "CL_INVALID_PROGRAM_EXECUTABLE";
+	case CL_INVALID_KERNEL_NAME:
+		return "CL_INVALID_KERNEL_NAME";
+	case CL_INVALID_KERNEL_DEFINITION:
+		return "CL_INVALID_KERNEL_DEFINITION";
+	case CL_INVALID_KERNEL:
+		return "CL_INVALID_KERNEL";
+	case CL_INVALID_ARG_INDEX:
+		return "CL_INVALID_ARG_INDEX";
+	case CL_INVALID_ARG_VALUE:
+		return "CL_INVALID_ARG_VALUE";
+	case CL_INVALID_ARG_SIZE:
+		return "CL_INVALID_ARG_SIZE";
+	case CL_INVALID_KERNEL_ARGS:
+		return "CL_INVALID_KERNEL_ARGS";
+	case CL_INVALID_WORK_DIMENSION:
+		return "CL_INVALID_WORK_DIMENSION";
+	case CL_INVALID_WORK_GROUP_SIZE:
+		return "CL_INVALID_WORK_GROUP_SIZE";
+	case CL_INVALID_WORK_ITEM_SIZE:
+		return "CL_INVALID_WORK_ITEM_SIZE";
+	case CL_INVALID_GLOBAL_OFFSET:
+		return "CL_INVALID_GLOBAL_OFFSET";
+	case CL_INVALID_EVENT_WAIT_LIST:
+		return "CL_INVALID_EVENT_WAIT_LIST";
+	case CL_INVALID_EVENT:
+		return "CL_INVALID_EVENT";
+	case CL_INVALID_OPERATION:
+		return "CL_INVALID_OPERATION";
+	case CL_INVALID_GL_OBJECT:
+		return "CL_INVALID_GL_OBJECT";
+	case CL_INVALID_BUFFER_SIZE:
+		return "CL_INVALID_BUFFER_SIZE";
+	case CL_INVALID_MIP_LEVEL:
+		return "CL_INVALID_MIP_LEVEL";
+	case CL_INVALID_GLOBAL_WORK_SIZE:
+		return "CL_INVALID_GLOBAL_WORK_SIZE";
+#ifdef CL_VERSION_1_1
+	case CL_MISALIGNED_SUB_BUFFER_OFFSET:
+		return "CL_MISALIGNED_SUB_BUFFER_OFFSET";
+	case CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST:
+		return "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST";
+	case CL_INVALID_PROPERTY:
+		return "CL_INVALID_PROPERTY";
+#endif
+#ifdef CL_VERSION_1_2
+	case CL_COMPILE_PROGRAM_FAILURE:
+		return "CL_COMPILE_PROGRAM_FAILURE";
+	case CL_LINKER_NOT_AVAILABLE:
+		return "CL_LINKER_NOT_AVAILABLE";
+	case CL_LINK_PROGRAM_FAILURE:
+		return "CL_LINK_PROGRAM_FAILURE";
+	case CL_DEVICE_PARTITION_FAILED:
+		return "CL_DEVICE_PARTITION_FAILED";
+	case CL_KERNEL_ARG_INFO_NOT_AVAILABLE:
+		return "CL_KERNEL_ARG_INFO_NOT_AVAILABLE";
+	case CL_INVALID_IMAGE_DESCRIPTOR:
+		return "CL_INVALID_IMAGE_DESCRIPTOR";
+	case CL_INVALID_COMPILER_OPTIONS:
+		return "CL_INVALID_COMPILER_OPTIONS";
+	case CL_INVALID_LINKER_OPTIONS:
+		return "CL_INVALID_LINKER_OPTIONS";
+	case CL_INVALID_DEVICE_PARTITION_COUNT:
+		return "CL_INVALID_DEVICE_PARTITION_COUNT";
+#endif
+#ifdef CL_VERSION_2_0
+	case CL_INVALID_PIPE_SIZE:
+		return "CL_INVALID_PIPE_SIZE";
+	case CL_INVALID_DEVICE_QUEUE:
+		return "CL_INVALID_DEVICE_QUEUE";
+#endif
+#ifdef CL_VERSION_2_2
+	case CL_INVALID_SPEC_ID:
+		return "CL_INVALID_SPEC_ID";
+	case CL_MAX_SIZE_RESTRICTION_EXCEEDED:
+		return "CL_MAX_SIZE_RESTRICTION_EXCEEDED";
+#endif
+	default:
+		return "OpenCL unknown error";
+	}
+}
diff --git a/src/farbfeld.h b/src/util_cl.h
index 295c2a1..a6446e3 100644
--- a/src/farbfeld.h
+++ b/src/util_cl.h
@@ -13,8 +13,14 @@ GNU General Public License for more details.
 You should have received a copy of the GNU General Public License
 along with this program.  If not, see <http://www.gnu.org/licenses/>. */
 
-#include <stdint.h>
-#include <stdio.h>
+#include <CL/cl.h>
+#include <CL/cl_platform.h>
+#include <stddef.h>
 
-void farbfeld_write(FILE *fd, uint16_t *canvas, size_t canvas_size,
-                    uint32_t width, uint32_t height);
+cl_platform_id *get_platforms(cl_uint *num_platforms);
+cl_device_id *get_devices(cl_platform_id platform, cl_uint *num_devices);
+
+cl_program compile_spirv_program(cl_context context, cl_device_id device,
+                                 const void *spirv_start, size_t spirv_size);
+
+const char *cl_strerror(cl_int err);