char texture

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi everyone,

Can texture memory work with characters? I get weird output.

I’m supposed to get the elements in order.

I created an array with elements A B C D E F G H I J K L M N O P Q.

but I’m getting elements A E I M…

texture<char, 1, cudaReadModeElementType> tex;

__global__

void uncompress(float *a, char *b, char *s, int N )

{

	int i = blockIdx.x * blockDim.x + threadIdx.x;

	int idx = (int) a[i];

	if ( i < N )

	{

		b[i] = tex1Dfetch(tex, idx);

	}

}

Ted

2

I’m guessing you haven’t informed it to only fetch byte (8bit) data, so it’s grabbing 32bits at a time instead. When you bind the texture, try something like:
[indent]cudaBindTexture(0, tex, charBuf, cudaCreateChannelDesc(8, 0, 0, 0, cudaChannelFormatKindUnsigned), BUFSIZE);[/indent]
where charBuf is a pointer to the char buffer, and BUFSIZE is the count of bytes in the buffer. Use “cudaChannelFormatKindSigned” if you really want signed “char” types rather than unsigned.

Regards,
Paul

3

I’m guessing you haven’t informed it to only fetch byte (8bit) data, so it’s grabbing 32bits at a time instead. When you bind the texture, try something like:
[indent]cudaBindTexture(0, tex, charBuf, cudaCreateChannelDesc(8, 0, 0, 0, cudaChannelFormatKindUnsigned), BUFSIZE);[/indent]
where charBuf is a pointer to the char buffer, and BUFSIZE is the count of bytes in the buffer. Use “cudaChannelFormatKindSigned” if you really want signed “char” types rather than unsigned.

Regards,
Paul