CUDA纹理使用方法.doc

一、例子1:fluidsGL中:
(一)host端
1)声明实际纹理数据
float2 *hvfield = NULL;
float2 *dvfield = NULL;
2) 对host数据和device数据进行初始化
hvfield = (cData*)malloc(sizeof(cData) * DS);
memset(hvfield, 0, sizeof(cData) * DS);
// Allocate and initialize device data
cudaMallocPitch((void**)&dvfield, &tPitch, sizeof(cData)*DIM, DIM);
cudaMemcpy(dvfield, hvfield, sizeof(cData) * DS,
cudaMemcpyHostToDevice);
调用device端函数,建立并绑定纹理数据
setupTexture(DIM, DIM);
bindTexture(); //注意,其中将array和纹理标识texref绑定
(二)device端
1)声明纹理对象标识,以及纹理实际数据数组
texture<float2, 2> texref;
static cudaArray *array = NULL;
2)实现纹理建立、绑定函数
void setupTexture(int x, int y) {
// Wrap mode appears to be the new default
= cudaFilterModeLinear;
cudaChannelFormatDesc desc = cudaCreateChannelDesc<float2>();
cudaMallocArray(&array, &desc, y, x);
CUT_CHECK_ERROR("cudaMalloc failed");
}
void bindTexture(void) {
cudaBindTextureToArray(texref, array);
CUT_CHECK_ERROR("cudaBindTexture failed");
}
void unbindTexture(void) {
cudaUnbindTexture(texref);
CUT_CHECK_ERROR("cudaUnbindTexture failed");
}
//传入数据
void updateTexture(cData *data, size_t wib, size_t h, size_t pitch) {
cudaMemcpy2DToArray(array, 0, 0, data, pitch, wib, h, cudaMemcpyDeviceToDevice);
CUT_CHECK_ERROR("cudaMemcpy failed");
}
void deleteTexture(void) {
cudaFreeArray(array);
CUT_CHECK_ERROR("cudaFreeArray failed");
}
3)访问纹理