Grayscale Texture

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi gys,

Below is my code for conversoin of rgba image to greyscale. I am using textures but the output isnt coming what is expected.

kernel code:

float elapsedtime;

texture<uchar4, 2, cudaReadModeElementType> texSrc;

__global__ void convert(unsigned char *iin_1)

{

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

	int ty = threadIdx.y + blockIdx.y * blockDim.y;

	int offset = tx + ty * blockDim.x * gridDim.x;

const float  x = (float)tx + 0.5f;

    const float  y = (float)ty + 0.5f;

	if(offset < 320*240)

        return;

	uchar4 temp;

	temp = tex2D(texSrc,x,y);

	float color = 0.3 * temp.x + 0.6 * temp.y + 0.1 * temp.z ;

	iin_1[offset] = color;

}

function code:

float tograyscale(unsigned char *in, unsigned char * in_1)

{

	unsigned char *iin_1;

	cudaMalloc((void **)&iin_1, (240*320*sizeof(unsigned char)));

	cudaArray *src;

cudaChannelFormatDesc floatTex = cudaCreateChannelDesc(8, 8, 8, 8, cudaChannelFormatKindUnsigned);

    cudaMallocArray(&src, &floatTex, 240, 320);

    cudaMemcpyToArray(src, 0, 0, in, 4 * 240 * 320, cudaMemcpyHostToDevice);

    cudaBindTextureToArray(texSrc, src, floatTex);

	dim3 grid(18,18);

	dim3 block(16,16);

	convert<<<grid,block>>>(iin_1);

	cudaMemcpy( in_1, iin_1, (240*320*sizeof(unsigned char)), cudaMemcpyDeviceToHost);

	return elapsedtime;	

}

The output is all messed up. Can anyone tell me whats is it that i am doing wrong…

Thnx…

EDIT: The in buffer is in rgba format and in_1 buffer is in 1 channel format.

2

Hi gys,

Below is my code for conversoin of rgba image to greyscale. I am using textures but the output isnt coming what is expected.

kernel code:

float elapsedtime;

texture<uchar4, 2, cudaReadModeElementType> texSrc;

__global__ void convert(unsigned char *iin_1)

{

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

	int ty = threadIdx.y + blockIdx.y * blockDim.y;

	int offset = tx + ty * blockDim.x * gridDim.x;

const float  x = (float)tx + 0.5f;

    const float  y = (float)ty + 0.5f;

	if(offset < 320*240)

        return;

	uchar4 temp;

	temp = tex2D(texSrc,x,y);

	float color = 0.3 * temp.x + 0.6 * temp.y + 0.1 * temp.z ;

	iin_1[offset] = color;

}

function code:

float tograyscale(unsigned char *in, unsigned char * in_1)

{

	unsigned char *iin_1;

	cudaMalloc((void **)&iin_1, (240*320*sizeof(unsigned char)));

	cudaArray *src;

cudaChannelFormatDesc floatTex = cudaCreateChannelDesc(8, 8, 8, 8, cudaChannelFormatKindUnsigned);

    cudaMallocArray(&src, &floatTex, 240, 320);

    cudaMemcpyToArray(src, 0, 0, in, 4 * 240 * 320, cudaMemcpyHostToDevice);

    cudaBindTextureToArray(texSrc, src, floatTex);

	dim3 grid(18,18);

	dim3 block(16,16);

	convert<<<grid,block>>>(iin_1);

	cudaMemcpy( in_1, iin_1, (240*320*sizeof(unsigned char)), cudaMemcpyDeviceToHost);

	return elapsedtime;	

}

The output is all messed up. Can anyone tell me whats is it that i am doing wrong…

Thnx…

EDIT: The in buffer is in rgba format and in_1 buffer is in 1 channel format.