OpenCL* for CPU
Ask questions and share information on Intel® SDK for OpenCL™ Applications and OpenCL™ implementations for Intel® CPU
Announcements
This forum covers OpenCL* for CPU only. OpenCL* for GPU questions can be asked in the GPU Compute Software forum. Intel® FPGA SDK for OpenCL™ questions can be ask in the FPGA Intel® High Level Design forum.
1663 Discussions

Kernel not working anymore with latest driver.

LLess
Beginner
137 Views

I would like to check if anybody else is seeing things like that.

I have a simple kernel doing a simple texture copy (npower2 to p2) conversion that was working fine with OpenCL 1.1 and with OpenCL 1.2 I am getting a black image.

Other kernels seems to be working fine so that points at a bug somewhere but I have strictly no idea where.

//Texturing into a power of 2 target
float4 RunPass ( __read_only image2d_t input_texture, int2 pos, __constant ImageFilterParams *params)
{
    int2 input_dim =get_image_dim(input_texture);
    pos.x = pos.x - params->outputTextureCenterX + (input_dim.x/2);
    pos.y = pos.y - params->outputTextureCenterY + (input_dim.y/2);
    if ( pos.x < 0 || pos.x >= input_dim.x || pos.y < 0 || pos.y >input_dim.y )
        return (float4)(0.0f);   
    return read_imagef(input_texture,int_nearest_sampler2, pos );
}


__kernel void ImageFilterKernel(__read_only image2d_t texture, __write_only image2d_t output, __constant ImageFilterParams *params)
{
    int x = get_global_id(0);
    int y = get_global_id(1);
    int2 image_dim =get_image_dim(output);
    if  (( y < image_dim.y ) && (x < image_dim.x ))
    {
        int2 position = (int2)(x,y);
        float4 color = RunPass(texture,position,params);
        write_imagef(output,position,color);
    }
}


No failure reported by the functions that I am calling.

Laurent

0 Kudos
4 Replies
LLess
Beginner
137 Views

//Texturing into a power of 2 target
float4 RunPass ( __read_only image2d_t input_texture, int2 pos, int2 outputCenter)
{
    int2 input_dim =get_image_dim(input_texture);
    int2 p;
    p.x = pos.x - outputCenter.x + (input_dim.x/2);
    p.y = pos.y - outputCenter.y + (input_dim.y/2);
    if ( p.x < 0 || p.x >= input_dim.x || p.y < 0 || p.y >input_dim.y )
        return (float4)(0.0f);
   
    return read_imagef(input_texture,int_nearest_sampler2, p );
}

__kernel void ImageFilterKernel(__read_only image2d_t texture, __write_only image2d_t output)
{
    int x = get_global_id(0);
    int y = get_global_id(1);
    int2 image_dim =get_image_dim(output);
    int2 image_center_dim = (int2)(image_dim.x/2,image_dim.y/2);
    if  (( y < image_dim.y ) && (x < image_dim.x ))
    {
        int2 position = (int2)(x,y);
        float4 color = RunPass(texture,position,image_center_dim);
        write_imagef(output,position,color);
    }
}

But this works...

Go figure.

Raghupathi_M_Intel
137 Views

Laurent,

The kernel outputs incorrect image if you pass the ImageFilterParams* as a parameter but works ok if you compute the center in your kernel and pass to RunPass? Can you share your host code too?

Thanks,
Raghu

LLess
Beginner
137 Views

I can probably do that.

I will email you the code of that class and the other one needed but it is will be difficult to send you something that compiles since it would need almost the full project.

Laurent

LLess
Beginner
137 Views

I can now confirm that reverting back to the driver supporting only OpenCL 1.1 on the HD4000 fixed all my troubles.

Laurent.

Reply