**1.**openclModel.cpp
#include <string.h>
#include <stdio.h>
#include "bufferCpy.h"
#include "CLInit.h"
#include "InitKernel.h"
#include "MemBuffer.h"
int main()
{
initCLEnv();
InitKernelList();
InitMem(10, 10);
bufferCopy(MEM_input, MEM_output);
UnitMem();
UnitKernelList();
unitCLEnv();
getchar();
return 0;
}
**2.**CLInit.h CLInit.hc
#ifndef _CLINIT_H_
#define _CLINIT_H_
#include <CL\cl.h>
typedef struct _CLENV_{
cl_platform_id * platform;
cl_uint numOfPlatform;
cl_device_id * device;
cl_uint numOfDevice;
cl_context ctx;
cl_command_queue * commandQueue;
int inited;
}clEnv;
#ifdef __cplusplus
extern "C"
{
#endif
clEnv * getCLEnv();
int initCLEnv();
int unitCLEnv();
#ifdef __cplusplus
}
#endif
#endif
#include <malloc.h>
#include <stdio.h>
#include <string.h>
#include "CLInit.h"
#pragma warning(disable : 4996)
static clEnv gCLEnv = {0};
clEnv * getCLEnv(){
if(gCLEnv.inited == 1)
{
return &gCLEnv;
}else
{
return NULL;
}
}
static int initPlatform(){
cl_int err = 0;
err = clGetPlatformIDs(0, NULL, &(gCLEnv.numOfPlatform));
if(err != CL_SUCCESS || 0 == gCLEnv.numOfPlatform)
{
printf("clGetPlatformIDs error. error code = %d\n", err);
return -1;
}
gCLEnv.platform = (cl_platform_id*)malloc(sizeof(cl_platform_id) * gCLEnv.numOfPlatform);
if(NULL == gCLEnv.platform){
printf("malloc platform error.\n");
return -2;
}
err = clGetPlatformIDs(gCLEnv.numOfPlatform, gCLEnv.platform, NULL);
if(CL_SUCCESS != err)
{
printf("clGetPlatformIDs error. error code = %d\n", err);
return -3;
}
return 0;
}
static int initDevice(){
cl_int err = 0;
cl_uint i = 0;
cl_platform_id temp;
for(i = 0; i < gCLEnv.numOfPlatform; i ++)
{
char platformInfo[2048];
size_t num = 0;
clGetPlatformInfo(gCLEnv.platform[i], CL_PLATFORM_VENDOR, 2048, platformInfo, &num);
platformInfo[num] = '\0';
printf("platform number = %d , %s\n", i, platformInfo);
}
for(i = 0; i < gCLEnv.numOfPlatform; i ++){
char platformInfo[2048];
size_t num = 0;
clGetPlatformInfo(gCLEnv.platform[i], CL_PLATFORM_VENDOR, 2048, platformInfo, &num);
platformInfo[num] = '\0';
if(strcmp(platformInfo, "Advanced Micro Devices, Inc."))
{
continue;
}
err = clGetDeviceIDs(gCLEnv.platform[i], CL_DEVICE_TYPE_GPU, 0, NULL, &gCLEnv.numOfDevice);
if(0 != gCLEnv.numOfDevice)
{
gCLEnv.device = (cl_device_id*)malloc(gCLEnv.numOfDevice * sizeof(cl_device_id));
if(NULL == gCLEnv.device){
printf("malloc devices error.\n");
return -1;
}
err = clGetDeviceIDs(gCLEnv.platform[i], CL_DEVICE_TYPE_GPU, gCLEnv.numOfDevice, gCLEnv.device, 0);
if(CL_SUCCESS != err){
printf("clGetDeviceIDs error. error code = %d\n", err);
return -2;
}
temp = gCLEnv.platform[0];
gCLEnv.platform[0] = gCLEnv.platform[i];
gCLEnv.platform[i] = temp;
{
char deviceInfo[2048];
size_t num = 0;
clGetDeviceInfo (gCLEnv.device[0],CL_DEVICE_NAME,2048, deviceInfo, &num);
deviceInfo[num] = '\0';
printf("select device name = %s\n", deviceInfo);
}
{
char platformInfo[2048];
size_t num = 0;
clGetPlatformInfo(gCLEnv.platform[0],CL_PLATFORM_VENDOR, 2048, platformInfo, &num);
platformInfo[num] = '\0';
printf("select platform name = %s\n", platformInfo);
}
break;
}
}
if(0 != gCLEnv.numOfDevice)
{
return 0;
}
for(i = 0; i < gCLEnv.numOfPlatform; i ++){
err = clGetDeviceIDs(gCLEnv.platform[i], CL_DEVICE_TYPE_CPU, 0, NULL, &gCLEnv.numOfDevice);
if(0 != gCLEnv.numOfDevice){
gCLEnv.device = (cl_device_id*)malloc(gCLEnv.numOfDevice * sizeof(cl_device_id));
if(NULL == gCLEnv.device){
printf("malloc device error.\n");
return -1;
}
err = clGetDeviceIDs(gCLEnv.platform[i], CL_DEVICE_TYPE_CPU, gCLEnv.numOfDevice, gCLEnv.device, 0);
if(CL_SUCCESS != err){
printf("clGetDeviceIDs error. error code = %d\n", err);
return -2;
}
temp = gCLEnv.platform[0];
gCLEnv.platform[0] = gCLEnv.platform[i];
gCLEnv.platform[i] = temp;
{
char deviceInfo[2048];
size_t num = 0;
clGetDeviceInfo (gCLEnv.device[0],CL_DEVICE_NAME,2048, deviceInfo, &num);
deviceInfo[num] = '\0';
printf("select device name = %s\n", deviceInfo);
}
{
char platformInfo[2048];
size_t num = 0;
clGetPlatformInfo(gCLEnv.platform[0],CL_PLATFORM_VENDOR, 2048, platformInfo, &num);
platformInfo[num] = '\0';
printf("select platform name = %s\n", platformInfo);
}
break;
}
}
return 0;
}
static int initContext(){
cl_int err = 0;
cl_context_properties prop[3] = {CL_CONTEXT_PLATFORM, (cl_context_properties)gCLEnv.platform[0], 0};
gCLEnv.ctx = clCreateContext(prop, gCLEnv.numOfDevice, gCLEnv.device, NULL, NULL, &err);
if(CL_SUCCESS != err){
printf("clCreateContext error. error code = %d\n", err);
return -1;
}
return 0;
}
static int initCommandQueue(){
int err = 0;
cl_uint i = 0;
gCLEnv.commandQueue = (cl_command_queue*)malloc(sizeof(cl_command_queue) * gCLEnv.numOfDevice);
if(NULL == gCLEnv.commandQueue){
printf("malloc commandQueue error\n");
return -1;
}
for(i = 0; i < gCLEnv.numOfDevice; i ++){
gCLEnv.commandQueue[i] = clCreateCommandQueue(gCLEnv.ctx, gCLEnv.device[i], CL_QUEUE_PROFILING_ENABLE, &err);
if(CL_SUCCESS != err){
printf("clCreateCommandQueue error. error code = %d\n", err);
return -2;
}
}
return 0;
}
int initCLEnv(){
int err = 0;
if((err = initPlatform()) != 0)
{
printf("initplatform error.\n");
return -1;
}
if((err = initDevice()) != 0)
{
printf("initDevice error.\n");
return -2;
}
if((err = initContext()) != 0)
{
printf("initContext error.\n");
return -3;
}
if((err = initCommandQueue()) != 0)
{
printf("initCommandQueue error.\n");
return -4;
}
gCLEnv.inited = 1;
return 0;
}
int unitCLEnv(){
cl_uint i = 0;
for(i = 0; i < gCLEnv.numOfDevice; i ++){
clReleaseCommandQueue(gCLEnv.commandQueue[i]);
}
free(gCLEnv.commandQueue);
clReleaseContext(gCLEnv.ctx);
for(i = 0; i < gCLEnv.numOfDevice; i ++){
clReleaseDevice(gCLEnv.device[i]);
}
free(gCLEnv.device);
free(gCLEnv.platform);
memset(&gCLEnv, 0, sizeof(clEnv));
return 0;
}
**3.**InitKernel.h InitKernel.c
#ifndef _INITKERNEL_H_
#define _INITKERNEL_H_
//
//Auto generated file
//
typedef enum _KERNEL_List_{
KERNEL_Start = -1,
KERNEL_buffercpy,
KERNEL_Number
}KERNEL_List;
#ifdef __cplusplus
extern "C"
{
#endif
cl_kernel* getKernelList();
int InitKernelList();
int UnitKernelList();
#ifdef __cplusplus
}
#endif
//
//end
//
#endif
#include <CL\cl.h>
#include <stdio.h>
#include "InitKernel.h"
#include "CLInit.h"
#include "kernelSrc.c"
static cl_kernel kernelList[KERNEL_Number] = {0};
cl_kernel* getKernelList(){
return kernelList;
}
int InitKernelList(){
cl_int err = 0;
int i = 0;
clEnv * env = getCLEnv();
cl_program m_program = clCreateProgramWithSource(env->ctx, 1, &kernelsrc, NULL, &err);
if(CL_SUCCESS != err){
printf("clCreateProgramWidthSource error. error code = %d\n", err);
return -1;
}
err = clBuildProgram(m_program, env->numOfDevice, env->device, NULL, NULL, NULL);
if(CL_SUCCESS != err){
printf("clBuildProgram error. error code = %d\n", err);
return err;
}
for(i = 0; i < KERNEL_Number; i ++){
kernelList[i] = clCreateKernel(m_program, kernelName[i], &err);
if(CL_SUCCESS != err){
printf("clCreateKernel %s error. error code = %d\n", kernelName[i], err);
return err;
}
}
clReleaseProgram(m_program);
return 0;
}
int UnitKernelList(){
int i = 0;
for(i = 0; i < KERNEL_Number; i ++){
clReleaseKernel(kernelList[i]);
kernelList[i] = 0;
}
return 0;
}
**4.**MemBuffer.h MemBuffer.c
#ifndef _MEMBUFFER_H_
#define _MEMBUFFER_H_
#include <CL\cl.h>
typedef enum _MEMLIST_{
MEM_START = -1,
MEM_input,
MEM_output,
MEM_Number,
}MEM_LIST;
#ifdef __cplusplus
extern "C"
{
#endif
cl_mem * getMemList();
int InitMem(int width, int height);
int UnitMem();
int loadMemData(MEM_LIST memBuffer, int width, int height, float * data);
#ifdef __cplusplus
}
#endif
#endif
#include <stdio.h>
#include "MemBuffer.h"
#include "CLInit.h"
static cl_mem memList[MEM_Number] = {0};
cl_mem* getMemList(){
return memList;
}
int InitMem(int width, int height){
cl_int err = 0;
clEnv * env = NULL;
env = getCLEnv();
memList[MEM_input] = clCreateBuffer(env->ctx, CL_MEM_READ_WRITE, width * height * sizeof(float), NULL, &err);
if(CL_SUCCESS != err){
goto ERROR;
}
memList[MEM_output] = clCreateBuffer(env->ctx, CL_MEM_READ_WRITE, width * height * sizeof(float), NULL, &err);
if(CL_SUCCESS != err){
goto ERROR;
}
return 0;
ERROR:
printf("clCreateBuffer error.error code = %d", err);
return -1;
return 0;
}
int UnitMem(){
int i = 0;
for(i = 0; i < MEM_Number; i ++){
if(memList[i] != NULL){
clReleaseMemObject(memList[i]);
memList[i] = 0;
}
}
return 0;
}
int loadMemData(MEM_LIST memBuffer, int width, int height, float * data){
cl_int err = 0;
clEnv * env = NULL;
env = getCLEnv();
err = clEnqueueWriteBuffer(env->commandQueue[0], memList[MEM_input], CL_FALSE, 0, width * height * sizeof(float), data, 0, NULL, NULL);
if(CL_SUCCESS != err){
printf("write buffer error. error code = %d\n", err);
return -1;
}
return 0;
}
5. bufferCpy.h bufferCpy.c
#ifndef _BUFFERCPY_H_
#define _BUFFERCPY_H_
#include "MemBuffer.h"
#ifdef __cplusplus
extern "C"
{
#endif
int bufferCopy(MEM_LIST input, MEM_LIST output);
#ifdef __cplusplus
}
#endif
#endif
#include <malloc.h>
#include <stdio.h>
#include "bufferCpy.h"
#include "CLInit.h"
#include "InitKernel.h"
int bufferCopy(MEM_LIST input, MEM_LIST output){
int width = 10, height = 10;
float * inData = NULL;
float * outData = NULL;
size_t mGlobal;
float data = 0.0f;
int i, j;
cl_int err = 0;
cl_mem * memList = NULL;
cl_kernel * kernelList = NULL;
clEnv * env = NULL;
inData = (float*)malloc(width * height * sizeof(float));
outData = (float*)malloc(width * height * sizeof(float));
for(i = 0; i < height; i ++){
for(j = 0; j < width; j ++){
inData[i * width + j] = data;
printf("%f ", inData[i * width + j]);
data += 1.0f;
}
printf("\n");
}
printf("***************************************************************************\n");
memList = getMemList();
kernelList = getKernelList();
env = getCLEnv();
err = clEnqueueWriteBuffer(env->commandQueue[0], memList[MEM_input], CL_TRUE, 0, width * height * sizeof(float), inData, 0, NULL, NULL);
if(CL_SUCCESS != err){
printf("bufferCopy clEnqueueWriteBuffer error. error code = %d\n", err);
return err;
}
err = clSetKernelArg(kernelList[KERNEL_buffercpy], 0, sizeof(cl_mem), &memList[MEM_input]);
err += clSetKernelArg(kernelList[KERNEL_buffercpy], 1, sizeof(cl_mem), &memList[MEM_output]);
if(CL_SUCCESS != err){
printf("bufferCopy clSetKernelArg error. error code = %d\n", err);
return err;
}
mGlobal = width * height;
err = clEnqueueNDRangeKernel(env->commandQueue[0], kernelList[KERNEL_buffercpy], 1, NULL, &mGlobal, NULL, 0, NULL, NULL);
if(CL_SUCCESS != err){
printf("bufferCopy clEnqueueNDRangeKernel error. error code = %d\n", err);
return err;
}
err = clEnqueueReadBuffer(env->commandQueue[0], memList[MEM_output], CL_TRUE, 0, width * height * sizeof(float), outData, 0, NULL, NULL);
if(CL_SUCCESS != err){
printf("bufferCopy clEnqueueMapBuffer error. error code = %d\n", err);
return err;
}
for(i = 0; i < height; i ++){
for(j = 0; j < width; j ++){
printf("%f ", outData[i * width + j]);
}
printf("\n");
}
return 0;
}
**6.**buffercpy.cl
/*
* name : buffercpy
* @param in : input
* @param out : output
*/
__kernel void buffercpy(__global float * in, __global float * out){
int id = get_global_id(0);
out[id] = in[id];
}
**7.**PrepareCLKernel.bat 改文件是把文件中kernel,转换成字符串
@echo off
echo 开始删除文件。。。
@del kernelSrc.c
@del InitKernel.h
echo 删除文件完成。。。
echo #ifndef _KernelSrc_C_ >> kernelSrc.c
echo #define _KernelSrc_C_ >> kernelSrc.c
echo // >> kernelSrc.c
echo //Auto generated file >> kernelSrc.c
echo // >> kernelSrc.c
echo static const char* kernelName[] = { >> kernelSrc.c
echo #ifndef _INITKERNEL_H_ >> InitKernel.h
echo #define _INITKERNEL_H_ >> InitKernel.h
echo // >> InitKernel.h
echo //Auto generated file >> InitKernel.h
echo // >> InitKernel.h
echo typedef enum _KERNEL_List_{ >> InitKernel.h
echo KERNEL_Start = -1, >> InitKernel.h
@setlocal enabledelayedexpansion
For /F "delims==" %%i in ('dir /b *.cl') do (
echo "%%~nsi", >> kernelSrc.c
echo KERNEL_%%~nsi, >> InitKernel.h
)
@setlocal disabledelayedexpansion
echo }; >> kernelSrc.c
echo KERNEL_Number >> InitKernel.h
echo }KERNEL_List; >> InitKernel.h
echo #ifdef __cplusplus >> InitKernel.h
echo extern "C" >> InitKernel.h
echo { >> InitKernel.h
echo #endif >> InitKernel.h
echo cl_kernel* getKernelList(); >> InitKernel.h
echo int InitKernelList(); >> InitKernel.h
echo int UnitKernelList(); >> InitKernel.h
echo #ifdef __cplusplus >> InitKernel.h
echo } >> InitKernel.h
echo #endif >> InitKernel.h
echo // >> InitKernel.h
echo //end >> InitKernel.h
echo // >> InitKernel.h
echo #endif >> InitKernel.h
echo static const char* kernelsrc = >> kernelSrc.c
@setlocal enabledelayedexpansion
set var=
For /F "delims==" %%i in ('dir /b *.cl') do (
For /F "delims=" %%l in ('type "%%~nsi.cl"') do (
echo "%%l\n"\>> kernelSrc.c
)
echo "\n"\>> kernelSrc.c
)
@setlocal disabledelayedexpansion
echo "";>> kernelSrc.c
echo // >> kernelSrc.c
echo //end >> kernelSrc.c
echo // >> kernelSrc.c
echo #endif >> kernelSrc.c
move kernelSrc.c ..\kernelInc\kernelSrc.c
move InitKernel.h ..\..\InitKernel.h
@echo Done
如有疑问可留言。