Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Something wrong about using cuda texture processing colorful image #1

Open
github2016-yuan opened this issue Nov 16, 2017 · 0 comments

Comments

@github2016-yuan
Copy link

github2016-yuan commented Nov 16, 2017

I have learned your code and write down some code to do some image processing. But my code could get right result and when debugging it I found that my code even can't show the colorful image.I tried a few days but still cannot get the right result image. May I post the code here to get some help for you?
(I delete the code of image processing and just I/O operation remained)
global void Kernel2(float* output, int width, int height)
{
int x = blockIdx.x * blockDim.x + threadIdx.x;
int y = blockIdx.y * blockDim.y + threadIdx.y;
// int idx = (y * width + x) * 4;
int idx = y * width + x;

if(x < width && y < height)
{

// float u = x / (float)width;
// float v = y / (float)height;

	 output[idx] = float(tex2D(texRef, x + 0.5,y + 0.5)) ;
	 output[idx+1] = float(tex2D(texRef, x + 0.5 + 1,y + 0.5));
	 output[idx+2] = float(tex2D(texRef,x + 0.5 + 2,y + 0.5));
	 output[idx+3] = float(tex2D(texRef, x + 0.5 + 3, y + 0.5));
}

}
const int width = 512;
const int height = 512;
int main()
{
cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc(32, 0, 0, 0, cudaChannelFormatKindFloat);
cudaArray* cuArray;
cudaMallocArray(&cuArray, &channelDesc, width*4, height);

Mat image = imread("lena.jpg");
imshow("source-texture1", image);
cvtColor(image,image,CV_BGR2RGBA);
//convert from uchar to float
image.convertTo(image,CV_32FC4);

cudaMemcpyToArray(cuArray, 0, 0, image.data, width*height*4*sizeof(float), cudaMemcpyHostToDevice);

   //ste texture
texRef.addressMode[0] = cudaAddressModeWrap; 
texRef.addressMode[1] = cudaAddressModeWrap;
texRef.filterMode = cudaFilterModePoint; 
texRef.normalized = false; 

   //bind
cudaBindTextureToArray(texRef, cuArray, channelDesc);



Mat imageOutput = Mat(Size(width, height), CV_32FC4);
float * output = NULL;

cudaMalloc((void**)&output, width * height * sizeof(float)*4);

dim3 dimBlock(16, 16);
dim3 dimGrid((width + dimBlock.x - 1) / dimBlock.x, (height + dimBlock.y - 1) / dimBlock.y);

Kernel2<<<dimGrid,dimBlock>>>(output,width,height);

cudaMemcpy(imageOutput.data, output, 4*height*width*sizeof(float), cudaMemcpyDeviceToHost);

imageOutput.convertTo(imageOutput,CV_8UC4);
cvtColor(imageOutput,imageOutput,CV_RGBA2BGR);
imwrite("texture1.jpg",imageOutput);
imshow("texture1", imageOutput);
waitKey();

cudaFreeArray(cuArray);
cudaFree(output);

}

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
None yet
Development

No branches or pull requests

1 participant