Hi all

This is a tough question to explain.

I am using SDL, Opengl and opencl.

I have a simple window that is rendering a quad over the full screen and am changing its texture each frame.

I am running the simple kernel below which is supposed to paint the screen with a nice sky blue gradient. It works, until i move the mouse or hit a key. It is the strangest thing. When i push a key or move the mouse, the colors change. It goes from blue to purple to black to pink and so on. I have implemented no key input code and none of my drawing code is linked in any way to key presses so i have no idea.

I know this might be tough to answer without seeing my full code. The only thing i can think of is that i dont know enough about opencl so maybe my kernel below is using some kind of system defined variable like U and V and that is causing it??

Or maybe its an sdl thing but my code is super simple, literally just a copy and paste from an sdl tutorial that sets up an opengl window.

Any help would be greatly appreciated!

Thanks

struct Sphere{	float3 pos;	float radius;};struct Ray{	float3 origin;	float3 dir;};float3 Colour(struct Ray *ray){	//lerp	float3 unit_direction = normalize(ray->dir);	float d = 0.5*(unit_direction.y + 1.0);	float3 ones = (float3)(1.0,1.0,1.0);	float3 blue = (float3)(0.5,0.7,1.0);	float3 tempa = ones * (1.0 - d);	float3 tempb = blue *d;	return (tempa + tempb);}__kernel void __main(__global float4* data, uint width, uint height) {	struct Sphere ball;	ball.pos = (float3)(0,0,-1);	ball.radius = 1.0f;	struct Ray ray;	ray.origin = (float3)(0,0,0);	float u, v;	int id = get_global_id(0);	float3 lowerLeft = (float3)(-2.0, -2.0, -1.0);	float3 horizontal = (float3)(4.0, 0.0, 0.0);	float3 vertical = (float3)(0.0, 4.0, 0.0);	float x = (float)(get_global_id(0) % width) / (float)(width);	float y = (float)(get_global_id(0) / width) / (float)(height);	float aspect = (float)(width) / (float)(height);	x = (x -0.5f)*aspect;	y = y -0.5f;	u = x / (float)width;	v = y / (float)height;	float3 U, V, COMB;	U = horizontal*u;	V = vertical*v;	COMB = U+V;	ray.dir = normalize(lowerLeft+COMB);	float3 colour = Colour(&ray);		//data[get_global_id(0)] = (float4)(0.5,0.5,0.5,1);	//data[get_global_id(0)] += (float4)(0.5,0.5,0,0);			data[get_global_id(0)] += (float4)(colour.x,colour.y,colour.z,1.0f);}

It may be a waste of time to game programmers.

I am writing a realtime raytracer. Modern opengl is overkill. A simple sdl window with opengl 2 buffer is all i require from opengl. Opencl is a small api that only does a handful of things.

I think learning to think in terms of parallelism is whats important and will transfer to any api in the future.

As for the code above, my mistake was data[get_global_id(0)] += (float4)(colour.x,colour.y,colour.z,1.0f);}

I was adding the color each frame lol.

I would tell you that learning OpenCL is currently a waste of time when it comes to game programming.

I think this is a very bad advise.

OpenCL is the easiest way to do and learn GPU compute - you need one line of code to upload to GPU memory, with Vulkan you need 50.

Same ratio for the simple thing of executing a kernel.

OpenGL compute shaders are also a lot more cumbersome to use than OpenCL,

and at least in the past OpenCL was twice(!) as fast on Nvidia, and a bit faster on AMD.

You would not make such a statement if you would have taken the time to test this yourself.

OpenCL is not popular in game dev, but that's simply our own fault.

AFAIK SpirV is core with OpenCL 2.1, NV is at 1.2, AMD is at 2.0.

Only Vulkan really uses it now, but no one writes SpirV directly - it's just intermediate byte code used by compilers, so not relevant for learning.

Feature wise OpenCL 1.2 language and GLSL are very similar - anything you learn can be adapted to the other easily.

OpenCL is the easiest way to do and learn GPU compute - you need one line of code to upload to GPU memory, with Vulkan you need 50.

Same ratio for the simple thing of executing a kernel.

OpenGL compute shaders are also a lot more cumbersome to use than OpenCL,

and at least in the past OpenCL was twice(!) as fast on Nvidia, and a bit faster on AMD.

You would not make such a statement if you would have taken the time to test this yourself.

OpenCL is not popular in game dev, but that's simply our own fault.

AFAIK SpirV is core with OpenCL 2.1, NV is at 1.2, AMD is at 2.0.

Only Vulkan really uses it now, but no one writes SpirV directly - it's just intermediate byte code used by compilers, so not relevant for learning.

Feature wise OpenCL 1.2 language and GLSL are very similar - anything you learn can be adapted to the other easily.

Very interesting. Thanks for the sharing.

I have one question about this. I always was believing (it seems wrongly) that OpenCL was not used in games because it implies a lot of lags due to having to send data and computations, having to wait for them to fulfill and then having to retrieve the computation results back. Is that wrong ? Is it worth now to do some heavy 3D computations with OpenCL even with having to synchronize all and having the GPU heavily working for the rendering calls, rather than using 2 or 3 CPU cores in a separate thread ? Of course the answer will vary depending on the GPU and CPU... But what are the common feedback here (if any) ?

Yes, the sync lags are a very big problem. For my work mostly i've had to read back at least a single number after a batch, just to set the amount of work for the next.

This resulted in almost 50% of time lost for nothing (i did not try to compensate with multiple queues, nor did i look at 2.0, nor did i check if rendering can fill those bubbles).

OpenGL was better and already had indirect dispatch - but it still was slower than OpenCL for me.

The only explantation is that OpenCL compilers were much better - Kernels executed faster although my code was 1:1 for CL and GLSL.

My experience about this is 2 years old.

Today i know that GLSL->SpirV->Vulkan is faster than OpenCL on AMD (approx 20% but up to 400%, and not counting the bubble lags).

With Vulkan or DX12 there is no point to use OpenCL if you want to do async compute and fine grained work sheduling.

But for something different like physics it may still makes sense in some cases.

If this physics stuff saturates the GPU good enough, there is no win from async compute anyways, so a single switch between Graphics API and OpenCL should not hurt.

Also the feture level of OpenCL 2.0 is beyond current Graphics APIs, but as long as there is no support from Nvidia you need a second CUDA code path for them.

However, i'm perfectly happy with Vulkan/DX12 prerecording command buffers and indirect dispatch - it really fixes the lag problem you mentioned.

So i agree there is no more need for OpenCL in shipping games, but it's still great for learning and prototyping, and i would quickly change my mind if the next driver update brings its performance above Vulkan.