void Cleanup(int iExitCode)
{
// cleanup allocated objects
//shrLog("\nStarting Cleanup...\n\n");
if(cPathAndName)free(cPathAndName);
if(cSourceCL)free(cSourceCL);
if(ScalseKernel)clReleaseKernel(ScalseKernel);
if(TransformKernel)clReleaseKernel(TransformKernel);
if(LongToShortKernel)clReleaseKernel(LongToShortKernel);
if(cpProgram)clReleaseProgram(cpProgram);
if(volumeSamplerLinear)clReleaseSampler(volumeSamplerLinear);
if(volumeSamplerNearest)clReleaseSampler(volumeSamplerNearest);
if(transferFuncSampler)clReleaseSampler(transferFuncSampler);
if(d_volumeArray)clReleaseMemObject(d_volumeArray);
if(d_transferFuncArray)clReleaseMemObject(d_transferFuncArray);
if(pbo_cl)clReleaseMemObject(pbo_cl);
if(d_transferpoint)clReleaseMemObject(d_transferpoint);
if(d_transpoint)clReleaseMemObject(d_transpoint);
if(d_scalepoint)clReleaseMemObject(d_scalepoint);
if(d_invViewMatrix)clReleaseMemObject(d_invViewMatrix);
if(cqCommandQueue)clReleaseCommandQueue(cqCommandQueue);
if(cxGPUContext)clReleaseContext(cxGPUContext);
if(d_boxmax)clReleaseMemObject(d_boxmax);
if (d_boxmin)clReleaseMemObject(d_boxmin);
// finalize logs and leave
shrLogEx(LOGBOTH | CLOSELOG, 0, "%s Exiting...\nPress <Enter> to Quit\n", cExecutableName);
#ifdef WIN32
getchar();
#endif
exit (iExitCode);
}
///
// Cleanup any created OpenCL resources
//
void Cleanup(cl_context context, cl_command_queue commandQueue,
cl_program program, cl_kernel kernel, cl_mem imageObjects[2],
cl_sampler sampler)
{
for (int i = 0; i < 2; i++)
{
if (imageObjects[i] != 0)
clReleaseMemObject(imageObjects[i]);
}
if (commandQueue != 0)
clReleaseCommandQueue(commandQueue);
if (kernel != 0)
clReleaseKernel(kernel);
if (program != 0)
clReleaseProgram(program);
if (sampler != 0)
clReleaseSampler(sampler);
if (context != 0)
clReleaseContext(context);
}
cl_int WINAPI wine_clReleaseSampler(cl_sampler sampler)
{
cl_int ret;
TRACE("\n");
ret = clReleaseSampler(sampler);
return ret;
}
~image_sampler()
{
if(m_sampler){
BOOST_COMPUTE_ASSERT_CL_SUCCESS(
clReleaseSampler(m_sampler)
);
}
}
void cleanKill(int errNumber){
clReleaseMemObject(inputImage);
clReleaseMemObject(outputImage);
clReleaseProgram(program);
clReleaseSampler(sampler);
clReleaseKernel(kernel);
clReleaseCommandQueue(commands);
clReleaseContext(context);
exit(errNumber);
}
image_sampler& operator=(image_sampler&& other) noexcept
{
if(m_sampler){
clReleaseSampler(m_sampler);
}
m_sampler = other.m_sampler;
other.m_sampler = 0;
return *this;
}
// Function to clean up and exit
//*****************************************************************************
void Cleanup(int iExitCode)
{
// cleanup allocated objects
//shrlog("\nStarting Cleanup...\n\n");
//if(cPathAndName)free(cPathAndName);
if(cSourceCL)free(cSourceCL);
if(ckKernel)clReleaseKernel(ckKernel);
if(cpProgram)clReleaseProgram(cpProgram);
if(volumeSamplerLinear)clReleaseSampler(volumeSamplerLinear);
if(volumeSamplerNearest)clReleaseSampler(volumeSamplerNearest);
if(transferFuncSampler)clReleaseSampler(transferFuncSampler);
if(d_volumeArray)clReleaseMemObject(d_volumeArray);
if(d_transferFuncArray)clReleaseMemObject(d_transferFuncArray);
if(pbo_cl)clReleaseMemObject(pbo_cl);
if(d_invViewMatrix)clReleaseMemObject(d_invViewMatrix);
if(cqCommandQueue)clReleaseCommandQueue(cqCommandQueue);
if(cxGPUContext)clReleaseContext(cxGPUContext);
glDeleteBuffersARB(1, &pbo);
//shrQAFinish2(bQATest, *pArgc, (const char **)pArgv, (iExitCode == 0) ? QA_PASSED : QA_FAILED);
// finalize logs and leave
if (bNoPrompt || bQATest)
{
//shrlogEx(LOGBOTH | CLOSELOG, 0, "%s Exiting...\n", cExecutableName);
}
else
{
//shrlogEx(LOGBOTH | CLOSELOG, 0, "%s Exiting...\nPress <Enter> to Quit\n", cExecutableName);
#ifdef WIN32
getchar();
#endif
}
exit (iExitCode);
}
image_sampler& operator=(const image_sampler &other)
{
if(this != &other){
if(m_sampler){
clReleaseSampler(m_sampler);
}
m_sampler = other.m_sampler;
if(m_sampler){
clRetainSampler(m_sampler);
}
}
return *this;
}
int main(int argc, char **argv)
{
/* test name */
char name[] = "test_image_query_funcs";
size_t global_work_size[1] = { 1 }, local_work_size[1]= { 1 };
size_t srcdir_length, name_length, filename_size;
char *filename = NULL;
char *source = NULL;
cl_device_id devices[1];
cl_context context = NULL;
cl_command_queue queue = NULL;
cl_program program = NULL;
cl_kernel kernel = NULL;
cl_int err;
/* image parameters */
cl_uchar4 *imageData;
cl_image_format image_format;
cl_image_desc image2_desc, image3_desc;
printf("Running test %s...\n", name);
memset(&image2_desc, 0, sizeof(cl_image_desc));
image2_desc.image_type = CL_MEM_OBJECT_IMAGE2D;
image2_desc.image_width = 2;
image2_desc.image_height = 4;
memset(&image3_desc, 0, sizeof(cl_image_desc));
image3_desc.image_type = CL_MEM_OBJECT_IMAGE3D;
image3_desc.image_width = 2;
image3_desc.image_height = 4;
image3_desc.image_depth = 8;
image_format.image_channel_order = CL_RGBA;
image_format.image_channel_data_type = CL_UNSIGNED_INT8;
imageData = (cl_uchar4*)malloc (4 * 4 * sizeof(cl_uchar4));
TEST_ASSERT (imageData != NULL && "out of host memory\n");
memset (imageData, 1, 4*4*sizeof(cl_uchar4));
/* determine file name of kernel source to load */
srcdir_length = strlen(SRCDIR);
name_length = strlen(name);
filename_size = srcdir_length + name_length + 16;
filename = (char *)malloc(filename_size + 1);
TEST_ASSERT (filename != NULL && "out of host memory\n");
snprintf(filename, filename_size, "%s/%s.cl", SRCDIR, name);
/* read source code */
source = poclu_read_file (filename);
TEST_ASSERT (source != NULL && "Kernel .cl not found.");
/* setup an OpenCL context and command queue using default device */
context = poclu_create_any_context();
TEST_ASSERT (context != NULL && "clCreateContextFromType call failed\n");
cl_sampler external_sampler = clCreateSampler (
context, CL_FALSE, CL_ADDRESS_NONE, CL_FILTER_NEAREST, &err);
CHECK_OPENCL_ERROR_IN ("clCreateSampler");
CHECK_CL_ERROR (clGetContextInfo (context, CL_CONTEXT_DEVICES,
sizeof (cl_device_id), devices, NULL));
queue = clCreateCommandQueue (context, devices[0], 0, &err);
CHECK_OPENCL_ERROR_IN ("clCreateCommandQueue");
/* Create image */
cl_mem image2
= clCreateImage (context, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR,
&image_format, &image2_desc, imageData, &err);
CHECK_OPENCL_ERROR_IN ("clCreateImage image2");
cl_mem image3
= clCreateImage (context, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR,
&image_format, &image3_desc, imageData, &err);
CHECK_OPENCL_ERROR_IN ("clCreateImage image3");
unsigned color[4] = { 2, 9, 11, 7 };
size_t orig[3] = { 0, 0, 0 };
size_t reg[3] = { 2, 4, 1 };
err = clEnqueueFillImage (queue, image2, color, orig, reg, 0, NULL, NULL);
CHECK_OPENCL_ERROR_IN ("clCreateImage image3");
/* create and build program */
program = clCreateProgramWithSource (context, 1, (const char **)&source,
NULL, &err);
CHECK_OPENCL_ERROR_IN ("clCreateProgramWithSource");
err = clBuildProgram (program, 0, NULL, NULL, NULL, NULL);
CHECK_OPENCL_ERROR_IN ("clBuildProgram");
/* execute the kernel with give name */
kernel = clCreateKernel (program, name, NULL);
CHECK_OPENCL_ERROR_IN ("clCreateKernel");
err = clSetKernelArg (kernel, 0, sizeof (cl_mem), &image2);
CHECK_OPENCL_ERROR_IN ("clSetKernelArg 0");
err = clSetKernelArg (kernel, 1, sizeof (cl_mem), &image3);
CHECK_OPENCL_ERROR_IN ("clSetKernelArg 1");
err = clSetKernelArg (kernel, 2, sizeof (cl_sampler), &external_sampler);
CHECK_OPENCL_ERROR_IN ("clSetKernelArg 2");
err = clEnqueueNDRangeKernel (queue, kernel, 1, NULL, global_work_size,
local_work_size, 0, NULL, NULL);
CHECK_OPENCL_ERROR_IN ("clEnqueueNDRangeKernel");
err = clFinish (queue);
CHECK_OPENCL_ERROR_IN ("clFinish");
clReleaseMemObject (image2);
clReleaseMemObject (image3);
clReleaseKernel (kernel);
clReleaseProgram (program);
clReleaseCommandQueue (queue);
clReleaseSampler (external_sampler);
clUnloadCompiler ();
clReleaseContext (context);
free (source);
free (filename);
free (imageData);
printf("OK\n");
return 0;
}
int main()
{
cl_platform_id platform = NULL;
cl_device_id device = NULL;
cl_context context = NULL;
cl_command_queue command_queue = NULL;
cl_program program = NULL;
cl_kernel kernel = NULL;
cl_int status = 0;
cl_event task_event, map_event;
cl_device_type dType = CL_DEVICE_TYPE_GPU;
cl_int image_width, image_height;
cl_float4 *result;
int i, j;
cl_mem clImage, out;
cl_bool support;
int pixels_read = 8;
//Setup the OpenCL Platform,
//Get the first available platform. Use it as the default platform
status = clGetPlatformIDs(1, &platform, NULL);
LOG_OCL_ERROR(status, "clGetPlatformIDs Failed" );
//Get the first available device
status = clGetDeviceIDs (platform, dType, 1, &device, NULL);
LOG_OCL_ERROR(status, "clGetDeviceIDs Failed" );
/*Check if the device support images */
clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof(support), &support, NULL);
if (support != CL_TRUE) {
std::cout <<"IMAGES not supported\n";
return 1;
}
//Create an execution context for the selected platform and device.
cl_context_properties contextProperty[3] =
{
CL_CONTEXT_PLATFORM,
(cl_context_properties)platform,
0
};
context = clCreateContextFromType(
contextProperty,
dType,
NULL,
NULL,
&status);
LOG_OCL_ERROR(status, "clCreateContextFromType Failed" );
/*Create command queue*/
command_queue = clCreateCommandQueue(context,
device,
0,
&status);
LOG_OCL_ERROR(status, "clCreateCommandQueue Failed" );
/* Create Image Object */
//Create OpenCL device input image with the format and descriptor as below
cl_image_format image_format;
image_format.image_channel_data_type = CL_FLOAT;
image_format.image_channel_order = CL_R;
//We create a 5 X 5 2D image
image_width = 5;
image_height = 5;
cl_image_desc image_desc;
image_desc.image_type = CL_MEM_OBJECT_IMAGE2D;
image_desc.image_width = image_width;
image_desc.image_height = image_height;
image_desc.image_depth = 1;
image_desc.image_array_size = 1;
image_desc.image_row_pitch = image_width*sizeof(float);
image_desc.image_slice_pitch = 25*sizeof(float);
image_desc.num_mip_levels = 0;
image_desc.num_samples = 0;
image_desc.buffer = NULL;
/* Create output buffer */
out = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(cl_float4)*pixels_read, NULL, &status);
LOG_OCL_ERROR(status, "clCreateBuffer Failed" );
size_t origin[] = {0,0,0}; /* Transfer target coordinate*/
size_t region[] = {image_width,image_height,1}; /* Size of object to be transferred */
float *data = (float *)malloc(image_width*image_height*sizeof(float));
float pixels[] = { /* Transfer Data */
10, 20, 10, 40, 50,
10, 20, 20, 40, 50,
10, 20, 30, 40, 50,
10, 20, 40, 40, 50,
10, 20, 50, 40, 50
};
memcpy(data, pixels, image_width*image_height*sizeof(float));
clImage = clCreateImage(context, CL_MEM_READ_ONLY|CL_MEM_USE_HOST_PTR, &image_format, &image_desc, pixels, &status);
LOG_OCL_ERROR(status, "clCreateImage Failed" );
/* If the image was not created using CL_MEM_USE_HOST_PTR, then you can write the image data to the device using the
clEnqueueWriteImage function. */
//status = clEnqueueWriteImage(command_queue, clImage, CL_TRUE, origin, region, 5*sizeof(float), 25*sizeof(float), data, 0, NULL, NULL);
//LOG_OCL_ERROR(status, "clCreateBuffer Failed" );
/* Build program */
program = clCreateProgramWithSource(context, 1, (const char **)&sample_image_kernel,
NULL, &status);
LOG_OCL_ERROR(status, "clCreateProgramWithSource Failed" );
// Build the program
status = clBuildProgram(program, 1, &device, "", NULL, NULL);
LOG_OCL_ERROR(status, "clBuildProgram Failed" );
if(status != CL_SUCCESS)
{
if(status == CL_BUILD_PROGRAM_FAILURE)
LOG_OCL_COMPILER_ERROR(program, device);
LOG_OCL_ERROR(status, "clBuildProgram Failed" );
}
printf("Printing the image pixels\n");
for (i=0; i<image_height; i++) {
for (j=0; j<image_width; j++) {
printf("%f ",data[i*image_width +j]);
}
printf("\n");
}
//Create kernel and set the kernel arguments
kernel = clCreateKernel(program, "image_test", &status);
clSetKernelArg(kernel, 0, sizeof(cl_mem), (void*)&clImage);
clSetKernelArg(kernel, 1, sizeof(cl_mem), (void*)&out);
/*********Image sampler with image repeated at every 1.0 normalized coordinate***********/
/*If host side sampler is not required the sampler objects can also be created on the kernel code.
Don't pass the thirsd argument to the kernel and create a sample object as shown below in the kernel code*/
//const sampler_t sampler = CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;
cl_sampler sampler = clCreateSampler (context,
CL_TRUE,
CL_ADDRESS_REPEAT,
CL_FILTER_NEAREST,
&status);
clSetKernelArg(kernel, 2, sizeof(cl_sampler), (void*)&sampler);
//Enqueue the kernel
status = clEnqueueTask(command_queue, kernel, 0, NULL, &task_event);
LOG_OCL_ERROR(status, "clEnqueueTask Failed" );
/* Map the result back to host address */
result = (cl_float4*)clEnqueueMapBuffer(command_queue, out, CL_TRUE, CL_MAP_READ, 0, sizeof(cl_float4)*pixels_read, 1, &task_event, &map_event, &status);
printf(" SAMPLER mode set to CL_ADDRESS_REPEAT | CL_FILTER_NEAREST\n");
printf("\nPixel values retreived based on the filter and Addressing mode selected\n");
printf("(float2)(0.5f,0.5f) = %f,%f,%f,%f\n",result[0].s[0],result[0].s[1],result[0].s[2],result[0].s[3]);
printf("(float2)(0.8f,0.5f) = %f,%f,%f,%f\n",result[1].s[0],result[1].s[1],result[1].s[2],result[1].s[3]);
printf("(float2)(1.3f,0.5f) = %f,%f,%f,%f\n",result[2].s[0],result[2].s[1],result[2].s[2],result[2].s[3]);
printf("(float2)(0.5f,0.5f) = %f,%f,%f,%f\n",result[3].s[0],result[3].s[1],result[3].s[2],result[3].s[3]);
printf("(float2)(0.5f,0.8f) = %f,%f,%f,%f\n",result[4].s[0],result[4].s[1],result[4].s[2],result[4].s[3]);
printf("(float2)(0.5f,1.3f) = %f,%f,%f,%f\n",result[5].s[0],result[5].s[1],result[5].s[2],result[5].s[3]);
printf("(float2)(4.5f,0.5f) = %f,%f,%f,%f\n",result[5].s[0],result[5].s[1],result[5].s[2],result[5].s[3]);
printf("(float2)(5.0f,0.5f) = %f,%f,%f,%f\n",result[7].s[0],result[7].s[1],result[7].s[2],result[7].s[3]);
clEnqueueUnmapMemObject(command_queue, out, result, 0, NULL, NULL);
clReleaseSampler(sampler);
/*********Image sampler with image mirrored at every 1.0 normalized coordinate***********/
sampler = clCreateSampler (context,
CL_TRUE,
CL_ADDRESS_MIRRORED_REPEAT,
CL_FILTER_LINEAR,
&status);
clSetKernelArg(kernel, 2, sizeof(cl_sampler), (void*)&sampler);
//Enqueue the kernel
status = clEnqueueTask(command_queue, kernel, 0, NULL, &task_event);
LOG_OCL_ERROR(status, "clEnqueueTask Failed" );
/* Map the result back to host address */
result = (cl_float4*)clEnqueueMapBuffer(command_queue, out, CL_TRUE, CL_MAP_READ, 0, sizeof(cl_float4)*pixels_read, 1, &task_event, &map_event, &status);
printf(" SAMPLER mode set to CL_ADDRESS_MIRRORED_REPEAT | CL_FILTER_LINEAR\n");
printf("\nPixel values retreived based on the filter and Addressing mode selected\n");
printf("(float2)(0.5f,0.5f) = %f,%f,%f,%f\n",result[0].s[0],result[0].s[1],result[0].s[2],result[0].s[3]);
printf("(float2)(0.8f,0.5f) = %f,%f,%f,%f\n",result[1].s[0],result[1].s[1],result[1].s[2],result[1].s[3]);
printf("(float2)(1.3f,0.5f) = %f,%f,%f,%f\n",result[2].s[0],result[2].s[1],result[2].s[2],result[2].s[3]);
printf("(float2)(0.5f,0.5f) = %f,%f,%f,%f\n",result[3].s[0],result[3].s[1],result[3].s[2],result[3].s[3]);
printf("(float2)(0.5f,0.8f) = %f,%f,%f,%f\n",result[4].s[0],result[4].s[1],result[4].s[2],result[4].s[3]);
printf("(float2)(0.5f,1.3f) = %f,%f,%f,%f\n",result[5].s[0],result[5].s[1],result[5].s[2],result[5].s[3]);
printf("(float2)(4.5f,0.5f) = %f,%f,%f,%f\n",result[5].s[0],result[5].s[1],result[5].s[2],result[5].s[3]);
printf("(float2)(5.0f,0.5f) = %f,%f,%f,%f\n",result[7].s[0],result[7].s[1],result[7].s[2],result[7].s[3]);
clEnqueueUnmapMemObject(command_queue, out, result, 0, NULL, NULL);
clReleaseSampler(sampler);
/********************/
//Free All OpenCL objects.
clReleaseMemObject(out);
clReleaseMemObject(clImage);
clReleaseKernel(kernel);
clReleaseProgram(program);
clReleaseCommandQueue(command_queue);
clReleaseContext(context);
return 0;
}
/*
* To ease testing, each kernel will be a Task kernel taking a pointer to an
* integer and running built-in functions. If an error is encountered, the
* integer pointed to by the arg will be set accordingly. If the kernel succeeds,
* this integer is set to 0.
*/
static uint32_t run_kernel(const char *source, TestCaseKind kind)
{
cl_platform_id platform = 0;
cl_device_id device;
cl_context ctx;
cl_command_queue queue;
cl_program program;
cl_int result;
cl_kernel kernel;
cl_event event;
cl_mem rs_buf;
cl_sampler sampler;
cl_mem mem1, mem2, mem3;
cl_image_format fmt;
unsigned char image2d_data[3*3*4] = {
255, 0, 0, 0, 0, 255, 0, 0, 128, 128, 128, 0,
0, 0, 255, 0, 255, 255, 0, 0, 0, 128, 0, 0,
255, 128, 0, 0, 128, 0, 255, 0, 0, 0, 0, 0
};
uint32_t rs = 0;
result = clGetDeviceIDs(platform, CL_DEVICE_TYPE_DEFAULT, 1, &device, 0);
if (result != CL_SUCCESS) return 65536;
ctx = clCreateContext(0, 1, &device, 0, 0, &result);
if (result != CL_SUCCESS) return 65537;
queue = clCreateCommandQueue(ctx, device, 0, &result);
if (result != CL_SUCCESS) return 65538;
program = clCreateProgramWithSource(ctx, 1, &source, 0, &result);
if (result != CL_SUCCESS) return 65539;
result = clBuildProgram(program, 1, &device, "", 0, 0);
if (result != CL_SUCCESS)
{
// Print log
char *log = 0;
size_t len = 0;
clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 0, 0, &len);
log = (char *)std::malloc(len);
clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, len, log, 0);
std::cout << log << std::endl;
std::free(log);
return 65540;
}
kernel = clCreateKernel(program, "test_case", &result);
if (result != CL_SUCCESS) return 65541;
// Create the result buffer
rs_buf = clCreateBuffer(ctx, CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR,
sizeof(rs), &rs, &result);
if (result != CL_SUCCESS) return 65542;
result = clSetKernelArg(kernel, 0, sizeof(cl_mem), &rs_buf);
if (result != CL_SUCCESS) return 65543;
// Kind
switch (kind)
{
case NormalKind:
break;
case SamplerKind:
sampler = clCreateSampler(ctx, 1, CL_ADDRESS_MIRRORED_REPEAT, CL_FILTER_NEAREST, &result);
if (result != CL_SUCCESS) return 65546;
result = clSetKernelArg(kernel, 1, sizeof(cl_sampler), &sampler);
if (result != CL_SUCCESS) return 65547;
break;
case ImageKind:
fmt.image_channel_data_type = CL_UNORM_INT8;
fmt.image_channel_order = CL_RGBA;
mem1 = clCreateImage2D(ctx, CL_MEM_WRITE_ONLY, &fmt, 4, 4, 0, 0, &result);
if (result != CL_SUCCESS) return 65548;
mem3 = clCreateImage2D(ctx, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
&fmt, 3, 3, 0, image2d_data, &result);
if (result != CL_SUCCESS) return 65548;
fmt.image_channel_data_type = CL_SIGNED_INT16;
mem2 = clCreateImage2D(ctx, CL_MEM_WRITE_ONLY, &fmt, 4, 4, 0, 0, &result);
if (result != CL_SUCCESS) return 65548;
result = clSetKernelArg(kernel, 1, sizeof(cl_mem), &mem1);
if (result != CL_SUCCESS) return 65549;
result = clSetKernelArg(kernel, 2, sizeof(cl_mem), &mem2);
if (result != CL_SUCCESS) return 65549;
result = clSetKernelArg(kernel, 3, sizeof(cl_mem), &mem3);
if (result != CL_SUCCESS) return 65549;
break;
default:
break;
}
if (kind == BarrierKind)
{
size_t local_size = 64;
size_t global_size = 64;
result = clEnqueueNDRangeKernel(queue, kernel, 1, 0, &global_size,
&local_size, 0, 0, &event);
if (result != CL_SUCCESS) return 65544;
}
else
{
result = clEnqueueTask(queue, kernel, 0, 0, &event);
if (result != CL_SUCCESS) return 65544;
}
result = clWaitForEvents(1, &event);
if (result != CL_SUCCESS) return 65545;
if (kind == SamplerKind) clReleaseSampler(sampler);
if (kind == ImageKind)
{
clReleaseMemObject(mem1);
clReleaseMemObject(mem2);
clReleaseMemObject(mem3);
}
clReleaseEvent(event);
clReleaseMemObject(rs_buf);
clReleaseKernel(kernel);
clReleaseProgram(program);
clReleaseCommandQueue(queue);
clReleaseContext(ctx);
return rs;
}
/**
* \brief Releases this Sampler.
*/
void ocl::Sampler::release()
{
if(this->_id == nullptr) throw std::runtime_error("Could not release Sampler.");
cl_int status = clReleaseSampler (this->_id);
OPENCL_SAFE_CALL(status);
}
int main(int argc, char *argv[])
{
CSKContext csk_context;
csk_context.init();
//显示device 扩展
char *device_extension = nullptr;
size_t extension_length = 0;
clGetDeviceInfo(csk_context._device, CL_DEVICE_EXTENSIONS, 0, nullptr,&extension_length);
device_extension = new char[extension_length + 4];
clGetDeviceInfo(csk_context._device, CL_DEVICE_EXTENSIONS, extension_length, device_extension,nullptr);
int index_j = 0;
int index_i = 0;
char extension[128];
while (index_j < extension_length)
{
if (device_extension[index_j] != ' ')
extension[index_i++] = device_extension[index_j];
else
{
extension[index_i] = '\0';
index_i = 0;
printf("%s\n", extension);
}
++index_j;
}
delete[] device_extension;
device_extension = nullptr;
int width = 0, height = 0;
cl_mem image_source = create_texture(csk_context._context, "image/global_img_heng4.tga",&width,&height);
CL_ASSERT(image_source != nullptr,"create_texture error.");
/*
*创建目标纹理
*/
int result = 0;
cl_image_format image_format;
image_format.image_channel_order = CL_RGBA;
image_format.image_channel_data_type = CL_UNORM_INT8;
cl_mem target_image = clCreateImage2D(csk_context._context, CL_MEM_WRITE_ONLY,&image_format,width,height,0,nullptr,&result);
CL_ASSERT(target_image!=nullptr && result ==CL_SUCCESS,"clCreateImage2D error.");
/*
*创建采样器对象
*/
cl_sampler sampler = clCreateSampler(csk_context._context,CL_FALSE,CL_ADDRESS_CLAMP_TO_EDGE,CL_FILTER_LINEAR,&result);
CL_ASSERT(sampler!=nullptr && result == CL_SUCCESS,"clCreateSampler error");
/*
*创建程序对象与内核
*/
cl_program program = csk_context.createCLProgram("kernel/image_filter.cl");
CL_ASSERT(program != nullptr,"createCLProgram error.");
cl_kernel kernel= clCreateKernel(program, "gaussion_filter", &result);
CL_ASSERT(kernel!=nullptr && result == CL_SUCCESS,"clCreateKernel error.");
/*设置内核参数
*/
result = clSetKernelArg(kernel, 0, sizeof(cl_mem), &image_source);
result |= clSetKernelArg(kernel, 1, sizeof(cl_mem), &target_image);
result |= clSetKernelArg(kernel, 2, sizeof(cl_sampler),&sampler);
result |= clSetKernelArg(kernel, 3, sizeof(cl_int), &width);
result |= clSetKernelArg(kernel, 4, sizeof(cl_int), &height);
CL_ASSERT(result == CL_SUCCESS,"clSetKernelArg error");
//派发队列
size_t localWorkSize[2] = {16,16};
size_t globalWorkSize[2] = {ceil(width/16.0f) * 16,ceil(height/16.0f) * 16};
cl_event event = nullptr;
result = clEnqueueNDRangeKernel(csk_context._commandQueue, kernel, 2, nullptr, globalWorkSize, localWorkSize, 0, nullptr, &event);
clWaitForEvents(1, &event);
char *buffer = new char[width * height * 4];
size_t origin_location[3] = {
0,0,0
};
size_t region_rect[3] = {
width,height,1,
};
result = clEnqueueReadImage(csk_context._commandQueue, target_image, CL_FALSE, origin_location, region_rect, 0, 0, buffer, 0, nullptr, &event);
clWaitForEvents(1, &event);
save_texture(csk_context._context, buffer, "image/global_img_heng4.png", width, height);
delete[] buffer;
buffer = nullptr;
clReleaseSampler(sampler);
clReleaseMemObject(image_source);
clReleaseMemObject(target_image);
clReleaseKernel(kernel);
clReleaseProgram(program);
getchar();
return 0;
}
void releaseSampler(v8::FunctionArgs args)
{
clReleaseSampler(cl_sampler(UnwrapPointer(args[0])));
}
