I have been following along with a tutorial online using OpenCL, where I am doing everything using python and pyOpenCL. As a stripped down example of my problem, I need to pass an array of C structs as an argument to an OpenCL kernel.
Here is an artificial example of OpenCL code:
typedef struct Test{
float a;
float3 b;
} Test;
__kernel void render_kernel(__constant Test *tests, const int width, const int height, const int num_structs, __global float3* output)
{
unsigned int work_item_id = get_global_id(0);
unsigned int x_coord = work_item_id % width;
unsigned int y_coord = work_item_id / width;
Test test = tests[0];
output[work_item_id] = test.b;
}
This will do something silly which is just give one of the float3 arrays as the output, but I just need to know I am actually getting the data through to the kernel properly.
I am trying to mimic an array of these structures on the python side with the following code:
class Test(ctypes.Structure):
_fields_ = [
("a", ctypes.c_float),
("b", (ctypes.c_float * 4))
]
class Test_Array(ctypes.Structure):
_fields_ = [("TEST_ARRAY", ctypes.POINTER(Test))]
def __init__(self, num_structs):
elems = (Test * num_structs)()
self.TEST_ARRAY = ctypes.cast(elems, ctypes.POINTER(Test))
self.elements = num_structs
for num in range(0, num_structs):
self.TEST_ARRAY[num].a = 1.0
self.TEST_ARRAY[num].b = (1.0, 0.0, 0.0, 1.0)
num_structs = 2
test_arr = Test_Array(num_structs)
#host buffer
color_out = np.empty((win.width * win.height, 4), dtype=np.float32)
cl_prog = CL()
cl_prog.load_program("shaders/struct.cl")
#device buffers
cl_structs = cl_prog.create_input_buffer(num_structs * ctypes.sizeof(Test))
cl_output = cl_prog.create_output_buffer(color_out.nbytes)
cl.enqueue_fill_buffer(cl_prog.queue, cl_structs, test_arr.TEST_ARRAY,
0, num_structs * ctypes.sizeof(Test))
global_work_size = (win.width * win.height,)
cl_prog.program.render_kernel(cl_prog.queue, global_work_size, None,
cl_structs, np.int32(win.width), np.int32(win.height),
np.int32(num_structs), cl_output)
cl_prog.retrieve_data(color_out, cl_output)
print(color_out)
This isn't really relevant as the functions in this class are just wrappers around pyOpenCL functions, but here is the CL class which is instantiated.
class CL:
def __init__(self):
self.platform = cl.get_platforms()[0]
self.device = self.platform.get_devices()[0]
self.ctx = cl.Context([self.device])
self.queue = cl.CommandQueue(self.ctx)
def load_program(self, file_path):
with open(file_path) as f:
src = f.read()
self.program = cl.Program(self.ctx, src).build()
def create_output_buffer(self, size):
"""
creates and returns a write only cl.Buffer of size bytes.
"""
mf = cl.mem_flags
return cl.Buffer(self.ctx, mf.WRITE_ONLY, size)
def create_input_buffer(self, size):
"""
returns a read only cl.Buffer of size bytes.
"""
mf = cl.mem_flags
return cl.Buffer(self.ctx, mf.READ_ONLY, size)
def retrieve_data(self, host_buffer, device_buffer):
"""
retrieves data from a buffer on the device, device_buffer, and copies it
over to host_buffer
"""
cl.enqueue_copy(self.queue, host_buffer, device_buffer)
def fill_buffer(self, memory, pattern, offset, size, wait_for=None):
"""
A wrapper around cl.enqueue_fill_buffer which uses self.queue
"""
cl.enqueue_fill_buffer(self.queue, memory, pattern, offset, size, wait_for)
def enqueue_copy(self, device_buffer, host_buffer):
cl.enqueue_copy(self.queue, device_buffer, host_buffer)
When I run the above code, it compiles and runs fine, but the information I get back from the buffer is just garbage that was already in memory. I can't tell if my problem is with alignment of the data, the way I am creating the array of ctypes structs, or what?
I am not attached to using a C array of C structs. I suspect there is a way to do this with numpy arrays, but I can't figure it out. Any way to properly get the data from the host to the device would be greatly appreciated.
