CUDA与OpenGL交互实现YUV视频转换为RGB后的渲染问题

chunjinglinglan 2011-05-18 09:48:36

请教大家了！我写的一个程序将本地YUV视频文件读取到CPU内存当中，并拷贝给GPU上的内存，在 GPU上进行YUV到RGB的转换后，再将转换后的数据拷贝出来，在CPU端窗口上显示。这些都实现了，能正确地播放视频画面。
现在我想做的是，在GPU上YUV转RGB后，不进行拷贝，直接把RGB数据绑定给一个二维纹理，然后在CPU端利用OpenGL的缓存对象PBO和纹理缓冲区对象进行映射，然后进行渲染播放。可是运行程序总会出现“GLUT: Warning in d:\XXXX.exe: GL error : 无效列举”的问题，不知道问题出在哪里，求教！
以下是我的程序：
RGB数据绑定给一个二维纹理
__global__ void rgb_tex( unsigned char* d_odata, int width, int height)
{
// calculate normalized texture coordinates
unsigned int x = blockIdx.x*blockDim.x + threadIdx.x;
unsigned int y = blockIdx.y*blockDim.y + threadIdx.y;

if (x >= width)
return; //i = 720 - 1;
if (y >= height)
return; // j = 384 - 1;

float u = (float)x;
float v = (float)y;

// read from texture and write to global memory
d_odata[y*720*3 + x] = tex2D(tex, u, v);
}

extern "C"
void initTexture(int width, int height, unsigned char *d_rgb）
{//实现将yuv2rgb后在global memory中的rgb数据与纹理数组绑定

int size = width * height;

// allocate array and copy image data
cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc(8, 0, 0, 0, cudaChannelFormatKindUnsigned);
//cudaArray* cu_array;
cutilSafeCall( cudaMallocArray( &cu_array, &channelDesc, width, height ));
cutilSafeCall( cudaMemcpyToArray( cu_array, 0, 0, d_rgb, size, cudaMemcpyDeviceToDevice));

// set texture parameters
tex.addressMode[0] = cudaAddressModeWrap;
tex.addressMode[1] = cudaAddressModeWrap;
tex.filterMode = cudaFilterModePoint;
tex.normalized = false; // access with normalized texture coordinates

// Bind the array to the texture
cutilSafeCall( cudaBindTextureToArray( tex, cu_array, channelDesc));
}

extern "C"
void run_yuv2rgb(unsigned char * in_data,unsigned char *out_data)
{
//设置线程块等的大小，对于720*384的图像
dim3 dimGrid(45,24);
dim3 dimBlock(16,16);

//调用kernel函数执行
yuv2rgb<<<dimGrid,dimBlock>>>(in_data,out_data);
}

extern "C"
void runtexture(unsigned char* d_odata, int width_rgb, int height)
{
//设置线程块等的大小，对于720*384的RGB图像
dim3 Grid(135,24);
dim3 Block(16,16);

rgb_tex<<<Grid,Block>>>(d_odata,width_rgb,height);
}

int main(int argc, char* argv[])
{
if(!InitCUDA()) {
return 0;
}
int size = 0;
//unsigned char* d_yuv_data;
//unsigned char* d_rgb_data;
cutilSafeCall(cudaMalloc((void**) &d_yuv_data, yuv_size));
cutilSafeCall(cudaMalloc((void**) &d_rgb_data, rgb_size));

// initialize GLUT
glutInit(&argc, argv);
glutInitDisplayMode(GLUT_RGB | GLUT_DOUBLE);
glutInitWindowSize(window_width, window_height);
glutCreateWindow("yuv2rgb_render");

printf(">> initGL() creating window [%d x %d]\n", window_width, window_height);

glewInit();
if (!glewIsSupported( "GL_VERSION_1_5 GL_ARB_vertex_buffer_object GL_ARB_pixel_buffer_object" )) {
fprintf(stderr, "Error: failed to get minimal extensions for demo\n");
fprintf(stderr, "This sample requires:\n");
fprintf(stderr, " OpenGL version 1.5\n");
fprintf(stderr, " GL_ARB_vertex_buffer_object\n");
fprintf(stderr, " GL_ARB_pixel_buffer_object\n");
return true;
}

glutDisplayFunc(display);

FILE* YuvFile=fopen("E:/camera0_420.yuv", "rb");
if (YuvFile==NULL)
{
printf("Can not open the yuv file!");
exit(0);
}

YuvData=(unsigned char*)malloc(yuv_size);
size=fread(YuvData,sizeof(unsigned char),yuv_size,YuvFile);
if(size!=yuv_size)
{
printf("the yuv422 video format is wrong!");
exit(0);
}

cutilSafeCall(cudaMemcpy(d_yuv_data,YuvData, yuv_size, cudaMemcpyHostToDevice) );

run_yuv2rgb(d_yuv_data,d_rgb_data);//运行yuv2rgb内核函数
printf("yuv2rgb has been run!\n");

//unsigned char *d_out_data_final;
cutilSafeCall(cudaMalloc((void**) &d_rgb_data_final, rgb_size));

initTexture(image_width_rgb,image_height,d_rgb_data);//RGB数据与二维纹理绑定

initOpenGL();

runtest();
printf("runtest has been run!\n");

atexit(deletePBO);

glutMainLoop();
}

void display()
{
//unsigned char *d_rgb_data_final;
cudaGLMapBufferObject((void**)&d_rgb_data_final,Image_pbo);

//run_yuv2rgb(d_yuv_data,d_rgb_data);
runtexture(d_rgb_data_final,image_width_rgb,image_height);
printf("rgb_tex has been run!\n");

cudaGLUnmapBufferObject(Image_pbo);

glClear(GL_COLOR_BUFFER_BIT);
glEnable(GL_TEXTURE_2D);

// load texture from pbo
glBindBufferARB (GL_PIXEL_UNPACK_BUFFER_ARB, Image_pbo);
glBindTexture (GL_PIXEL_UNPACK_BUFFER_ARB, Image_tex);
glTexSubImage2D (GL_PIXEL_UNPACK_BUFFER_ARB, 0, 0, 0, image_width_rgb, image_height, GL_RGB, GL_UNSIGNED_BYTE, 0/*d_rgb_data_final*/);
CUT_CHECK_ERROR_GL();
glBindBufferARB (GL_PIXEL_UNPACK_BUFFER_ARB, 0);

// fragment program is required to display floating point texture
glBindProgramARB(GL_FRAGMENT_PROGRAM_ARB, gl_shader_);
glEnable(GL_FRAGMENT_PROGRAM_ARB);
glDisable(GL_DEPTH_TEST);

glBegin(GL_QUADS);
glTexCoord2f(0, 1); glVertex2f(0, 0);
glTexCoord2f(1, 1); glVertex2f(1, 0);
glTexCoord2f(1, 0); glVertex2f(1, 1);
glTexCoord2f(0, 0); glVertex2f(0, 1);
glEnd();

glBindTexture(GL_PIXEL_UNPACK_BUFFER_ARB, 0);
glDisable(GL_FRAGMENT_PROGRAM_ARB);

glutSwapBuffers();
glutReportErrors();

}

...全文