Archives Discussions

[nou]

i try OGL OCL interoperability but i get some errors.

i create OGL context.

create CL context.

create VBO buffer and fill with some data

when i try get CL object from GL buffer i get error -34 from which mean CL_INVALID_CONTEXT when i use GPU.

when i try run code with CPU device i succesfully get CL object and get correct size. but it segafult when i set it as kernel argument and run it. well it segfault only if i try read or write into GL object.

gl_context = glXGetCurrentContext();//get gl context display = glXGetCurrentDisplay();//geting display handle cl_context_properties cl_properties[] = {CL_GL_CONTEXT_KHR, (cl_context_properties)gl_context, CL_GLX_DISPLAY_KHR, (cl_context_properties)display, CL_CONTEXT_PLATFORM, (cl_context_properties)platform, 0};//pass it to properties cl_c = clCreateContextFromType(cl_properties, CL_DEVICE_TYPE_GPU, NULL, NULL, &err_code);//create context glGenBuffers(1, &vbo);//create vbo buffer and fill it with some data glBindBuffer(GL_ARRAY_BUFFER, vbo); glBufferData(GL_ARRAY_BUFFER, sizeof(float)*grid_size*grid_size*4, grid, GL_STREAM_DRAW); //create CL buffer from GL buffer cl_m = clCreateFromGLBuffer(cl_c, CL_MEM_READ_WRITE, vbo, &err_code); size_t size; clGetMemObjectInfo(cl_m, CL_MEM_SIZE, sizeof(size_t), &size, NULL); printf("%p %p\n", cl_m, size);//when it is CPU it print out correnct size. with GPU both is (null) if(err_code) printf("Unable share GL object %d\n", err_code);//with GPU -34, with CPU CL_SUCCES

[genaganna]

nou,

       It looks like you are getting error(CL_INVALID_CONTEXT) from clCreateFromGLBuffer for GPU.  Could please look at the post  http://forums.amd.com/devforum/messageview.cfm?catid=390&threadid=124232 and see what is the difference b/w that code and your code?   Send a testcase if you donot see any difference.

[nou]

today i get deliverde radeon 5850. and when i try my code it works when i run on GPU. i still get segfaul when i try run on CPU device. it crash on friends 4870x2. but thanks for the answer.

[genaganna]

nou,

could you please post your code?  i will try on my system.

[nou]

sure.

g++ main.cpp -I$ATISTREAMROOT/include -L$ATISTREAMROOT/lib/x86_64 -lGL -lGLU -lOpenCL -D GL_GLEXT_PROTOTYPES -lSDL

