at 29.10.2011 18:15, Andreas Kloeckner wrote:
This is unlikely to be an issue with PyOpenCL, I think--it's more likely that the hardware or the driver is at fault. My first try would be to replace the image write with a write to a regular 2D array. Hope that helps, Andreas
Hi Andreas,

thx for your fast reply.
I'm sorry i didn't wrote that i allready tested your proposal.
But i changed the script this way:

import pyopencl as cl
import numpy as np
import cv2 # OpenCV 2.3.1

Img = cv2.imread("Test.jpg")
Img = cv2.cvtColor(Img, cv2.COLOR_BGR2GRAY)

ctx = cl.create_some_context()
queue = cl.CommandQueue(ctx)
mf = cl.mem_flags

dev_Img = cl.Image(ctx,
                    mf.READ_ONLY | mf.COPY_HOST_PTR,
cl.ImageFormat(cl.channel_order.R, cl.channel_type.UNSIGNED_INT8),
                    hostbuf=Img)
OutImg = np.empty(shape=Img.shape, dtype=np.uint8) # create Output-Image
dev_OutImg = cl.Buffer(ctx,
                        mf.WRITE_ONLY | mf.ALLOC_HOST_PTR,
                        size=OutImg.nbytes)

prg = cl.Program(ctx, """
const sampler_t smp = CLK_NORMALIZED_COORDS_TRUE | CLK_FILTER_LINEAR | CLK_ADDRESS_NONE;

    __kernel void ImageCopy(__read_only image2d_t Img, __global uchar* Out)
    {
        const int2 dims = get_image_dim(Img);
        const int2 Coords = (int2)(get_global_id(0), get_global_id(1));
const float2 NormCoords = (convert_float2(Coords) + (float2)(0.5f) + (float2)(0.4f)) / convert_float2(dims); // = (x + 0.5 + 0.4)/w ; (y + 0.5 + 0.2)/h

        uint4 Pixel = read_imageui(Img, smp, NormCoords);
        Out[Coords.x + dims.x * Coords.y] = Pixel.x;
    }
    """).build()

prg.ImageCopy(queue, Img.shape, None, dev_Img, dev_OutImg)
cl.enqueue_read_buffer(queue, dev_OutImg, OutImg).wait()
cv2.imwrite("Out.jpg", OutImg)


The problem still exists. It seems that the problem is the reading from the picture - not the writing. Because of my various tests it seems for me that the problem can be that the shape of the picture is inverted when reading. The image-array seems to be treated with width=height and height=width. At bilinear interpolation 4 pixels where read from 2 different lines. It seems that the 2 pixels from current line read out correct and the 2 other pixels (at the other line) are read out shifted.
That would explain these strange effect.

Anyway your answer is a big help for me. You think that is a hardware or driver bug too so i don't need to try further. It's not the problem to write a own bilinear interpolation but the image2d-read-out should be faster - that's a pity.
I posted this because i didn't know if this effect is my fault or a bug.
You acknowledged my assumption - that helps.
But i still favour of other good ideas to this problem. :-)

Thx for your reply and thx for your good work with pyopencl

Andreas

_______________________________________________
PyOpenCL mailing list
[email protected]
http://lists.tiker.net/listinfo/pyopencl

Reply via email to