From fc4daa1fe40d0127dbd1e9fdbd4031bade1f7522 Mon Sep 17 00:00:00 2001 From: Ludovic Pouzenc Date: Mon, 4 Mar 2013 22:38:40 +0000 Subject: Ok, première version qui affiche le mesh correctement initializé avec OpenCL et qui affiche même un sinus radial sur un clic de souris mouse1. MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Reste a capturer les evenements mouse wheel et/ou clavier. git-svn-id: file:///var/svn/2013-gpudataviz/trunk@19 371a6b4a-a258-45f8-9dcc-bdd82ce0ac9d --- src/boring_parts.cpp | 57 ------------------ src/boring_parts.hpp | 24 -------- src/compil.sh | 1 - src/gpudataviz.cpp | 129 ++++++++++++++++++++++++++--------------- src/gtk_includes.hpp | 19 ------ src/gtk_win_main.hpp | 1 - src/my_gtk_gl_scene_widget.hpp | 4 +- src/opencl_mesh_kit.cpp | 61 +++++++++++-------- src/opencl_mesh_kit.hpp | 11 +++- 9 files changed, 132 insertions(+), 175 deletions(-) delete mode 100644 src/boring_parts.cpp delete mode 100644 src/boring_parts.hpp delete mode 100644 src/gtk_includes.hpp diff --git a/src/boring_parts.cpp b/src/boring_parts.cpp deleted file mode 100644 index 447ae60..0000000 --- a/src/boring_parts.cpp +++ /dev/null @@ -1,57 +0,0 @@ -#include "boring_parts.hpp" - -// TODO : only need OpenGL things, not GTK ones for now -//#include "gtk_includes.hpp" - -/* From http://stackoverflow.com/questions/4317062/opengl-how-to-check-if-the-user-supports-glgenbuffers -#ifdef WIN32 - #include - #define glGetProcAddress(a) wglGetProcAddress(a) -#endif -#ifdef X11 - #define glGetProcAddress(a) glXGetProcAddress ( \ - reinterpret_cast(a) \ - ) -#endif - -#ifndef GetExtension - #define GetExtension(Type, ExtenName) \ - ExtenName = (Type) \ - glGetProcAddress(STRINGIFY(ExtenName)); \ - if(!ExtenName) \ - { \ - std:cout << "Your Computer Does Not " \ - << "Support GL Extension: " \ - << STRINGIFY(ExtenName) \ - << std::endl; \ - exit(1); \ - } \ - else \ - { \ - std::cout << "Loaded Extension: " \ - << STRINGIFY(ExtenName) \ - << std::endl; \ - } -#endif -*/ - -bool updateGLProjectionMatrix(Glib::RefPtr glCtx, Glib::RefPtr glWin, int width, int height) { - - GLdouble aspect = (GLdouble) width/height; - - // *** OpenGL BEGIN *** - if (!glWin->gl_begin(glCtx)) return false; - - glViewport(0, 0, width, height); - glMatrixMode(GL_PROJECTION); - glLoadIdentity(); - gluPerspective(60.0, aspect, 0.1, 10.0); - glMatrixMode(GL_MODELVIEW); - - glWin->gl_end(); - // *** OpenGL END *** - - return true; - -} - diff --git a/src/boring_parts.hpp b/src/boring_parts.hpp deleted file mode 100644 index 09404d8..0000000 --- a/src/boring_parts.hpp +++ /dev/null @@ -1,24 +0,0 @@ -#include - -#include /* Must included before X11/Xlib.h */ -//#include /* Seems included by GL/glxew.h */ -//#include /* For GLXContext, must appear before gl.h, includes X11/Xlib.h */ -#include /* Includes a GL/gl.h (the right one ??) */ - -//#include -#include -//#include - -#ifdef HAS_OPENCL -// Blabla -#else -// Quick and dirty cl_float4 replacement -typedef union { - float s[4]; - struct{ float x, y, z, w; }; - struct{ float s0, s1, s2, s3; }; -} cl_float4; -#endif /*HAS_OPENCL*/ - -bool updateGLProjectionMatrix(Glib::RefPtr glCtx, Glib::RefPtr glWin, int width, int height); - diff --git a/src/compil.sh b/src/compil.sh index 1f5db4c..b0a1bad 100755 --- a/src/compil.sh +++ b/src/compil.sh @@ -47,7 +47,6 @@ rm -v $BUILD_PATH/* || true build_cxx opencl_mesh_kit.o opencl_mesh_kit.cpp build_cxx gpudataviz.o gpudataviz.cpp -build_cxx boring_parts.o boring_parts.cpp build_cxx gtk_win_main.o gtk_win_main.cpp link_cxx gpudataviz $BUILD_PATH/*.o diff --git a/src/gpudataviz.cpp b/src/gpudataviz.cpp index 72ee472..2b2386a 100644 --- a/src/gpudataviz.cpp +++ b/src/gpudataviz.cpp @@ -1,12 +1,7 @@ #include - #include -//#include "gtk_includes.h" #include "gtk_win_main.hpp" -//#include "my_gtk_gl_scene_widget.hpp" -#include "boring_parts.hpp" - #include // X11 specific // Macro to make things readable in main() function @@ -59,7 +54,7 @@ MyGTKGLSceneWidget::MyGTKGLSceneWidget(Glib::RefPtr &glconfig) set_gl_capability(glconfig); Gdk::EventMask mask = Gdk::POINTER_MOTION_MASK | Gdk::BUTTON_MOTION_MASK | Gdk::BUTTON_PRESS_MASK | Gdk::BUTTON_RELEASE_MASK; set_events(mask); // The containing window should have those attributes too - this->camera.rx = 0.0f; this->camera.ry = 0.0f; this->camera.tz = -3.0f; + this->camera.rx = -64.0f; this->camera.ry = -16.0f; this->camera.tz = -1.0f; } MyGTKGLSceneWidget::~MyGTKGLSceneWidget() { } @@ -97,49 +92,54 @@ void MyGTKGLSceneWidget::on_realize() { glBufferData(GL_ARRAY_BUFFER, gl_vbo_data_size, NULL, GL_DYNAMIC_DRAW); gl_res=glGetError(); if ( gl_res != GL_NO_ERROR ) { - std::cerr << "glBufferData(). Unable to allocate " << gl_vbo_data_size << "bytes in VRAM" << std::endl; + std::cerr << "glBufferData(). Unable to allocate " << gl_vbo_data_size << " bytes in VRAM" << std::endl; std::cerr << gluErrorString(gl_res); EXIT_IF_FAIL(5, false); } -//#ifdef HAS_OPENCL -// #ifdef X11 +//#ifdef X11 intptr_t gl_context = (intptr_t)glXGetCurrentContext(); intptr_t gl_display = (intptr_t)glXGetCurrentDisplay(); - - int cl_res = clKit.initCL(gl_display, gl_context, gl_vbo, mesh_width, mesh_height, group_size); - EXIT_IF_FAIL(cl_res, cl_res==0); - -// std::cerr << "DEBUG : begin initOpenCL()" << std::endl; -// int cl_res = initOpenCL(gl_display, gl_context, gl_vbo); /* See boring_parts.cpp */ -// EXIT_IF_FAIL(cl_res, cl_res==0); -// #else -// #error initOpenCL works only for X11 systems for now -// #endif -// } +//#else +// #error initOpenCL works only for X11 systems for now //#endif + int cl_res = this->clKit.initCL(gl_display, gl_context, gl_vbo, mesh_width, mesh_height, group_size); + EXIT_IF_FAIL(cl_res, cl_res==0); + this->clKit.resetVBO(); // XXX Just for displaying a flat mesh at start + + const char source[]=STRINGIFY( + /* This is OpenCL kernel code (Syntax like C but it's a different language) */ + __kernel void water1(__global float4 *pos, unsigned int width, unsigned int height, float time) { + unsigned int nx = get_global_id(0); + unsigned int ny = get_global_id(1); + /* calculate uv coordinates of the mesh point [0.0;1.0] */ + float u = nx / (float) width; + float v = ny / (float) height; + /* calculate centered coordinates [-1.0;1.0] */ + float x = u*2.0-1.0; + float y = v*2.0-1.0; + /* We only use normalized quaterinons here */ + float w = 1.0; + /* Calculate the desirated value of the mesh point */ + float freq = 8.0 * 3.14; + float amp = 1.0 / 10.0; /* 0.1 does NOT works ! WTF !!! */ + float speed = 1.0; + float dist = sqrt(x*x+y*y); + float z = amp * sin( freq * dist - speed * time ) / dist ; + + /* Write output vertex (centered) */ + pos[ny*width+nx] = (float4)(x, y, z, w); + } + ); + + std::list knames; + knames.push_back("water1"); + cl_res = this->clKit.compileKernels(knames, source, sizeof(source)); + + EXIT_IF_FAIL(6, cl_res==0); + knames.clear(); - // Programmatically create rendering lists : opengl will able to replay that efficiently - GLUquadricObj* qobj = gluNewQuadric(); - gluQuadricDrawStyle(qobj, GLU_FILL); - glNewList(1, GL_COMPILE); - //gluSphere(qobj, 1.0, 20, 20); - gluSphere(qobj, 1.0, 5, 5); - glEndList(); - - // Setup scene envrionnement - static GLfloat light_diffuse[] = {1.0, 0.0, 0.0, 1.0}; - static GLfloat light_position[] = {1.0, 1.0, 1.0, 0.0}; - glLightfv(GL_LIGHT0, GL_DIFFUSE, light_diffuse); - glLightfv(GL_LIGHT0, GL_POSITION, light_position); - glEnable(GL_LIGHTING); - glEnable(GL_LIGHT0); - glEnable(GL_DEPTH_TEST); - - glClearColor(1.0, 1.0, 1.0, 1.0); - glClearDepth(1.0); - // Projection setup is done at on_configure_event // Camera setup (ie initial MODELVIEW matrix) is done at on_expose_event @@ -149,8 +149,28 @@ void MyGTKGLSceneWidget::on_realize() { bool MyGTKGLSceneWidget::on_configure_event(GdkEventConfigure* event) { CALL_TRACE ; // This one runs mainly when GTK GL Widget is resized - // See boring_parts.cpp. In short : gluPerspective(60.0, aspect, 0.1, 10.0); - return updateGLProjectionMatrix(get_gl_context(), get_gl_window(), get_width(), get_height()); + + float h=this->get_height(); + float w=this->get_width(); + Glib::RefPtr glwindow = get_gl_window(); + + // *** OpenGL BEGIN *** + + //FIXME could segfault if get_gl_window() has failed + if (!glwindow->gl_begin(get_gl_context())) { + std::cerr << "Oups : glwindow->gl_begin(get_gl_context())" << std::endl; + return false; + } + + glViewport(0, 0, w, h); + glMatrixMode(GL_PROJECTION); + glLoadIdentity(); + gluPerspective(60.0, w/h, 0.1, 10.0); + + glwindow->gl_end(); + // *** OpenGL END *** + + return true; } bool MyGTKGLSceneWidget::on_expose_event(GdkEventExpose* event) { @@ -169,12 +189,18 @@ bool MyGTKGLSceneWidget::on_expose_event(GdkEventExpose* event) { glLoadIdentity(); glTranslatef(0.0, 0.0, this->camera.tz); glRotatef(this->camera.rx, 1.0, 0.0, 0.0); - glRotatef(this->camera.ry, 0.0, 1.0, 0.0); + glRotatef(this->camera.ry, 0.0, 0.0, 1.0); - // Drawing all the stuff glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT); - glCallList(1); + // Drawing all the stuff + glColor4f(0.40f,0.78f,0.97f,1.0f); + glBindBuffer(GL_ARRAY_BUFFER, this->clKit.getGLVBO()); + glVertexPointer(4, GL_FLOAT, 0, (GLvoid *) 0); + glEnableClientState(GL_VERTEX_ARRAY); + glDrawArrays(GL_POINTS, 0, this->clKit.getMeshItemCount()); + glDisableClientState(GL_COLOR_ARRAY); + glBindBuffer(GL_ARRAY_BUFFER, 0); glwindow->gl_end(); // *** OpenGL END *** @@ -243,11 +269,12 @@ bool MyGTKGLSceneWidget::do_mouse_logic(GdkEventType type, guint state, guint x, bool redraw=false; // Setting it to true will queue a redraw to the widget (invalidate) -// std::cout << "event type " << type << " state " << state << " on (" << x << "," << y << ") " << std::endl; + std::cout << "event type " << type << " state " << state << " on (" << x << "," << y << ") " << std::endl; /* *** BEGIN event filtering *** */ MOUSE_DRAG_START(GDK_BUTTON2_MASK) { drag_x=x; drag_y=y; + //std::cout << "this->camera == {" << this->camera.rx << ", " << this->camera.ry << ", " << this->camera.tz << "}" << std::endl; } MOUSE_DRAGING(GDK_BUTTON2_MASK) { @@ -262,8 +289,18 @@ bool MyGTKGLSceneWidget::do_mouse_logic(GdkEventType type, guint state, guint x, redraw=true; } + MOUSE_CLIC(GDK_BUTTON4_MASK, 0, 0) { + this->camera.tz += 1.0f; + } + + MOUSE_CLIC(GDK_BUTTON5_MASK, 0, 0) { + this->camera.tz -= 1.0f; + } + MOUSE_CLIC(GDK_BUTTON1_MASK, 0, 0) { //TODO + this->clKit.execKernel("water1", 0.0f); + redraw=true; } diff --git a/src/gtk_includes.hpp b/src/gtk_includes.hpp deleted file mode 100644 index 5a60f04..0000000 --- a/src/gtk_includes.hpp +++ /dev/null @@ -1,19 +0,0 @@ -#ifndef GTK_INCLUDES_H -#define GTK_INCLUDES_H - - -#include -//#include "gl_core_3_1.hpp" -#include -//#include "gl_core_1_5_vbo.hpp" - -#ifdef G_OS_WIN32 -#define WIN32_LEAN_AND_MEAN 1 -#include -#endif - -//#include -#include - -#endif /* GTK_INCLUDES_H */ - diff --git a/src/gtk_win_main.hpp b/src/gtk_win_main.hpp index 2a42beb..9833fb0 100644 --- a/src/gtk_win_main.hpp +++ b/src/gtk_win_main.hpp @@ -1,4 +1,3 @@ -#include "gtk_includes.hpp" #include "my_gtk_gl_scene_widget.hpp" #define WIN_MAIN_TITLE "GPU Data Viz v0.1" diff --git a/src/my_gtk_gl_scene_widget.hpp b/src/my_gtk_gl_scene_widget.hpp index b6fbe73..4d50dd6 100644 --- a/src/my_gtk_gl_scene_widget.hpp +++ b/src/my_gtk_gl_scene_widget.hpp @@ -1,7 +1,9 @@ #ifndef MY_GTK_GL_SCENE_H #define MY_GTK_GL_SCENE_H -#include "gtk_includes.hpp" +#include +#include +#include #include "opencl_mesh_kit.hpp" // Class that will contain all the OpenGL logic for displaying the OpenCL computed data diff --git a/src/opencl_mesh_kit.cpp b/src/opencl_mesh_kit.cpp index a3eb6dc..a9010e8 100644 --- a/src/opencl_mesh_kit.cpp +++ b/src/opencl_mesh_kit.cpp @@ -20,6 +20,7 @@ cl_int OpenCLMeshKit::initCL(intptr_t gl_display, intptr_t gl_context, intptr_t this->meshWidth = meshWidth; this->meshHeight = meshHeight; this->groupSize = groupSize; + this->gl_vbo = gl_vbo; // Get platform count CL_RETURN_VAL_IF_FAIL(10, @@ -154,46 +155,52 @@ cl_int OpenCLMeshKit::compileKernels(std::list names, const char so 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 OpenCLMeshKit::execKernel(std::string kernelName, float time) { cl_int res; cl_event eventND[1]; + size_t globalWorkSize[2], localWorkSize[2]; + cl_kernel kernel; + + std::map::iterator ii=this->kernels.find(kernelName); + if ( ii==this->kernels.end() ) { + std::cerr << "execKernel(\"" << kernelName << "\", " << time << ") failed : no kernel found with this name" << std::endl; + return -1; + } + + kernel=this->kernels[kernelName]; // 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; + globalWorkSize[0]=this->meshWidth; + globalWorkSize[1]=this->meshHeight; + localWorkSize[0]=this->groupSize; localWorkSize[1]=1; - res=clEnqueueAcquireGLObjects(commandQueue, 1, &cl_vbo, 0, 0, NULL); + res=clEnqueueAcquireGLObjects(this->cl_cq, 1, &(this->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); + res=clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&(this->cl_vbo)); // float4 *pos + res=clSetKernelArg(kernel, 1, sizeof(cl_uint), (void *)&(this->meshWidth)); + res=clSetKernelArg(kernel, 2, sizeof(cl_uint), (void *)&(this->meshHeight)); + res=clSetKernelArg(kernel, 3, sizeof(float), (void *)&time); // Execute kernel on given device - res=clEnqueueNDRangeKernel(commandQueue, kernel, 2, NULL, globalWorkSize, localWorkSize, 0, NULL, eventND); + res=clEnqueueNDRangeKernel(this->cl_cq, kernel, 2, NULL, globalWorkSize, localWorkSize, 0, NULL, eventND); if ( res != CL_SUCCESS ) { - fputs("Failed to clEnqueueNDRangeKernel()\n", stderr); + std::cerr << "clEnqueueNDRangeKernel() failed" << std::endl; return 1; } - res=clFlush(commandQueue); + //TODO : return values checking + res=clFlush(this->cl_cq); + + // (CPU) Wait until GPU kernel execution end 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); + res=clEnqueueReleaseGLObjects(this->cl_cq, 1, &(this->cl_vbo), 0, 0, 0); + res=clFinish(this->cl_cq); return CL_SUCCESS; } -*/ - return 0; -} void OpenCLMeshKit::releaseKernels() { for (std::map::iterator ii = kernels.begin(); ii != kernels.end(); ++ii ) { @@ -209,7 +216,7 @@ cl_int OpenCLMeshKit::resetVBO() { std::list 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"); + if(res==0) res = execKernel("zero_z", 0.0f); releaseKernels(); kernels=user_kernels; @@ -217,9 +224,13 @@ cl_int OpenCLMeshKit::resetVBO() { return res; } -void OpenCLMeshKit::setGroupSize(size_t groupSize) { - this->groupSize=groupSize; -} +size_t OpenCLMeshKit::getMeshWidth() { return this->meshWidth; } +size_t OpenCLMeshKit::getMeshHeight() { return this->meshHeight; } +size_t OpenCLMeshKit::getMeshItemCount() { return this->meshWidth * this->meshHeight; } +size_t OpenCLMeshKit::getGroupSize() { return this->groupSize; } +intptr_t OpenCLMeshKit::getGLVBO() { return this->gl_vbo; } + +void OpenCLMeshKit::setGroupSize(size_t groupSize) { this->groupSize=groupSize; } OpenCLMeshKit::~OpenCLMeshKit() { } diff --git a/src/opencl_mesh_kit.hpp b/src/opencl_mesh_kit.hpp index aa3a013..f99c6c2 100644 --- a/src/opencl_mesh_kit.hpp +++ b/src/opencl_mesh_kit.hpp @@ -14,8 +14,14 @@ class OpenCLMeshKit //RAII is violated but it is really triky to do differently cl_int initCL(intptr_t gl_display, intptr_t gl_context, intptr_t gl_vbo, size_t meshWidth, size_t meshHeight, size_t groupSize); cl_int compileKernels(std::list names, const char source[], size_t sourceLen); - cl_int execKernel(std::string kernelName); + cl_int execKernel(std::string kernelName, float time); void releaseKernels(); + + size_t getMeshWidth(); + size_t getMeshHeight(); + size_t getMeshItemCount(); + size_t getGroupSize(); + intptr_t getGLVBO(); void setGroupSize(size_t groupSize); // Quick and dirty function to initialize a test mesh @@ -23,6 +29,7 @@ class OpenCLMeshKit virtual ~OpenCLMeshKit(); + protected: size_t meshWidth; size_t meshHeight; @@ -33,6 +40,8 @@ class OpenCLMeshKit cl_command_queue cl_cq; cl_mem cl_vbo; + intptr_t gl_vbo; // Save this pointer for convinence (for data display code) + std::map kernels; }; -- cgit v1.2.3