Skip to content

Commit

Permalink
Merge branch 'windows'
Browse files Browse the repository at this point in the history
  • Loading branch information
@hughperkins
hughperkins committed Feb 1, 2015
2 parents 6c89c5e + 5110563 commit 2f184f5
Show file tree
Hide file tree
Showing 23 changed files with 652 additions and 381 deletions.
176 changes: 176 additions & 0 deletions CLKernel.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -12,6 +12,33 @@ using namespace std;
#include "CLArrayInt.h"
#include "CLArray.h"

#include "OpenCLHelper_export.h"

template<typename T>
std::string CLKernel::toString(T val) {
std::ostringstream myostringstream;
myostringstream << val;
return myostringstream.str();
}

CLKernel::CLKernel(OpenCLHelper *openclhelper, cl_program program, cl_kernel kernel) {
this->openclhelper = openclhelper;
nextArg = 0;
error = CL_SUCCESS;
this->program = program;
this->kernel = kernel;
}
CLKernel::CLKernel(const CLKernel &kernel) {
throw std::runtime_error("can't assign CLKernel");
}
CLKernel &CLKernel::operator=(const CLKernel &kernel) {
throw std::runtime_error("can't assign CLKernel");
}
CLKernel::~CLKernel() {
clReleaseProgram(program);
clReleaseKernel(kernel);
}

CLKernel *CLKernel::input( CLArray *clarray1d ) {
assert( clarray1d != 0 );
if( !clarray1d->isOnDevice() ) {
Expand Down Expand Up @@ -92,3 +119,152 @@ CLKernel *CLKernel::output( CLWrapper *wrapper ) {
return this;
}

CLKernel *CLKernel::localFloats(int count) {
error = clSetKernelArg(kernel, nextArg, count * sizeof(cl_float), 0);
openclhelper->checkError(error);
nextArg++;
return this;
}
CLKernel *CLKernel::localInts(int count) {
error = clSetKernelArg(kernel, nextArg, count * sizeof(cl_int), 0);
openclhelper->checkError(error);
nextArg++;
return this;
}
CLKernel *CLKernel::local(int N) {
return localFloats(N);
}

template<typename T> CLKernel *CLKernel::input(int N, const T *data) {
cl_mem buffer = clCreateBuffer(*(openclhelper->context), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(T) * N, (void *)data, &error);
openclhelper->checkError(error);
error = clSetKernelArg(kernel, nextArg, sizeof(cl_mem), &buffer);
openclhelper->checkError(error);
buffers.push_back(buffer);
nextArg++;
return this;
}
template<typename T>
CLKernel *CLKernel::in(int N, const T *data) {
return input(N, data);
}
CLKernel *CLKernel::input(int value) {
inputArgInts.push_back(value);
error = clSetKernelArg(kernel, nextArg, sizeof(int), &(inputArgInts[inputArgInts.size() - 1]));
openclhelper->checkError(error);
nextArg++;
return this;
}
CLKernel *CLKernel::in(int value) {
return input(value);
}
CLKernel *CLKernel::input(float value) {
inputArgFloats.push_back(value);
error = clSetKernelArg(kernel, nextArg, sizeof(float), &(inputArgFloats[inputArgFloats.size() - 1]));
openclhelper->checkError(error);
nextArg++;
return this;
}
CLKernel *CLKernel::in(float value) {
return input(value);
}
template<typename T>
CLKernel *CLKernel::output(int N, T *data) {
cl_mem buffer = clCreateBuffer(*(openclhelper->context), CL_MEM_WRITE_ONLY, sizeof(T) * N, 0, &error);
openclhelper->checkError(error);
error = clSetKernelArg(kernel, nextArg, sizeof(cl_mem), &buffer);
buffers.push_back(buffer);
//outputArgNums.push_back(nextArg);
outputArgBuffers.push_back(buffer);
outputArgPointers.push_back( (void *)data);
outputArgSizes.push_back(sizeof(T) * N);
nextArg++;
return this;
}
template<typename T>
CLKernel *CLKernel::out(int N, T *data) {
return output(N, data);
}
template<typename T>
CLKernel *CLKernel::inout(int N, T *data) {
cl_mem buffer = clCreateBuffer(*(openclhelper->context), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(T) * N, (void *)data, &error);
openclhelper->checkError(error);
error = clSetKernelArg(kernel, nextArg, sizeof(cl_mem), &buffer);
openclhelper->checkError(error);
buffers.push_back(buffer);
outputArgBuffers.push_back(buffer);
outputArgPointers.push_back( (void *)( data ) );
outputArgSizes.push_back(sizeof(T) * N);
nextArg++;
return this;
}
void CLKernel::run_1d(int global_worksize, int local_worksize) {
size_t global_ws = global_worksize;
size_t local_ws = local_worksize;
run(1, &global_ws, &local_ws);
}

void CLKernel::run(int ND, const size_t *global_ws, const size_t *local_ws) {
//cout << "running kernel" << std::endl;
error = clEnqueueNDRangeKernel(*(openclhelper->queue), kernel, ND, NULL, global_ws, local_ws, 0, NULL, NULL);
switch (error) {
case 0:
break;
case -4:
throw std::runtime_error("Memory object allocation failure, code -4");
break;
case -5:
throw std::runtime_error("Out of resources, code -5");
break;
case -11:
throw std::runtime_error("Program build failure, code -11");
break;
case -46:
throw std::runtime_error("Invalid kernel name, code -46");
break;
case -52:
throw std::runtime_error("Invalid kernel args, code -52");
break;
case -54:
throw std::runtime_error("Invalid work group size, code -54");
break;
default:
throw std::runtime_error("Something went wrong, code " + toString(error));
}
openclhelper->checkError(error);
// error = clFinish(openclhelper->queue);
// openclhelper->checkError(error);
//}

//void retrieveresultsandcleanup() {
for (int i = 0; i < (int)outputArgBuffers.size(); i++) {
clEnqueueReadBuffer(*(openclhelper->queue), outputArgBuffers[i], CL_TRUE, 0, outputArgSizes[i], outputArgPointers[i], 0, NULL, NULL);
}
// std::cout << "done" << std::endl;

for (int i = 0; i < (int)buffers.size(); i++) {
clReleaseMemObject(buffers[i]);
}
buffers.clear();
outputArgBuffers.clear();
outputArgPointers.clear();
outputArgSizes.clear();
inputArgInts.clear();
inputArgFloats.clear();
nextArg = 0;
}

// template class std::vector<cl_mem>;


template OpenCLHelper_EXPORT CLKernel *CLKernel::input(int N, const float *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::input(int N, const int *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::in(int N, const float *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::in(int N, const int *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::output(int N, float *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::output(int N, int *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::out(int N, float *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::out(int N, int *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::inout(int N, float *data);
template OpenCLHelper_EXPORT CLKernel *CLKernel::inout(int N, int *data);

Loading

0 comments on commit 2f184f5

Please sign in to comment.