CUDA和OpenCV实现的图像GAMMA变换

分手后的思念是犯贱 2022-08-08 05:06 161阅读 0赞

**\[cpp\]** [view plain][] [copy][view plain]

1.  \#include <stdlib.h>
2.  \#include <stdio.h>
3.  \#include <cv.h>
4.  \#include <highgui.h>
5.  \#include "cutil\_inline.h"
6.  
7.  \#define GAMMA 0.4
8.  
9.  **void** runTest(**int** argc, **char**\*\* argv); 
10. 
11. \_\_global\_\_ **void** testKernel(**float**\* d\_idata, **float**\* d\_odata, **int** width, **int** height, **float** gamma) 
12. \{ 
13.  unsigned **int** tid\_in\_grid\_x = blockDim.x\*blockIdx.x + threadIdx.x; 
14.  unsigned **int** tid\_in\_grid\_y = blockDim.y\*blockIdx.y + threadIdx.y; 
15.  unsigned **int** tid\_in\_grid = tid\_in\_grid\_y\*width + tid\_in\_grid\_x; 
16. 
17.  d\_odata\[tid\_in\_grid\] = powf(d\_idata\[tid\_in\_grid\], gamma); 
18. \} 
19. 
20. **int** main(**int** argc, **char**\*\* argv) 
21. \{ 
22.  runTest(argc, argv); 
23.  CUT\_EXIT(argc, argv); 
24. \} 
25. 
26. **void** runTest(**int** argc, **char**\*\* argv) 
27. \{ 
28. **if**( cutCheckCmdLineFlag(argc, (**const****char**\*\*)argv, "device") ) //设备初始化并设定计时器。
29.  cutilDeviceInit(argc, argv); 
30. **else**
31.  cudaSetDevice(cutGetMaxGflopsDeviceId() ); 
32. 
33.  unsigned **int** timer = 0; 
34.  cutilCheckError(cutCreateTimer(&timer)); 
35.  cutilCheckError(cutStartTimer(timer)); 
36. 
37.  IplImage\* pImg; 
38. **if**((pImg = cvLoadImage("lena\_gray.jpg", CV\_LOAD\_IMAGE\_GRAYSCALE)) != 0 ) 
39.  \{ 
40.  cvNamedWindow("Image", 1); 
41.  cvShowImage("Image", pImg); 
42. 
43.  unsigned **int** num\_blocks\_x = pImg->width/16; 
44.  unsigned **int** num\_blocks\_y = pImg->height/16; 
45. 
46.  unsigned **int** mem\_size = **sizeof**(**float**)\*pImg->widthStep\*pImg->height; 
47. 
48. **float**\* h\_idata = (**float**\*)malloc(mem\_size); 
49. **float**\* h\_odata = (**float**\*)malloc(mem\_size); 
50. 
51. **float**\* d\_idata; 
52.  CUDA\_SAFE\_CALL(cudaMalloc((**void**\*\*)&d\_idata, mem\_size)); 
53. **float**\* d\_odata; 
54.  CUDA\_SAFE\_CALL(cudaMalloc((**void**\*\*)&d\_odata, mem\_size)); 
55. 
56. **for**(**int** i = 0; i < pImg->widthStep\*pImg->height; i++) //从图像中复制数据时要特别注意变量类型，char类型的数据计算精度 
57. 
58.  h\_idata\[i\] = ((uchar)pImg->imageData\[i\])/255.0; //太低，但是图像中只能存储char类型，所以要进行转换
59. 
60.  CUDA\_SAFE\_CALL(cudaMemcpy(d\_idata, h\_idata, mem\_size, cudaMemcpyHostToDevice)); 
61. 
62.  dim3 grid(num\_blocks\_x, num\_blocks\_y, 1); 
63.  dim3 threads(16, 16, 1); 
64. 
65.  testKernel<<<grid, threads>>>(d\_idata, d\_odata, pImg->width, pImg->height, GAMMA); 
66. 
67.  CUT\_CHECK\_ERROR("Kernel execution failed"); 
68. 
69.  CUDA\_SAFE\_CALL(cudaMemcpy(h\_odata, d\_odata, mem\_size, cudaMemcpyDeviceToHost)); 
70. 
71.  IplImage\* oImg = cvCreateImage(cvSize(pImg->width,pImg->height), IPL\_DEPTH\_8U, 1); 
72. **for**(**int** i = 0; i < pImg->widthStep\*pImg->height; i++) 
73.  oImg->imageData\[i\] = (uchar)(**int**)(h\_odata\[i\]\*255); 
74. 
75.  cvNamedWindow("Result", CV\_WINDOW\_AUTOSIZE); 
76.  cvShowImage("Result", oImg); 
77. // cvSaveImage("result.jpg", oImg);
78. 
79.  cutilCheckError( cutStopTimer( timer)); 
80.  printf( "Processing time: %f (ms)/n", cutGetTimerValue( timer)); 
81.  cutilCheckError( cutDeleteTimer( timer)); 
82. 
83.  cvWaitKey(0); 
84.  cvDestroyWindow("Image"); 
85.  cvReleaseImage(&pImg); 
86.  cvDestroyWindow("Result"); 
87.  cvReleaseImage(&oImg); 
88. 
89.  free(h\_idata); 
90.  free(h\_odata); 
91.  CUDA\_SAFE\_CALL(cudaFree(d\_idata)); 
92.  CUDA\_SAFE\_CALL(cudaFree(d\_odata)); 
93. 
94.  \} 
95. \} 
96.

[view plain]: http://blog.csdn.net/augusdi/article/details/12209191#