summaryrefslogtreecommitdiff
diff options
context:
space:
mode:
authorAiden Gall <aiden@aidengall.xyz>2024-02-06 08:49:55 +0000
committerAiden Gall <aiden@aidengall.xyz>2024-02-06 09:03:40 +0000
commit3b90152d969056682069254ddcdcc2d04148402b (patch)
tree260ca18b1489e603ca4594405cba6dc24457eb37
parent71277d9af0a9c89ffeab47f01dcb1426f8ffa253 (diff)
switch canvas to floating point rgb
-rw-r--r--src/cl/spirt.cl9
-rw-r--r--src/spirt.c15
2 files changed, 11 insertions, 13 deletions
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");