Results 1 to 3 of 3

Thread: stripes over output image

  1. #1
    Junior Member
    Join Date
    Feb 2010
    Posts
    7

    stripes over output image

    i've decided to start new beginners thread because old one was in located inaccurate in 'technical discussions'.

    im trying to remove R and G color channel from 8bit/channel BMP file using image buffers.
    the problem is my output is covered in RGB vertical stripes (something like in CRT monitors).
    kernel code in my opinion is ok, the same as cl_image_format (CL_RGBA, CL_UNSIGNED_INT. i have doubts about passing image to imagebuffer, but i dont know how what causes the problem (i've checked the error flags and theres no error given).
    so my problem is how to remove this unwanted stripes.

    kernel.cl

    __kernel void copy(__read_only image2d_t imageIn,__write_only image2d_t imageOut)
    {
    const sampler_t sampler=CLK_NORMALIZED_COORDS_FALSE|CLK_ADDRESS_CL AMP|CLK_FI
    LTER_NEAREST;
    int gid0 = get_global_id(0);
    int gid1 = get_global_id(1);
    uint4 pixel;
    pixel=read_imageui(imageIn,sampler,(int2)(gid0,gid 1));
    pixel.x = 0;
    pixel.y = 0;
    write_imageui (imageOut,(int2)(gid0,gid1),pixel);
    }
    vec.cpp
    #include <oclUtils.h>

    const char* cSourceFile = "kernel.cl";

    // OpenCL Vars
    cl_context cxGPUContext; // OpenCL context
    cl_command_queue cqCommandQue; // OpenCL command que
    cl_device_id* cdDevices; // OpenCL device list
    cl_program cpProgram; // OpenCL program
    cl_kernel ckKernel; // OpenCL kernel
    cl_mem cmDevSrcA; // OpenCL device source buffer A
    cl_mem cmDevSrcB; // OpenCL device source buffer B
    cl_mem cmDevDst; // OpenCL device destination buffer

    size_t szGlobalWorkSize[2];
    size_t szLocalWorkSize[2];

    size_t szParmDataBytes; // Byte size of context information
    size_t szKernelLength; // Byte size of kernel code
    cl_int ciErr1, ciErr2; // Error code var
    char* cPathAndName = NULL; // var for full paths to data, src, etc.
    char* cSourceCL = NULL; // Buffer to hold source for compilation

    shrBOOL bNoPrompt = shrFALSE;


    // Main function
    // ************************************************** **********
    *********
    int main(int argc, char **argv)
    {

    // get command line arg for quick test, if provided
    bNoPrompt = shrCheckCmdLineFlag(argc, (const char**)argv, "noprompt");

    // start logs
    shrSetLogFileName ("vec.txt");

    size_t result=0;
    unsigned char header [54];

    FILE *input = fopen("in.bmp", "rb");
    result = fread(header,1,54,input);
    fseek (input, 54, SEEK_SET);

    unsigned char *tab;
    tab = (unsigned char*)malloc(512*512*4);
    result = fread(tab, 1, 3*512*512, input);
    fclose(input);
    void * image= (unsigned char *)tab;

    unsigned char *tab2;
    tab2 = (unsigned char*)malloc(512*512*4);
    void * image2= (unsigned char *)tab2;

    // Create the OpenCL context on a GPU device
    cxGPUContext = clCreateContextFromType(0, CL_DEVICE_TYPE_GPU, NULL, NULL, &ciErr1);
    shrLog(LOGBOTH, 0.0, "clCreateContextFromType...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clCreateContextFromType, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Get the list of GPU devices associated with context
    ciErr1 = clGetContextInfo(cxGPUContext, CL_CONTEXT_DEVICES, 0, NULL, &szParmDataBytes);
    cdDevices = (cl_device_id*)malloc(szParmDataBytes);
    ciErr1 |= clGetContextInfo(cxGPUContext, CL_CONTEXT_DEVICES, szParmDataBytes, cdDevices, NULL);
    shrLog(LOGBOTH, 0.0, "clGetContextInfo...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clGetContextInfo, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Create a command-queue
    cqCommandQue = clCreateCommandQueue(cxGPUContext, cdDevices[0], 0, &ciErr1);
    shrLog(LOGBOTH, 0.0, "clCreateCommandQueue...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clCreateCommandQueue, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Allocate the OpenCL buffer memory objects for source and result on the device GMEM

    size_t width = 512;
    size_t height = 512;
    size_t rowpitch = 0;

    cl_image_format format;
    format.image_channel_order = CL_RGBA;
    format.image_channel_data_type = CL_UNSIGNED_INT8;

    cl_mem_flags flags;
    flags = CL_MEM_READ_ONLY;
    shrLog(LOGBOTH, 0.0, "clCreateImage 1...\n");

    cl_mem myClImage = clCreateImage2D(
    cxGPUContext,
    flags,
    &format,
    width,
    height,
    rowpitch,
    0,
    &ciErr1
    );



    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clCreateImage2d 1, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    shrLog(LOGBOTH, 0.0, "clCreateImage 2...\n");
    cl_mem_flags flags2;
    flags2 = CL_MEM_WRITE_ONLY;

    cl_mem myClImage2 = clCreateImage2D(
    cxGPUContext,
    flags2,
    &format,
    width,
    height,
    rowpitch,
    0,//image2,
    &ciErr1
    );

    //ciErr1 |= ciErr2;
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clCreateImage2D 2, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Read the OpenCL kernel in from source file
    shrLog(LOGBOTH, 0.0, "oclLoadProgSource (%s)...\n", cSourceFile);
    cPathAndName = shrFindFilePath(cSourceFile, argv[0]);
    cSourceCL = oclLoadProgSource(cPathAndName, "", &szKernelLength);

    // Create the program
    cpProgram = clCreateProgramWithSource(cxGPUContext, 1, (const char **)&cSourceCL, &szKernelLength, &ciErr1);
    shrLog(LOGBOTH, 0.0, "clCreateProgramWithSource...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clCreateProgramWithSource, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Build the program
    ciErr1 = clBuildProgram(cpProgram, 0, NULL, NULL, NULL, NULL);
    shrLog(LOGBOTH, 0.0, "clBuildProgram...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clBuildProgram, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Create the kernel
    ckKernel = clCreateKernel(cpProgram, "copy", &ciErr1);
    shrLog(LOGBOTH, 0.0, "clCreateKernel (copy)...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clCreateKernel, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // Set the Argument values

    ciErr1 = clSetKernelArg(ckKernel, 0, sizeof(cl_mem), (void*)&myClImage);
    shrLog(LOGBOTH, 0.0, "clSetKernelArg 0...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clSetKernelArg, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    ciErr1 = clSetKernelArg(ckKernel, 1, sizeof(cl_mem), (void*)&myClImage2);
    shrLog(LOGBOTH, 0.0, "clSetKernelArg 1...\n");
    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clSetKernelArg, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // --------------------------------------------------------
    // Start Core sequence... copy input data to GPU, compute, copy results back

    size_t origin[3];
    origin[0] = 0;
    origin[1] = 0;
    origin[2] = 0;

    size_t region[3];
    region[0] = width;
    region[1] = height;
    region[2] = 1;

    shrLog(LOGBOTH, 0.0, "clEnqueueWriteImage...\n");

    ciErr1 = clEnqueueWriteImage (
    cqCommandQue,
    myClImage,
    CL_TRUE,
    origin,
    region,
    width*sizeof(char)*4, //size_t input_row_pitch,
    0, //width*sizeof(char)*height,//size_t input_slice_pitch,
    image, //const void * ptr,
    0,
    NULL,
    NULL
    );


    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clEnqueueWriteImage, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    // write of data to GPU device


    szGlobalWorkSize[0] = 512;
    szGlobalWorkSize[1] = 512;
    szLocalWorkSize[0] = 16;
    szLocalWorkSize[1] = 16;

    // Launch kernel
    ciErr1 = clEnqueueNDRangeKernel(cqCommandQue, ckKernel, 2, NULL, szGlobalWorkSize, szLocalWorkSize, 0, NULL, NULL);
    clFinish(cqCommandQue);
    shrLog(LOGBOTH, 0.0, "clEnqueueNDRangeKernel (VectorAdd)...\n");


    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "leRROR Error in clEnqueueNDRangeKernel, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}

    shrLog(LOGBOTH, 0.0, "clEnqueueReadImage ...\n");

    ciErr1 = clEnqueueReadImage (
    cqCommandQue,
    myClImage2,
    CL_TRUE,
    origin,
    region,
    0, //width*sizeof(char), //size_t row_pitch,
    0, //width*sizeof(char)*height,
    image2, //void *ptr,
    0,
    NULL,
    NULL
    );

    if (ciErr1 != CL_SUCCESS)
    {shrLog(LOGBOTH, 0.0, "Error in clEnqueueReadImage, Line %u in file %s !!!\n\n", __LINE__, __FILE__);}


    FILE *nk = fopen("outputt.bmp", "wb");
    fwrite(header,1,54,nk);
    fwrite(image2, 1, (3*512*512), nk);
    fclose(nk);

    shrLog(LOGBOTH, 0.0, "Finish success\n\n");

    }

    output file
    http://img641.imageshack.us/img641/9598/outputt.jpg
    zoomed output file
    http://img13.imageshack.us/img13/6292/zooms.jpg
    input file
    http://www.bilsen.com/aic/tests/lena/lena.bmp

  2. #2
    Junior Member
    Join Date
    Feb 2010
    Posts
    7

    Re: stripes over output image

    if anyone's interested - the problem was the file 24bpp, with 32bpp everything works perfectly.

  3. #3
    Senior Member
    Join Date
    Jul 2009
    Location
    Northern Europe
    Posts
    311

    Re: stripes over output image

    I was about to say that it sounded like you were missing a channel in your RGBA.

Similar Threads

  1. How to get point stripes function to with textured polygons?
    By mobilevisuals in forum OpenGL ES 1.X - fixed function hardware
    Replies: 3
    Last Post: 10-26-2009, 12:35 AM
  2. Vincent gives upside down mirror image output!!!
    By GreatApe in forum OpenGL ES general technical discussions
    Replies: 4
    Last Post: 04-26-2005, 04:09 PM

Posting Permissions

  • You may not post new threads
  • You may not post replies
  • You may not post attachments
  • You may not edit your posts
  •