This is on MacOS 10.6.1. Here's my kernel:

Code :
__kernel void Convolve(
	__read_only __global image2d_t img, 
	__write_only __global image2d_t newImg,
	__constant uint *filter, const int filterWidth, const int filterHeight,
	const int w, const int h)
	__private const sampler_t sampler = 
	uint4 val = read_imageui(img, sampler, (int2)(0, 0));

The problem is that the build log shows this:

Code :
Build Log:
Error while compiling the ptx module: CLH_ERROR_NO_BINARY_FOR_GPU
PTX Info log: 
PTX Error log:

And that's no fun! Commenting out the read_imageui line causes a successful compilation, but is hardly a solution. The specs say specifically (6.11. "Image memory objects that are being read by a kernel should be declared with the __read_only qualifier." Removing the __read_only qualifier from img also doesn't help. Removing the __global qualifier does help, but why should that be? Global is the only space that could possibly hold the image because of its size.

Should I file a bug with Apple? Or is it a case of ur doin it wrong?