// kernel.cl #pragma OPENCL EXTENSION cl_khr_gl_sharing : enable __kernel void Wave(float time, int grid_size, __global float *data) { int i = get_global_id(0); int o = get_global_id(1); data[o*grid_size*4+i*4+2] = sin((float)o/grid_size*6.283185308*2+time/4)+cos((float)i/grid_size*6.283185308*2+time/4);; } //main.cpp #include <stdlib.h> #include <SDL/SDL.h> #include <GL/gl.h> #include <GL/glu.h> #include <GL/glext.h> #include <GL/glx.h> #include <CL/cl.h> #include <CL/cl_gl.h> #include <math.h> GLXContext gl_context; Display *display; GLuint vbo; GLuint indicies; cl_context cl_c; cl_command_queue queue; cl_program program; cl_kernel kernel; cl_mem cl_m; cl_mem cl_m2; const int grid_size = 64; const char kernel_source[] = "#include \"kernel.cl\""; float rott = 0; const int w = 1024; const int h = 768; void InitGL() { glClearColor(0.0f, 0.5f, 0.0f, 0.0f); glViewport(0,0,w,h); glClearDepth(1.0f); glMatrixMode(GL_PROJECTION); glLoadIdentity(); //glFrustum(-0.5, 500, -0.5, 500.0, 1.0, 100.0); gluPerspective(40.0f, (float)w/h, 0.1f, 100.0f); glMatrixMode(GL_MODELVIEW); glLoadIdentity(); glGenBuffers(1, &vbo); glBindBuffer(GL_ARRAY_BUFFER, vbo); float *grid = new float[grid_size*grid_size*4]; for(int i=0;i<grid_size;i++) { for(int o=0;o<grid_size;o++) { grid[i*grid_size*4+o*4] = (float)o/grid_size*10.0-5; grid[i*grid_size*4+o*4+1] = (float)i/grid_size*10.0-5; grid[i*grid_size*4+o*4+2] = sin((float)o/grid_size*6.283185308*2)+cos((float)i/grid_size*6.283185308*2); grid[i*grid_size*4+o*4+3] = 1.0f; } } glBufferData(GL_ARRAY_BUFFER, sizeof(float)*grid_size*grid_size*4, grid, GL_STREAM_DRAW); delete [] grid; unsigned int *ind = new unsigned int[(grid_size-1)*(grid_size-1)*6]; for(int i=0;i<(grid_size-1);i++) { for(int o=0;o<(grid_size-1);o++) { ind[i*(grid_size-1)*6+o*6] = i*(grid_size)+o; ind[i*(grid_size-1)*6+o*6+1] = i*(grid_size)+o+1; ind[i*(grid_size-1)*6+o*6+2] = (i+1)*(grid_size)+o+1; ind[i*(grid_size-1)*6+o*6+3] = i*(grid_size)+o; ind[i*(grid_size-1)*6+o*6+4] = (i+1)*(grid_size)+o+1; ind[i*(grid_size-1)*6+o*6+5] = (i+1)*(grid_size)+o; } } glGenBuffers(1, &indicies); glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, indicies); glBufferData(GL_ELEMENT_ARRAY_BUFFER, sizeof(int)*(grid_size-1)*(grid_size-1)*6, ind, GL_STATIC_DRAW); delete [] ind; glPolygonMode(GL_BACK, GL_LINE); printf("GL inicializated %d\n", glGetError()); } void InitCL() { cl_int err_code; cl_platform_id platform; err_code = clGetPlatformIDs(1, &platform, NULL); if(err_code) { printf("Unable get platform %d\n", err_code); exit(1); } cl_context_properties cl_properties[] = {CL_GL_CONTEXT_KHR, (cl_context_properties)gl_context, CL_GLX_DISPLAY_KHR, (cl_context_properties)display, CL_CONTEXT_PLATFORM, (cl_context_properties)platform, 0}; cl_c = clCreateContextFromType(cl_properties, CL_DEVICE_TYPE_GPU, NULL, NULL, &err_code); if(err_code) { printf("Unable create context %d\n", err_code); exit(1); } cl_device_id device; clGetContextInfo(cl_c, CL_CONTEXT_DEVICES, sizeof(cl_device_id), &device, NULL); queue = clCreateCommandQueue(cl_c, device, NULL, &err_code); if(err_code) { printf("Unable create command queue %d\n", err_code); exit(1); } size_t source_len[] = { sizeof(kernel_source)-1 }; const char *s[] = {kernel_source , 0}; program = clCreateProgramWithSource(cl_c, 1, s, source_len, &err_code); if(err_code) { printf("Unable create program %d\n", err_code); exit(1); } err_code = clBuildProgram(program, 1, &device, "-I.", NULL, NULL); if(err_code) { printf("Unable build program %d\n", err_code); char log[10000]; clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 10000, log, NULL); printf("BUILD LOG:\n%s", log); exit(1); } kernel = clCreateKernel(program, "Wave", &err_code); cl_m2 = clCreateBuffer(cl_c, CL_MEM_READ_WRITE, sizeof(float)*grid_size*grid_size*4, NULL, &err_code); if(err_code) { printf("Unable create buffer %d\n", err_code); exit(1); } } void RunCL() { cl_int err_code; glFinish(); err_code = clEnqueueAcquireGLObjects(queue, 1, &cl_m, 0, NULL, NULL); if(err_code)printf("Unable acquire object %d\n", err_code); printf("acquire object\n"); err_code = clSetKernelArg(kernel, 0, sizeof(float), &rott); err_code |= clSetKernelArg(kernel, 1, sizeof(int), &grid_size); err_code |= clSetKernelArg(kernel, 2, sizeof(cl_mem), &cl_m); if(err_code)printf("Unable set arguments %d\n", err_code); size_t work[] = {grid_size, grid_size}; clFinish(queue); err_code = clEnqueueNDRangeKernel(queue, kernel, 2, NULL, work, NULL, 0, NULL, NULL); if(err_code) { printf("Unable run a kernel %d\n", err_code); exit(1); } err_code = clEnqueueReleaseGLObjects(queue, 1, &cl_m, 0, NULL, NULL); if(err_code)printf("Unable release object %d\n", err_code); clFinish(queue); } void DrawGL() { glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT); glLoadIdentity(); glTranslatef( 0.0f, 0.0f, -20.0f ); glRotatef(rott, 1.0f, 0.0f, 0.0f); rott += 0.4f; /*glBegin(GL_QUADS); glVertex3f(0.0f, 0.0f, 0.0f); glVertex3f(5.0f, 0.0f, 0.0f); glVertex3f(5.0f, 5.0f, 0.0f); glVertex3f(0.0f, 5.0f, 0.0f); glEnd();*/ glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, indicies); glBindBuffer(GL_ARRAY_BUFFER, vbo); glEnableClientState(GL_VERTEX_ARRAY); glVertexPointer(4, GL_FLOAT, 0, 0); glDrawElements(GL_TRIANGLES, (grid_size-1)*(grid_size-1)*6, GL_UNSIGNED_INT, NULL); glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, 0); glBindBuffer(GL_ARRAY_BUFFER, 0); glDisableClientState(GL_VERTEX_ARRAY); //printf("%d\n", glGetError()); SDL_GL_SwapBuffers(); glFinish(); } int main ( int argc, char** argv ) { // initialize SDL video if ( SDL_Init( SDL_INIT_VIDEO ) < 0 ) { printf( "Unable to init SDL: %s\n", SDL_GetError() ); return 1; } // make sure SDL cleans up before exit atexit(SDL_Quit); // create a new window SDL_Surface* screen = SDL_SetVideoMode(w, h, 32, SDL_OPENGL|SDL_GL_DOUBLEBUFFER|SDL_HWSURFACE); if ( !screen ) { printf("Unable to set 640x480 video: %s\n", SDL_GetError()); return 1; } SDL_WM_SetCaption("OpenCL - OpenGL interoperability example", NULL); gl_context = glXGetCurrentContext(); display = glXGetCurrentDisplay(); InitCL(); InitGL(); cl_int err_code; cl_m = clCreateFromGLBuffer(cl_c, CL_MEM_READ_WRITE, vbo, &err_code); size_t size; clGetMemObjectInfo(cl_m, CL_MEM_SIZE, sizeof(size_t), &size, NULL); printf("%p %p\n", cl_m, size); if(err_code) { printf("Unable share GL object %d\n", err_code); } // program main loop bool done = false; while (!done) { // message processing loop SDL_Event event; while (SDL_PollEvent(&event)) { // check for messages switch (event.type) { // exit if the window is closed case SDL_QUIT: done = true; break; // check for keypresses case SDL_KEYDOWN: { if(event.key.keysym.sym == SDLK_r) { RunCL(); } // exit if ESCAPE is pressed if (event.key.keysym.sym == SDLK_ESCAPE) done = true; break; } } // end switch } // end of message processing DrawGL(); RunCL(); SDL_Delay(10); } // end main loop return 0; }

