Texture绑定时遇到InvalidTexture

大家好,我刚接触cuda编程。我想写一个类似于cook的光线追踪程序,不过在读取纹理时遇到一些问题。运行时cuda提示cudaErrorInvalidTexture,我对比了几个例程依然没找出错误来。具体代码如下。
texture<float4, 2, cudaReadModeElementType> tex;
//read texture
void readTexture(char *filename, uint width, uint height) {
//Load ppm picture
unsigned char *h_data;
sdkLoadPPM4ub(filename, &h_data, &width, &height);

	float4 *h_output = new float4[width*height];

	for(int i = 0; i < width*height; i++) {
		h_output[i].x = h_data[4*i];
		h_output[i].y = h_data[4*i+1];
		h_output[i].z = h_data[4*i+2];
		h_output[i].w = h_data[4*i+3];
		//cout << h_output[i].x <<  " " << h_output[i].y << " " << h_output[i].z << " " << h_output[i].w << endl;
	}

	// Copy to device memory some data located at address h_data in host memory
	cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<float4>();
	cudaArray *cuArray;
	checkCudaErrors(cudaMallocArray(&cuArray, &channelDesc, width, height));
	cudaMemcpyToArray(cuArray, 0, 0, h_output, sizeof(h_output),cudaMemcpyHostToDevice);

	// Set texture parameters
	tex.addressMode[0] = cudaAddressModeWrap;
	tex.addressMode[1] = cudaAddressModeWrap;
	tex.filterMode = cudaFilterModeLinear;
	tex.normalized = true;    // access with normalized texture coordinates

	// Bind the array to the texture
	checkCudaErrors(cudaBindTextureToArray(tex, cuArray, channelDesc));

}

PS:如果要调用15个纹理,是全局定义15个纹理比较好,还是在每个类内定义纹理对象比较好呢?

楼主您好,通过阅读您的代码,未能发现任何问题。

您的texture<>声明、cudaMallocArray(), 以及cudaBindTextureToArray(), 均正确无误。

您的最后的cudaBindTextureToArray()应该返回cudaSuccess的。而不是cudaErrorInvalidTexture。

因为本人水平有限,错误在所难免,只能给出上述观点了。

我发现我工程配置的问题应该是。。感谢版主!
不过我读入一张512*1024的颜色为黄色的图片,贴出来的是红色的图。。输出h_output发现都是205,205,205.是这段代码的问题吗。。。?

如果您是普通的4B/像素的图片,那么您不应该使用float4作为元素类型的。
(那将会是16B!)

使用uchar4之类的也许不错。:slight_smile:

不好意思。。还得打扰。。
我发现纹理总是不显示,然后__device__里输出结果发现颜色都是0.。。。
printf(“tex: %.2f %.2f %.2f”, tex2D(tex, 0.1, 0.1).x, tex2D(tex, 0.1, 0.1).y, tex2D(tex, 0.1, 0.1).z);

请问我可能是哪里出了错误= =|||?

不知道。

以及你没有回答我上个帖子中的询问。
(也许人家只是uchar4, 外加cudaReadModeNormalizedFloat返回的float4, 而不您这样的直接定义元素问float4的)

是4B的图片。。。我还是改成uchar4吧。

错就是错,对就是对。你不认识我,你不必在乎面子。

既然是4B, 请火速改成uchar4。
以及我看你需要float4, 不要忘记最后cudaReadModeNormalizedFloat。

感谢配合工作。

我使用的是cook(光线追踪的桌球)的pool_1.ppm那张图,颜色的确是4B的。
我改了一下代码。
cudaArray *cuArray;
texture<uchar4, 2, cudaReadModeNormalizedFloat> tex;

//read texture
void readTexture(char *filename,const char *exe_path, uint width, uint height) {

	char *imgPath = sdkFindFilePath(filename, exe_path);
	//Load ppm picture
	unsigned char *h_data = new unsigned char[width*height*4];
	if (sdkLoadPPM4(imgPath, &h_data, &width, &height) == 0)
		cout << "Error load pic." << endl;

	uchar4 *h_output = new uchar4[width*height];

	for(int i = 0; i < width*height; i++) {
		h_output[i].x = h_data[4*i];
		h_output[i].y = h_data[4*i+1];
		h_output[i].z = h_data[4*i+2];
		h_output[i].w = h_data[4*i+3];
	}
	//cout << h_output[0].x <<  " " << h_output[0].y << " " << h_output[0].z << " " << h_output[0].w << endl;

	// Copy to device memory some data located at address h_data in host memory
	cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<uchar4>();
	checkCudaErrors(cudaMallocArray(&cuArray, &channelDesc, width, height));
	cudaMemcpyToArray(cuArray, 0, 0, h_output, sizeof(h_output),cudaMemcpyHostToDevice);

	// Set texture parameters
	tex.addressMode[0] = cudaAddressModeWrap;
	tex.addressMode[1] = cudaAddressModeWrap;
	tex.filterMode = cudaFilterModeLinear;
	tex.normalized = true;    // access with normalized texture coordinates

	// Bind the array to the texture
	checkCudaErrors(cudaBindTextureToArray(tex, cuArray, channelDesc));

}

cudaMemcpyToArray(cuArray, 0, 0, h_output, sizeof(h_output),cudaMemcpyHostToDevice);
这句sizeof(h_output),改成sizeof(uchar4)widthheight,不然你只copy了4B(或者8B,在64位系统)的数据