Source code for a problem I'm experiencing with OpenCL's write_imagef() function on my NVIDIA GPU. The program creates a 2D image (CL_RGBA, CL_UNORM_INT8), invokes an OpenCL kernel that writes (1.0, 0.0, 1.0, 1.0) to pixel (0,0), and reads the image back to host memory. The output on my system is: CL_PLATFORM_NAME = NVIDIA CUDA CL_PLATFORM_VERSI…

write_imagef_test.cpp

#include <CL/cl.h>
#include <stdint.h>
#include <stdio.h>
#include <stdlib.h>
#include <string.h>

const char *writeImagefTestSrc = "__kernel void writeImagefTest(__write_only image2d_t outImg)\n\
{\n\
  write_imagef(outImg, (int2)(0,0), (float4)(1,0,1,1));\n\
}";

int main(int , char **)
{
	cl_int clError = CL_SUCCESS;

	cl_uint numPlatforms = 0;
	cl_platform_id platforms[32];
	clGetPlatformIDs(32,platforms,&numPlatforms);
	cl_platform_id oclPlatform = platforms[0];
	cl_context_properties contextProperties[] = {
		CL_CONTEXT_PLATFORM, (cl_context_properties)oclPlatform,
		0
	};
	cl_context oclContext = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU, NULL, NULL, &clError);
	cl_uint numDevices = 0;
	cl_device_id devices[32];
	clError = clGetContextInfo(oclContext, CL_CONTEXT_NUM_DEVICES, sizeof(cl_uint), &numDevices, NULL);
	clError = clGetContextInfo(oclContext, CL_CONTEXT_DEVICES, sizeof(devices), devices, NULL);
	cl_device_id oclDevice = devices[0];

	char strBuf[256];
	strBuf[255] = 0;
	clGetPlatformInfo(oclPlatform, CL_PLATFORM_NAME,    255, strBuf, NULL); printf("CL_PLATFORM_NAME    = %s\n", strBuf);
	clGetPlatformInfo(oclPlatform, CL_PLATFORM_VERSION, 255, strBuf, NULL); printf("CL_PLATFORM_VERSION = %s\n", strBuf);
	clGetPlatformInfo(oclPlatform, CL_PLATFORM_VENDOR,  255, strBuf, NULL); printf("CL_PLATFORM_VENDOR  = %s\n", strBuf);
	clGetDeviceInfo(oclDevice,     CL_DEVICE_NAME,      255, strBuf, NULL); printf("CL_DEVICE_NAME      = %s\n", strBuf);
	clGetDeviceInfo(oclDevice,     CL_DRIVER_VERSION,   255, strBuf, NULL); printf("CL_DRIVER_VERSION   = %s\n", strBuf);

	cl_command_queue oclQueue = clCreateCommandQueue(oclContext, oclDevice, CL_QUEUE_PROFILING_ENABLE, &clError);
	cl_program writeImagefTestPgm = clCreateProgramWithSource(oclContext, 1, &writeImagefTestSrc, NULL, &clError);
	clError = clBuildProgram(writeImagefTestPgm, 1, &oclDevice, "-Werror -cl-std=CL1.1", NULL, NULL);
	cl_kernel writeImagefTestKnl = clCreateKernel(writeImagefTestPgm, "writeImagefTest", &clError);
	cl_image_format imageFormat;
	imageFormat.image_channel_data_type = CL_UNORM_INT8;
	imageFormat.image_channel_order = CL_RGBA;
	const size_t imageWidth = 512, imageHeight = 512;
	cl_mem image = clCreateImage2D(oclContext, CL_MEM_WRITE_ONLY, &imageFormat, imageWidth, imageHeight, 0, NULL, &clError);
	size_t imageRowPitch = 0, imageSlicePitch = 0;
	clError = clGetImageInfo(image, CL_IMAGE_ROW_PITCH, sizeof(size_t), &imageRowPitch, NULL);
	clError = clGetImageInfo(image, CL_IMAGE_SLICE_PITCH, sizeof(size_t), &imageSlicePitch, NULL);

	clError = clSetKernelArg(writeImagefTestKnl, 0, sizeof(cl_mem), &image);
	clError = clEnqueueTask(oclQueue, writeImagefTestKnl, 0,NULL, NULL);
	size_t origin[3] = {0,0,0};
	size_t region[3] = {imageWidth,imageHeight,1};
	uint32_t *pixels = (uint32_t*)malloc(imageRowPitch*imageHeight);
	memset(pixels,0xCD,imageRowPitch*imageHeight);
	clError = clEnqueueReadImage(oclQueue, image, CL_TRUE, origin, region, imageRowPitch, imageSlicePitch, pixels, 0,NULL, NULL);
	clError = clFinish(oclQueue);
	printf("pixels[0]: expected=0xFFFF00FF, actual=0x%08X\n", pixels[0]);
	return 0;
}