OpenCL - Empty result when loading a specific amount of input data

Right now I am programming some samples to understand OpenCL for future use. In the sample with the problem, I load some big 8bit Images and calculate pixelwise mean values.

Result[X,Y] = (Image1[X,Y] + Image2[X,Y] + ... ) / ImageCount

This works very well for 0 to 9 pictures. But when i load 10 or more images, the result is just a black image (all Pixels 0).

I thought that it may be a problem with the amount of memory. But with 10 pictures the image data is only 100MB. The graphic card is a 8600GTS with 256MB RAM.

Also i checked all of the error code returns and don't get anything different than a CL_SUCCESS.

Host program (Delphi, but I thing also C people can read it):

//Settings
MaxImg := 4;  //Images from 0..4 Count = 5
SetLength(InImgs,MaxImg+1);    //Array for images in Host memory
SetLength(GPUInMems,MaxImg+1); //Array for images in GPU memory

//Create Kernel
CLKernel := clCreateKernel(CLProgram, PChar('MainKernel'), @LastError);
//Create Queue
CLQueue := clCreateCommandQueue(CLContext, CLDevices[0].DeviceID, 0, @LastError);

//Load images
for I := 0 to MaxImg do
begin
  InImgs[I] := TImageMem.Create;
  InImgs[I].LoadFile('C:\Test\Img-' + IntToStr(I) + '.bmp');
  GPUInMems[I] := clCreateBuffer(CLContext, CL_MEM_READ_ONLY or CL_MEM_COPY_HOST_PTR, InImgs[I].MemSize, InImgs[I].Memory, @LastError);
end;

//Prepare Outputimage
OutImg := TImageMem.Create;
OutImg.LoadFile('C:\Test\CLTestOut.bmp');//Temporary solution to get right memory size and headers
GPUOutMem := clCreateBuffer(CLContext, CL_MEM_WRITE_ONLY, OutImg.MemSize, nil, @LastError);

//Set parameter for kernel call
LastError := clSetKernelArg(CLKernel, 0, sizeof(cl_mem), @GPUOutMem);   //Output image
LastError := clSetKernelArg(CLKernel, 1, sizeof(integer), @OutImg.Width);
LastError := clSetKernelArg(CLKernel, 2, sizeof(integer), @OutImg.Height);

//Add pointer to memory from images as parameters
for I := 0 to MaxImg do
begin
  LastError := clSetKernelArg(CLKernel, I+3, sizeof(cl_mem), @GPUInMems[I]);
end;

//Specify Group and Grid sizes
GlobalWSize[0]:= (OutImg.Width div 512 + 1) * 512; //Calc groups needed for resolution
LocalWSize[0] := 512; //Max WorkItems per group possible

//Execute and transfer ouput to host memory
LastError := clEnqueueNDRangeKernel(CLQueue, CLKernel, 1, nil, @GlobalWSize, @LocalWSize, 0, nil, nil);
LastError := clEnqueueReadBuffer(CLQueue, GPUOutMem, C开发者_如何学PythonL_TRUE, 0, OutImg.MemSize, OutImg.Memory, 0, nil, nil);

//Write output
OutImg.SaveFile('C:\Test\CLTestOut.bmp');

Kernel:

__kernel void MainKernel(
    __global uchar* ret,
    int xRes,
    int yRes,
    __global uchar* I0,
    __global uchar* I1,
    __global uchar* I2,
    __global uchar* I3,
    __global uchar* I4)
    {
            //Get line position
            int y = get_global_id(0);

            //Check inbound
            if (y >= yRes) return;

            //Set pointers to position
            ret += xRes * y;
            I0 += xRes * y;
            I1 += xRes * y;
            I2 += xRes * y;
            I3 += xRes * y;
            I4 += xRes * y;

            //Set val for each pixel in line
            for (int x = 0; x < xRes; ++x)
            {
              ret[x] = (I0[x] + I1[x] + I2[x] + I3[x] + I4[x]) / 5 ;
            }
    }

It would be great if somebody can tell me, why it is not working with more than 9 Images and why I don't get an errorcode.

Thanks for any help.

The Kernel Args SHOULD be static. Use a struct to load all your images, or maybe load all of them in an array form, and add some Args to the kernel setting the length of every image. To be able to separate inside the kernel each image.

I've seen lots of people that gets errors using 10+ Kernel Args.

Also as "Eric Bainville" said. You should add the images as a vector. Since you don't do any special procesing to the rows or cols.

It would be useful to check the error codes after each OpenCL call, so you can verify all buffer allocations are OK.

It will probably be faster to process your images by columns instead of rows: in your kernel, threads executing together will access memory at interval xRes, and memory access will be slow with this pattern. Running a 2D array of threads may be even faster.

EDIT. There may be an issue with the number of registers used, limiting the possible workgroup size. Check the max kernel workgroup size, and try to decrease workgroup size.