From 3b90152d969056682069254ddcdcc2d04148402b Mon Sep 17 00:00:00 2001
From: Aiden Gall <aiden@aidengall.xyz>
Date: Tue, 6 Feb 2024 08:49:55 +0000
Subject: switch canvas to floating point rgb

---
 src/cl/spirt.cl |  9 ++++-----
 src/spirt.c     | 15 +++++++--------
 2 files changed, 11 insertions(+), 13 deletions(-)

(limited to 'src')

diff --git a/src/cl/spirt.cl b/src/cl/spirt.cl
index 943f317..f55aa74 100644
--- a/src/cl/spirt.cl
+++ b/src/cl/spirt.cl
@@ -68,19 +68,18 @@ hit_sphere(const float3 center, const float radius, const struct ray ray)
 }
 
 void
-write_pixel(__global uchar *const pixel, float3 colour)
+write_pixel(__global float *const pixel, float3 colour)
 {
 	/* gamma correction */
-	colour = 256.0f * clamp(pow(colour, 1.0f/2.2f), 0.0f, 0.999f);
+	colour = clamp(pow(colour, 1.0f / 2.2f), 0.0f, 1.0f);
 
 	pixel[0] = colour.x;
 	pixel[1] = colour.y;
 	pixel[2] = colour.z;
-	pixel[3] = 255;
 }
 
 __kernel void
-ray_colour(__global uchar *const canvas, __global const float3 *const rays)
+ray_colour(__global float *const canvas, __global const float3 *const rays)
 {
 	struct ray ray;
 
@@ -114,5 +113,5 @@ ray_colour(__global uchar *const canvas, __global const float3 *const rays)
 		colour = (1.0f - a) * 1.0f + a * (float3)(0.5f, 0.7f, 1.0f);
 	}
 
-	write_pixel(canvas + 4 * (w * j + i), colour);
+	write_pixel(canvas + 3 * (w * j + i), colour);
 }
diff --git a/src/spirt.c b/src/spirt.c
index f9c6420..612c36e 100644
--- a/src/spirt.c
+++ b/src/spirt.c
@@ -51,9 +51,8 @@ struct camera {
 	cl_float3 corner00;
 };
 
-static unsigned char *kernel_render(struct kernel_context *runtime,
-                                    size_t image_width, size_t image_height,
-                                    struct camera cam);
+static float *kernel_render(struct kernel_context *runtime, size_t image_width,
+                            size_t image_height, struct camera cam);
 
 static struct kernel_context kernel_context_init(size_t platidx, size_t devidx);
 static void kernel_context_clean(struct kernel_context *runtime);
@@ -63,7 +62,7 @@ static struct opts get_args(int argc, char *argv[]);
 extern const char *SPIRV_START[];
 extern const char *SPIRV_END[];
 
-static unsigned char *
+static float *
 kernel_render(struct kernel_context *const runtime, const size_t image_width,
               const size_t image_height, const struct camera cam)
 {
@@ -71,11 +70,11 @@ kernel_render(struct kernel_context *const runtime, const size_t image_width,
 
 	size_t global_item_size[2];
 
-	unsigned char *h_canvas;
+	float *h_canvas;
 	cl_mem d_canvas, d_rays;
 	size_t rays_size, canvas_size;
 
-	canvas_size = (sizeof("RGBA") - 1) * image_width * image_height;
+	canvas_size = (sizeof("RGB") - 1) * image_width * image_height;
 	rays_size = 2 * image_width * image_height;
 
 	d_rays = clCreateBuffer(runtime->context, CL_MEM_READ_WRITE,
@@ -256,7 +255,7 @@ main(int argc, char *argv[])
 	struct opts opts;
 
 	struct kernel_context runtime;
-	unsigned char *h_canvas;
+	float *h_canvas;
 	struct camera cam;
 	float focal_length, viewport_width, viewport_height;
 
@@ -299,7 +298,7 @@ main(int argc, char *argv[])
 	img.data = h_canvas;
 	img.width = opts.width;
 	img.height = opts.height;
-	img.format = PIXELFORMAT_UNCOMPRESSED_R8G8B8A8;
+	img.format = PIXELFORMAT_UNCOMPRESSED_R32G32B32;
 	img.mipmaps = 1;
 
 	InitWindow(opts.width, opts.height, "spirt");
-- 
cgit v1.2.3