[genaganna]

nou,

      Please first call InitGL and then call InitCL.  There is following restriction in this preview feature

To use shared resources, the OpenGL® application must first create an OpenGL® context and then an OpenCL™ context. All resources created after the OpenCL™ context has been created can be shared between OpenGL® and OpenCL™. If resources are allocated before the OpenCL™ context is created, they cannot be shared between OpenGL® and OpenCL™.

[nou]

but i create CL context after GL context. GL context create function SDL_SetVideoMode()

InitCL() create CL context and progam with kernel.

InitGL() set some OpenGL stuff and create VBO buffers.

RunCL() acquire GL object and run kernel.

whole code works on my 5850. but it crash on the card 4870X2 my friend whom i send my code to test. my 5850 run on Ubuntu 9.04. 4870X2 run on 8.10.

[omkaranathan]

Originally posted by: nou

whole code works on my 5850. but it crash on the card 4870X2 my friend whom i send my code to test. my 5850 run on Ubuntu 9.04. 4870X2 run on 8.10.

This is because the code is assuming single device.(line 119) Updating the code to query the number of device present and use the devices accordingly would solve the issue. Adding the code snippet which will run the program in first device.

cl_device_id *device; size_t numDevices; clGetContextInfo(cl_c, CL_CONTEXT_DEVICES, NULL, NULL, &numDevices); device = (cl_device_id*)malloc(numDevices * sizeof(cl_device_id)); clGetContextInfo(cl_c, CL_CONTEXT_DEVICES, numDevices * sizeof(cl_device_id), device, 0); queue = clCreateCommandQueue(cl_c, device[0], NULL, &err_code);

