CODEWITHSUNDEEP
javacopenclraytracingjocl
Anderson Gonçalves
Anderson Gonçalves

Reputation: 21

How do I pass an Java Object to my OpenCL kernel?

I'm developing a Java RayTracer with OpenCL. Im currently using JOCL to access the API. So, now I'm trying to make the user interact with the camera, and my plan was to make a camera that would be global to the kernel, and if the user interacts with it, the host would update the camera values and send these new values to the kernel. But I'm struggling to make these happen.

Below is my implementation of the camera, both in Java and C. I tried to match the values to avoid issues with memory alignment:

public class Camera {
    public float aspect_ratio;
    public int image_width;
    public int image_height;
    public float view_fov;
}

typedef struct {
    float aspect_ratio;
    int image_width;
    int image_height;
    float view_fov;
} Camera;

The method to convert the camera into a buffer that contains its attributes:

public ByteBuffer toByteBuffer(int bufferSize) {
    ByteBuffer buffer = ByteBuffer.allocateDirect(bufferSize);
    buffer.putFloat(aspect_ratio);
    buffer.putInt(image_width);
    buffer.putInt(image_height);
    buffer.putFloat(view_fov);
    buffer.flip();
    return buffer;
}

Here is how I create the buffer and send it to the kernel:

        int cameraBufferSize = camera.getCameraSize();
        ByteBuffer cameraBuffer = camera.toByteBuffer(cameraBufferSize);
        cl_mem cameraMem = clCreateBuffer(hostManager.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, cameraBuffer.capacity(), Pointer.to(cameraBuffer), null);

        long[] globalWorkSize = new long[2];
        globalWorkSize[0] = width;
        globalWorkSize[1] = height;

        clSetKernelArg(hostManager.getKernel(), 0, Sizeof.cl_mem, Pointer.to(outputImageMem));
        cL.clSetKernelArg(hostManager.getKernel(), 1, Sizeof.cl_mem, Pointer.to(cameraMem));
        clEnqueueNDRangeKernel(hostManager.getCommandQueue(), hostManager.getKernel(), 2, null, globalWorkSize, null, 0, null, null);

Here is my kernel code:

__kernel void render_kernel(__global uint* output, __global Camera* camera) {
    printf("aspect ratio: %f\n", camera->aspect_ratio);
    printf("image width: %d\n", camera->image_width);
    printf("image height: %d\n", camera->image_height);
    printf("view_fov: %f\n", camera->view_fov);
}

I even used clEnqueueReadBuffer() to read the buffer after sending it to the kernel, and the values appeared correct on the host side.

So, for a camera initialized with the following atributes:

Camera camera = new Camera(16.0 / 9.0, 800, 450, 90.0);

When executed, the kernel gives the following output:

aspect ratio: 0.000273 //expected aspect ratio -> 1.7777
image width: 327680 //expected width -> 800
image height: -805175296 //expected height -> 450
view_fov: 0.000000 //expected view_fov -> 90

So, is there a problem with how I'm creating the buffer and sending it, or is it related to the kernel side?

Upvotes: 2

Views: 66

Answers (3)

Anderson Gonçalves
Anderson Gonçalves

Reputation: 21

The issue was related to the endianness of my devices. Specifically, there was a difference in endianness between the host (written in Java) and the device (my GPU using OpenCL). The data transmitted through the camera buffer had its bytes reversed. Once I identified the problem and corrected the byte order, everything worked smoothly.

Upvotes: 0

Puri
Puri

Reputation: 1

cl_mem cameraMem = clCreateBuffer(hostManager.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, cameraBuffer.capacity(), Pointer.to(cameraBuffer), null);

clSetKernelArg(hostManager.getKernel(), 1, Sizeof.cl_mem, Pointer.to(cameraMem));


__kernel void render_kernel(__global uint* output, __global Camera* camera) {
    printf("aspect ratio: %f\n", camera->aspect_ratio);
    printf("image width: %d\n", camera->image_width);
    printf("image height: %d\n", camera->image_height);
    printf("view_fov: %f\n", camera->view_fov);
}
clEnqueueWriteBuffer(hostManager.getCommandQueue(), cameraMem, CL_TRUE, 0, cameraBuffer.capacity(), Pointer.to(cameraBuffer), 0, null, null);

Upvotes: 0

ProjectPhysX
ProjectPhysX

Reputation: 5754

Pass the four values for the camera as a single int array, on host side

int[] camera_buffer = new int[4];
camera_buffer[0] = Float.floatToIntBits(camera.aspect_ratio);
camera_buffer[1] = camera.width;
camera_buffer[2] = camera.height;
camera_buffer[3] = Float.floatToIntBits(camera.fov);

and also in the kernel parameters:

__global int* camera

Image width and image height will be correct as-is. For aspect_ratio and fov you need to call as_float (comes built-in OpenCL C already), to make the compiler treat the 32 bits of the int number as if they were encoding a floating-point number. This way you can mix/unmix data types in an array as long as they are all 32-bit.

const float aspect_ratio = as_float(camera[0]);
const int width = camera[1];
const int height = camera[2];
const float fov = as_float(camera[3]);

Passing structs in theory is also possible, but you will get alignmemt problems as soon as one of the member is not 32-bit, like for example a boolean or a double, or if the host compiler uses different alignment even if you only have 32-bit values. Never trust the compiler :)

Upvotes: 0

Related Questions