1 files changed, 225 insertions, 0 deletions
diff --git a/src/opencl_mesh_kit.cpp b/src/opencl_mesh_kit.cpp
new file mode 100644
index 0000000..a3eb6dc
--- /dev/null
+++ b/src/opencl_mesh_kit.cpp
@@ -0,0 +1,225 @@
+#include "opencl_mesh_kit.hpp"
+
+// TODO : print streamsdk::getOpenCLErrorCodeStr(res)
+#define CL_RETURN_VAL_IF_FAIL(val, expr) do { \
+	cl_int res=(expr); \
+	if ( res != CL_SUCCESS ) {	\
+		std::cerr << "file " << __FILE__ << ": line " << __LINE__ << " (" << __PRETTY_FUNCTION__ \
+			<< "): '" << "expr" << "' failed (return code : " << res << ")" << std::endl; \
+		return val; \
+	} \
+} while(0)
+
+cl_int OpenCLMeshKit::initCL(intptr_t gl_display, intptr_t gl_context, intptr_t gl_vbo, size_t meshWidth, size_t meshHeight, size_t groupSize) {
+	cl_uint id, numPlatforms;
+	cl_int res;
+	char pbuf[100];
+	cl_platform_id *platforms, platform;
+	bool usableDeviceFound=false;
+
+	this->meshWidth = meshWidth;
+	this->meshHeight = meshHeight;
+	this->groupSize = groupSize;
+
+	// Get platform count
+	CL_RETURN_VAL_IF_FAIL(10,
+		clGetPlatformIDs(0, NULL, &numPlatforms)
+	);
+	
+	std::cout << "Detected " << numPlatforms << " platform(s)" << std::endl;
+	if ( ! ( numPlatforms > 0 ) ) return 2;
+
+	// Allocate room for all platform IDs
+	platforms = new cl_platform_id[numPlatforms];
+
+	// Get platform IDs
+	CL_RETURN_VAL_IF_FAIL(11,
+		clGetPlatformIDs(numPlatforms, platforms, &numPlatforms)
+	);
+
+	// Enumerate platforms and grab informations
+	for(id=0;id<numPlatforms;id++) {
+		platform=platforms[id];
+
+		CL_RETURN_VAL_IF_FAIL(12,
+			clGetPlatformInfo(platform, CL_PLATFORM_VENDOR, sizeof(pbuf), pbuf, NULL)
+		);
+		std::cout << "Platform " << id << " : " << pbuf << std::endl;
+
+		// Dynamically get the function pointer for clGetGLConetextInfoKHR
+		clGetGLContextInfoKHR_fn clGetGLContextInfoKHR_proc = (clGetGLContextInfoKHR_fn) clGetExtensionFunctionAddressForPlatform(platform, "clGetGLContextInfoKHR");
+		if (!clGetGLContextInfoKHR_proc) {
+			std::cerr << "clGetExtensionFunctionAddressForPlatform(platform, clGetGLContextInfoKHR) failed" << std::endl;
+			continue;
+		}
+
+		// Try to get the device corresponding to the GL context/display on this platform
+		cl_context_properties cpsGL[] = {
+			CL_CONTEXT_PLATFORM, (cl_context_properties)platform,
+			CL_GLX_DISPLAY_KHR, gl_display,
+			CL_GL_CONTEXT_KHR, gl_context,
+			0
+		};
+
+		std::cout << "cl_context_properties cpsGL :" << std::endl;
+		std::cout << "\tCL_CONTEXT_PLATFORM :" << (void *)cpsGL[1] << std::endl;
+		std::cout << "\tCL_GLX_DISPLAY_KHR :"  << (void *)cpsGL[3] << std::endl;
+		std::cout << "\tCL_GL_CONTEXT_KHR :"   << (void *)cpsGL[5] << std::endl;
+
+    size_t deviceSize=0;
+    // get deviceSize (should be 1*sizeof(cl_device_id) with CL_CURRENT_DEVICE_FOR_GL_CONTEXT_KHR)
+    res=clGetGLContextInfoKHR_proc(cpsGL,CL_CURRENT_DEVICE_FOR_GL_CONTEXT_KHR,0,NULL,&deviceSize);
+    if ( res!=CL_SUCCESS || deviceSize!=1*sizeof(cl_device_id)) {
+      std::cerr << "clGetGLContextInfoKHR_proc(cpsGL,CL_CURRENT_DEVICE_FOR_GL_CONTEXT_KHR,0,...) failed" << std::endl;
+			std::cerr << " (return code : " << res << ")" << std::endl;
+      continue;
+    }
+
+		cl_dev=0;
+		res=clGetGLContextInfoKHR_proc(cpsGL,CL_CURRENT_DEVICE_FOR_GL_CONTEXT_KHR,deviceSize,&cl_dev,NULL);
+		if ( res!=CL_SUCCESS || cl_dev==0 ) {
+			std::cerr << "clGetGLContextInfoKHR_proc(cpsGL,CL_CURRENT_DEVICE_FOR_GL_CONTEXT_KHR," << deviceSize << ",...) failed" << std::endl;
+			std::cerr << " (return code : " << res << ")" << std::endl;
+			continue;
+		}
+
+		std::cout << "cl_device :" << (void *)cl_dev << std::endl;
+
+		cl_ctx = clCreateContext(cpsGL,1,&cl_dev,0,0,&res);
+		if ( res!=CL_SUCCESS ) {
+			std::cerr << "clCreateContext() failed" << std::endl; 
+			std::cerr << " (return code : " << res << ")" << std::endl;
+			continue;
+		}
+
+		cl_cq = clCreateCommandQueue(cl_ctx,cl_dev,0,&res);
+		if ( res!=CL_SUCCESS ) {
+			std::cerr << "clCreateCommandQueue() failed" << std::endl; 
+			std::cerr << " (return code : " << res << ")" << std::endl;
+			continue;
+		}
+
+		usableDeviceFound=true;
+		break;
+	}
+
+	if (! usableDeviceFound) {
+		std::cerr << "No OpenCL device has been successfully initialized" << std::endl;
+		return 13;
+	}
+
+	cl_vbo = clCreateFromGLBuffer(cl_ctx, CL_MEM_WRITE_ONLY, gl_vbo, &res);
+  if ( res!=CL_SUCCESS ) {
+		std::cerr << "clCreateFromGLBuffer(..., gl_vbo, &res) failed" << std::endl;
+		return 14;
+	}
+
+	std::cout << "OpenCL initialization done." << std::endl;
+	return 0;
+	
+}
+
+cl_int OpenCLMeshKit::compileKernels(std::list<std::string> names, const char source[], size_t sourceLen) {
+	cl_int res=0;
+
+	const char *p_source=source;
+
+	cl_program program = clCreateProgramWithSource(cl_ctx, 1, &p_source, &sourceLen,&res);
+	if ( res!=CL_SUCCESS ) {
+		std::cerr << "Failed to clCreateProgramWithSource(<source of zero_z kernel>)" << std::endl;
+		return 21;
+	}
+
+	res = clBuildProgram(program, 1, &cl_dev, "", NULL, NULL);
+  if ( res!=CL_SUCCESS ) {
+    std::cerr << "Failed to clBuildProgram()"  << std::endl;
+		return 22;
+	}
+
+	for (std::list<std::string>::iterator ii = names.begin(); ii != names.end(); ++ii) {
+		std::string kName = (*ii);
+		char *kNameZTS = new char[kName.length()+1];
+		std::strcpy(kNameZTS, kName.c_str());
+
+		cl_kernel kernel = clCreateKernel(program,kNameZTS,&res);
+		delete [] kNameZTS;
+
+		if ( res!=CL_SUCCESS ) {
+			std::cerr << "Failed to clCreateKernel(program,\"" << kName << "\",&res);" << std::endl;
+			return 23;
+		}
+		kernels[kName]=kernel;
+	}
+
+	return 0;
+}
+
+cl_int OpenCLMeshKit::execKernel(std::string kernelName) {
+	/*TODO
+
+cl_int execKernel(cl_context cl_ctx, cl_command_queue commandQueue, cl_kernel kernel, size_t mesh_width, size_t mesh_height, size_t group_size, cl_mem cl_vbo, float time) {
+
+  cl_int res;
+  cl_event  eventND[1];
+
+  // Set local and global work group sizes
+  size_t globalWorkSize[2], localWorkSize[2];
+  globalWorkSize[0]=mesh_width;
+  globalWorkSize[1]=mesh_height;
+  localWorkSize[0]=group_size;
+  localWorkSize[1]=1;
+
+  res=clEnqueueAcquireGLObjects(commandQueue, 1, &cl_vbo, 0, 0, NULL);
+
+  res=clSetKernelArg(kernel, 0, sizeof(cl_mem), &cl_vbo); // float4 *pos
+  res=clSetKernelArg(kernel, 1, sizeof(cl_uint), (void *)&mesh_width);
+  res=clSetKernelArg(kernel, 2, sizeof(cl_uint), (void *)&mesh_height);
+  res=clSetKernelArg(kernel, 3, sizeof(float), &time);
+
+  // Execute kernel on given device
+  res=clEnqueueNDRangeKernel(commandQueue, kernel, 2, NULL, globalWorkSize, localWorkSize, 0, NULL, eventND);
+  if ( res != CL_SUCCESS ) {
+    fputs("Failed to clEnqueueNDRangeKernel()\n", stderr);
+    return 1;
+  }
+
+  res=clFlush(commandQueue);
+  res=clWaitForEvents(1,eventND); //XXX: SimpleGL utilise une attente active, pourquoi ?
+  res=clReleaseEvent(eventND[0]);
+  res=clEnqueueReleaseGLObjects(commandQueue, 1, &cl_vbo, 0, 0, 0);
+  res=clFinish(commandQueue);
+
+  return CL_SUCCESS;
+}
+*/
+	return 0;
+}
+
+void OpenCLMeshKit::releaseKernels() {
+	for (std::map<std::string,cl_kernel>::iterator ii = kernels.begin(); ii != kernels.end(); ++ii ) {
+		clReleaseKernel((*ii).second);
+	}
+	kernels.clear();
+}
+
+cl_int OpenCLMeshKit::resetVBO() {
+	cl_int res;
+	std::map<std::string, cl_kernel> user_kernels=kernels;
+
+	std::list<std::string> n; n.push_back("zero_z");
+	res = compileKernels(n, kernel_src_zero_z, sizeof(kernel_src_zero_z));
+
+	if(res==0) res = execKernel("zero_z");	
+	
+	releaseKernels();
+	kernels=user_kernels;
+
+	return res;
+}
+
+void OpenCLMeshKit::setGroupSize(size_t groupSize) {
+	this->groupSize=groupSize;
+}
+
+OpenCLMeshKit::~OpenCLMeshKit() { }
+