[david_aiken]

hi genaganna..

I tried your AMD equivalent of the NVidia SimpleGL example posted on an earlier thread and was just curious if you see a significant increase in framerate with glinterop on. On my system (Windows 7 64-bit/5850)  it actually slows down from 180fps to 150fps in a win32 release build with it on, which is not what i was expecting. Digging into it to understand why, but just curious whether other users are seeing the same thing.

thx

[genaganna]

Originally posted by: nou sure.

g++ main.cpp -I$ATISTREAMROOT/include -L$ATISTREAMROOT/lib/x86_64 -lGL -lGLU -lOpenCL -D GL_GLEXT_PROTOTYPES -lSDL

Nou,

       Please add clWaitForEvents after clEnqueueNDRangeKernel to run on CPU.  We caught this issue internally.  See modified code below

// kernel.cl const char kernel_source[] ="\n\ #pragma OPENCL EXTENSION cl_khr_gl_sharing : enable\n\ \n\ __kernel void Wave(float time, int grid_size, __global float *data)\n\ {\n\ int i = get_global_id(0);\n\ int o = get_global_id(1);\n\ data[o*grid_size*4+i*4+2] = sin((float)o/grid_size*6.283185308*2+time/4)+cos((float)i/grid_size*6.283185308*2+time/4);;\n\ }\n\ "; //main.cpp #include <stdlib.h> #include <SDL/SDL.h> #include <GL/gl.h> #include <GL/glu.h> #include <GL/glext.h> #include <GL/glx.h> #include <CL/cl.h> #include <CL/cl_gl.h> #include <math.h> GLXContext gl_context; Display *display; GLuint vbo; GLuint indicies; cl_context cl_c; cl_command_queue queue; cl_program program; cl_kernel kernel; cl_mem cl_m; cl_mem cl_m2; const int grid_size = 64; //const char kernel_source[] = "#include \"kernel.cl\""; float rott = 0; const int w = 1024; const int h = 768; void InitGL() { glClearColor(0.0f, 0.5f, 0.0f, 0.0f); glViewport(0,0,w,h); glClearDepth(1.0f); glMatrixMode(GL_PROJECTION); glLoadIdentity(); //glFrustum(-0.5, 500, -0.5, 500.0, 1.0, 100.0); gluPerspective(40.0f, (float)w/h, 0.1f, 100.0f); glMatrixMode(GL_MODELVIEW); glLoadIdentity(); glGenBuffers(1, &vbo); glBindBuffer(GL_ARRAY_BUFFER, vbo); float *grid = new float[grid_size*grid_size*4]; for(int i=0;i<grid_size;i++) { for(int o=0;o<grid_size;o++) { grid[i*grid_size*4+o*4] = (float)o/grid_size*10.0-5; grid[i*grid_size*4+o*4+1] = (float)i/grid_size*10.0-5; grid[i*grid_size*4+o*4+2] = sin((float)o/grid_size*6.283185308*2)+cos((float)i/grid_size*6.283185308*2); grid[i*grid_size*4+o*4+3] = 1.0f; } } glBufferData(GL_ARRAY_BUFFER, sizeof(float)*grid_size*grid_size*4, grid, GL_STREAM_DRAW); delete [] grid; unsigned int *ind = new unsigned int[(grid_size-1)*(grid_size-1)*6]; for(int i=0;i<(grid_size-1);i++) { for(int o=0;o<(grid_size-1);o++) { ind[i*(grid_size-1)*6+o*6] = i*(grid_size)+o; ind[i*(grid_size-1)*6+o*6+1] = i*(grid_size)+o+1; ind[i*(grid_size-1)*6+o*6+2] = (i+1)*(grid_size)+o+1; ind[i*(grid_size-1)*6+o*6+3] = i*(grid_size)+o; ind[i*(grid_size-1)*6+o*6+4] = (i+1)*(grid_size)+o+1; ind[i*(grid_size-1)*6+o*6+5] = (i+1)*(grid_size)+o; } } glGenBuffers(1, &indicies); glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, indicies); glBufferData(GL_ELEMENT_ARRAY_BUFFER, sizeof(int)*(grid_size-1)*(grid_size-1)*6, ind, GL_STATIC_DRAW); delete [] ind; glPolygonMode(GL_BACK, GL_LINE); printf("GL inicializated %d\n", glGetError()); } void InitCL() { cl_int err_code; cl_platform_id platform; err_code = clGetPlatformIDs(1, &platform, NULL); if(err_code) { printf("Unable get platform %d\n", err_code); exit(1); } cl_context_properties cl_properties[] = {CL_GL_CONTEXT_KHR, (cl_context_properties)gl_context, CL_GLX_DISPLAY_KHR, (cl_context_properties)display, CL_CONTEXT_PLATFORM, (cl_context_properties)platform, 0}; cl_c = clCreateContextFromType(cl_properties, CL_DEVICE_TYPE_CPU, NULL, NULL, &err_code); if(err_code) { printf("Unable create context %d\n", err_code); exit(1); } cl_device_id *device; size_t numDevices; err_code = clGetContextInfo(cl_c, CL_CONTEXT_DEVICES, NULL, NULL, &numDevices); if(err_code) { printf("Unable to get numDEvices \n", err_code); exit(1); } device = (cl_device_id*)malloc(numDevices * sizeof(cl_device_id)); err_code = clGetContextInfo(cl_c, CL_CONTEXT_DEVICES, numDevices * sizeof(cl_device_id), device, 0); if(err_code) { printf("Unable to get context %d\n", err_code); exit(1); } queue = clCreateCommandQueue(cl_c, device[0], NULL, &err_code); if(err_code) { printf("Unable create command queue %d\n", err_code); exit(1); } size_t source_len[] = { sizeof(kernel_source)-1 }; const char *s[] = {kernel_source , 0}; program = clCreateProgramWithSource(cl_c, 1, s, source_len, &err_code); if(err_code) { printf("Unable create program %d\n", err_code); exit(1); } err_code = clBuildProgram(program, 1, &device[0], "-I.", NULL, NULL); if(err_code) { printf("Unable build program %d\n", err_code); char log[10000]; clGetProgramBuildInfo(program, device[0], CL_PROGRAM_BUILD_LOG, 10000, log, NULL); printf("BUILD LOG:\n%s", log); exit(1); } kernel = clCreateKernel(program, "Wave", &err_code); cl_m2 = clCreateBuffer(cl_c, CL_MEM_READ_WRITE, sizeof(float)*grid_size*grid_size*4, NULL, &err_code); if(err_code) { printf("Unable create buffer %d\n", err_code); exit(1); } } void RunCL() { cl_int err_code; glFinish(); err_code = clEnqueueAcquireGLObjects(queue, 1, &cl_m, 0, NULL, NULL); if(err_code)printf("Unable acquire object %d\n", err_code); printf("acquire object\n"); err_code = clSetKernelArg(kernel, 0, sizeof(float), &rott); err_code |= clSetKernelArg(kernel, 1, sizeof(int), &grid_size); err_code |= clSetKernelArg(kernel, 2, sizeof(cl_mem), &cl_m); if(err_code)printf("Unable set arguments %d\n", err_code); size_t work[] = {grid_size, grid_size}; clFinish(queue); cl_event eventND[1]; err_code = clEnqueueNDRangeKernel(queue, kernel, 2, NULL, work, NULL, 0, NULL, eventND); if(err_code) { printf("Unable run a kernel %d\n", err_code); exit(1); } err_code = clWaitForEvents(1, eventND); if(err_code) { printf("clWaitForEvents failed : %d\n", err_code); exit(1); } clReleaseEvent(eventND[0]); err_code = clEnqueueReleaseGLObjects(queue, 1, &cl_m, 0, NULL, NULL); if(err_code)printf("Unable release object %d\n", err_code); clFinish(queue); } void DrawGL() { glClear(GL_COLOR_BUFFER_BIT | GL_DEPTH_BUFFER_BIT); glLoadIdentity(); glTranslatef( 0.0f, 0.0f, -20.0f ); glRotatef(rott, 1.0f, 0.0f, 0.0f); rott += 0.4f; /*glBegin(GL_QUADS); glVertex3f(0.0f, 0.0f, 0.0f); glVertex3f(5.0f, 0.0f, 0.0f); glVertex3f(5.0f, 5.0f, 0.0f); glVertex3f(0.0f, 5.0f, 0.0f); glEnd();*/ glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, indicies); glBindBuffer(GL_ARRAY_BUFFER, vbo); glEnableClientState(GL_VERTEX_ARRAY); glVertexPointer(4, GL_FLOAT, 0, 0); glDrawElements(GL_TRIANGLES, (grid_size-1)*(grid_size-1)*6, GL_UNSIGNED_INT, NULL); glBindBuffer(GL_ELEMENT_ARRAY_BUFFER, 0); glBindBuffer(GL_ARRAY_BUFFER, 0); glDisableClientState(GL_VERTEX_ARRAY); //printf("%d\n", glGetError()); SDL_GL_SwapBuffers(); glFinish(); } int main ( int argc, char** argv ) { // initialize SDL video if ( SDL_Init( SDL_INIT_VIDEO ) < 0 ) { printf( "Unable to init SDL: %s\n", SDL_GetError() ); return 1; } // make sure SDL cleans up before exit atexit(SDL_Quit); // create a new window SDL_GL_SetAttribute(SDL_GL_MULTISAMPLEBUFFERS, 1); SDL_GL_SetAttribute(SDL_GL_MULTISAMPLESAMPLES, 8); SDL_Surface* screen = SDL_SetVideoMode(w, h, 32, SDL_OPENGL|SDL_GL_DOUBLEBUFFER|SDL_HWSURFACE); if ( !screen ) { printf("Unable to set 640x480 video: %s\n", SDL_GetError()); return 1; } SDL_WM_SetCaption("OpenCL - OpenGL interoperability example", NULL); gl_context = glXGetCurrentContext(); display = glXGetCurrentDisplay(); InitCL(); InitGL(); cl_int err_code; cl_m = clCreateFromGLBuffer(cl_c, CL_MEM_READ_WRITE, vbo, &err_code); size_t size; clGetMemObjectInfo(cl_m, CL_MEM_SIZE, sizeof(size_t), &size, NULL); printf("%p %p\n", cl_m, size); if(err_code) { printf("Unable share GL object %d\n", err_code); } // program main loop bool done = false; while (!done) { // message processing loop SDL_Event event; while (SDL_PollEvent(&event)) { // check for messages switch (event.type) { // exit if the window is closed case SDL_QUIT: done = true; break; // check for keypresses case SDL_KEYDOWN: { if(event.key.keysym.sym == SDLK_r) { RunCL(); } // exit if ESCAPE is pressed if (event.key.keysym.sym == SDLK_ESCAPE) done = true; break; } } // end switch } // end of message processing DrawGL(); RunCL(); SDL_Delay(10); } // end main loop return 0; }

[nou]

i added clFinish() after clEnqueueNDRangeKernel and it works too. thanks