大牛们，你们好。 我 最近在做用CUDA编程做一个YUV视频优化的程序。
现在已经进展到这一步了，
YUV的视频文件读取到CPU内存当中，并拷贝给GPU上的内存，在 GPU上进行YUV到RGB的转换后，再将转换后的数据拷贝出来，在CPU端窗口上显示。这些都实现了，能正确地播放视频画面。
现在我想做的是， 在GPU上YUV转RGB后，不进行拷贝，直接把RGB数据绑定给一个二维纹理，然后在CPU端利用OpenGL的缓存对象进行映射，然后进行渲染播放。
以下是我写的接口
.cu文件
texture<float, 2, CudaReadModeElementType> tex;
// 这里纹理参考是该用float, 还是该用UINT? 因为我的纹理数据在GPU全局内存中，每个像素是3个BYTE
__global__static void d_render(float*d_output, int imageW, int imageH)
{
int x = blockIdx.x * blockDim.x + threadIdx.x;
int y = blockIdx.y * blockDim.y + threadIdx.y;
float u = x / (float)imageW;
float v = y / (float)imageH;
if(x<imageW && y<imageH)
{
int i = y * imageW + x;
d_output[i] = tex2D(tex, u, v);
}
__syncthreads();
}
void InitTexture(BYTE*RGBbuf, int width, int height)
{
cudaChannelFormatDesc channelDesc = cudaCreateChannel();
size_t size = width * height * sizeof(float) * 4;
//上面这句该不该这么写。还是该写成size = width *height * sizeof(float) * 3; 还是写成其他的？因为实际上只有width *height * sizeof(BYTE) * 3这么大的一块内存
tex.normalized = true;
tex.filterMode = cudaFilterModeLinear;
tex.addressMode[0] = cudaAddressModeClamp;
tex.addressMode[1] = cudaAddressModeClamp;
cudaBindTexture2D(0, &tex, (void*)RGBbuf, &channelDesc, width, height, size);
}
extern “C” void run_kernel(float*d_output, int width, int height)
{
.
.
.
// 到这里，gpu上已经完成YUV到RGB的转换，并且像素数据保存在gpu_RGB指向的全局内存快。
dim3 block(32, 16, 1);
dim3 grid(11, 18, 1);
// 块的划分是依据画面大小是352288 = 32161118, 让每个块运行512条线程，每个线程处理一个像素。
InitTexture(gpu_RGB, width, height);
d_render<<<grid, block, 0>>>(d_output, width, height, 0);
cudaError_t err = cudaGetLastError(); // 这里调试的时候，提示没有错误。
.
.
.
}
.cpp 中视窗的接口
// 像素缓存的初始化
void ChildWindow::InitPixBuffer()
{
.
.
.
// 在这之前，我确保OpenGL的窗口设置已经完成， 因为我直接调用openGL的一些函数能在窗口画出图来。
glewInit();
cudaGLsetDevice(cutGetMaxGflopsDeviceId());
glGenBuffersARB(1, &m_pbo); // m_pbo是GLuint类型
glBindbufferARB(GL_ARRAY_BUFFER, m_pbo);
glBufferDataARB(GL_ARRAY_BUFFER, iWidth * iHeight * sizeof(float) * 4, 0, GL_DYNAMIC_DRAW);
// 上面这句的疑问同红线标记2
cudaGLRegisterBufferObject(m_pbo);
}
void ChildWindow::render()
{
cudaGLMapBufferObject((void**)&m_ndevptr, m_pbo); // m_ndevptr是float*型
run_Cuda(Y, Cb, Cr, RGBbuf, m_ndevptr, width, height);
// 上面这个接口是内核程序DLL导出的接口，内部调用了run_kernel
cudaGLUnmapBufferObject(m_pbo);
glDisable(GL_DEPTH_TEST);
glRasterPos2i(0, 0);
glBindBufferARB(GL_PIXEL_UNPACK_BUFFER_ARB, m_pbo);
glDrawPixels(width, height, GL_RGB, GL_UNSIGNED_BYTE, 0);
glBindBufferARB(GL_PIXEL_UNPACK_BUFFER_ARB, 0);
}
在 窗口那端，先调用InitPixBuffer(),再在OnPaint()里调用render(),窗口显示带花线的黑屏。
代码里的width 是指图像的宽度 352， height是图像的高度288.
[ 本帖最后由 yinquan8241 于 2010-4-20 16:48 编辑 ]