Help for one little crash.

Accelerated Computing CUDA CUDA Programming and Performance
1

Hello.
I am trying to setup a little CUDA/GL interop example. I have looked around in the internet, so I found some tutorials with some helpful stuff.

All I want is, is to produce a texture in CUDA and draw it with OpenGL.

The source I have now is crashing my Macbook Pro every time I run it, so I thought that if somebody could take an eye on it, that would be really helpful.

GLuint viewGLTexture;
cudaGraphicsResource_t viewCudaResource;

void initGLandCUDA() {
    int argc = 0;
    char** argv = NULL;
    glutInit(&argc, argv);
    glutInitDisplayMode(GLUT_RGBA | GLUT_DOUBLE);
    glutInitWindowSize(window_width, window_height);
    glutCreateWindow("CUDA GL Interop");
 
    glewInit();
    
    glEnable(GL_TEXTURE_2D);
    glGenTextures(1, &viewGLTexture);
    glBindTexture(GL_TEXTURE_2D, viewGLTexture);
    {
        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
        glTexImage2D(GL_TEXTURE_2D, 0, GL_RGBA, 512, 512, 0, GL_RGBA, GL_UNSIGNED_BYTE, NULL);
    }
    glBindTexture(GL_TEXTURE_2D, 0);
    
    cudaGLSetGLDevice(gpuGetMaxGflopsDeviceId());
    cudaGraphicsGLRegisterImage(&viewCudaResource, viewGLTexture, GL_TEXTURE_2D, cudaGraphicsRegisterFlagsWriteDiscard);
}


__global__ void renderingKernel(cudaSurfaceObject_t image) {
    unsigned int x = blockIdx.x*blockDim.x + threadIdx.x;
    unsigned int y = blockIdx.y*blockDim.y + threadIdx.y;
 
    uchar4 color = make_uchar4(0.f, 0.f, 0.f, .3f);
    //if I write in 0, 0 and not x,y, the computer is not crashing, but there is no black pixel at 0,0
    surf2Dwrite(color, image, x, y, cudaBoundaryModeClamp);
}


void callCUDAKernel(cudaSurfaceObject_t image) {
    dim3 block(8, 1, 1);
    dim3 grid(8, 1, 1);
    renderingKernel<<< grid, block>>>(image);
}

void renderFrame() {
    cudaGraphicsMapResources(1, &viewCudaResource);
    {
        cudaArray_t viewCudaArray;
        checkCudaErrors(cudaGraphicsSubResourceGetMappedArray(&viewCudaArray, viewCudaResource, 0, 0));
        cudaResourceDesc viewCudaArrayResourceDesc;
        {
            viewCudaArrayResourceDesc.resType = cudaResourceTypeArray;
            viewCudaArrayResourceDesc.res.array.array = viewCudaArray;
        }
        cudaSurfaceObject_t viewCudaSurfaceObject;
        checkCudaErrors(cudaCreateSurfaceObject(&viewCudaSurfaceObject, &viewCudaArrayResourceDesc));
        
        callCUDAKernel(viewCudaSurfaceObject);
        
        checkCudaErrors(cudaDestroySurfaceObject(viewCudaSurfaceObject));
    }
    checkCudaErrors(cudaGraphicsUnmapResources(1, &viewCudaResource));
    checkCudaErrors(cudaStreamSynchronize(0));
    
    glBindTexture(GL_TEXTURE_2D, viewGLTexture);
    {
        glBegin(GL_QUADS);
        {
            glTexCoord2f(0.0f, 0.0f); glVertex2f(-1.0f, -1.0f);
            glTexCoord2f(1.0f, 0.0f); glVertex2f(+1.0f, -1.0f);
            glTexCoord2f(1.0f, 1.0f); glVertex2f(+1.0f, +1.0f);
            glTexCoord2f(0.0f, 1.0f); glVertex2f(-1.0f, +1.0f);
        }
        glEnd();
    }
    glBindTexture(GL_TEXTURE_2D, 0);
    glFinish();
}

int main(int argc, char **argv)
{
    initGLandCUDA();
    
    glutDisplayFunc(renderFrame);
    //glutKeyboardFunc(keyboard);
    //glutMouseFunc(mouse);
    glutMainLoop();
}

It seems like some kind of out-of-range error, but I am currently out of ideas (btw, this is cc 3.0, running to nVidia 650M).

Thanks.

2

A good fellow has fixed it here : c++ - Crash with CUDA/OGL interop - Stack Overflow

3

As a result of this fixed, here is what I got CUDA/OpenGL interop + simple CUDA raytrace - YouTube :)
Happy holidays.