Blame view

cpp/cudafunc.cu 2.74 KB
a744d027   Laila Saadatifard   upload the ivote3...
1
2
3
4
5
6
7
8
  /*#include "circle_check.cuh"
  
  void test_3(float* gpu_out, float* gpu_grad, float rmax, float phi, int n, int x, int y, int z){
  gpu_test3(gpu_out, gpu_grad, rmax, phi, n, x, y, z);
  }
  */
  
  
c9aba18a   Laila Saadatifard   ivote3 on the GPU...
9
10
  #include "gaussian_blur3.cuh"
  #include "gradient3.cuh"
f12505fb   Laila Saadatifard   upload the ivote ...
11
  #include "mag3.cuh"
a744d027   Laila Saadatifard   upload the ivote3...
12
  #include "vote3_atomic_aabb.cuh"
1f55a874   Laila Saadatifard   upload the fixed ...
13
  #include "update_dir3_aabb.cuh"
f12505fb   Laila Saadatifard   upload the ivote ...
14
  #include "local_max3.cuh"
c9aba18a   Laila Saadatifard   ivote3 on the GPU...
15
16
  
  
46820b25   Laila Saadatifard   create an obj fil...
17
  void ivote3(float* img, float sigma[], float anisotropy, float phi, float d_phi, unsigned int r[],
f12505fb   Laila Saadatifard   upload the ivote ...
18
19
  			int iter, float t, unsigned int conn[], unsigned int x, unsigned int y, unsigned int z){
  
94d437dd   Laila Saadatifard   ivote3 code compi...
20
  	
a744d027   Laila Saadatifard   upload the ivote3...
21
  	cudaSetDevice(1);
1f55a874   Laila Saadatifard   upload the fixed ...
22
23
  	
  	unsigned int bytes = x * y * z * sizeof(float);				// compute the number of bytes in the input data
f12505fb   Laila Saadatifard   upload the ivote ...
24
  
1f55a874   Laila Saadatifard   upload the fixed ...
25
  	float* gpuI0;											//assign memory on gpu for the input data
f12505fb   Laila Saadatifard   upload the ivote ...
26
  	cudaMalloc(&gpuI0, bytes);	
1f55a874   Laila Saadatifard   upload the fixed ...
27
  	cudaMemcpy(gpuI0, img, bytes, cudaMemcpyHostToDevice);				//copy the image data to the GPU.
f12505fb   Laila Saadatifard   upload the ivote ...
28
  
1f55a874   Laila Saadatifard   upload the fixed ...
29
30
  	
  	gpu_gaussian_blur3<float>(gpuI0, sigma, x, y, z);							//call the blurring function from the gpu.
f12505fb   Laila Saadatifard   upload the ivote ...
31
32
  	cudaDeviceSynchronize();
  		
1f55a874   Laila Saadatifard   upload the fixed ...
33
  	float* gpu_grad;												//assign memory on the gpu for the gradient along the X, y, z.
c9aba18a   Laila Saadatifard   ivote3 on the GPU...
34
35
  	cudaMalloc(&gpu_grad, bytes*3);
  	
1f55a874   Laila Saadatifard   upload the fixed ...
36
  	gpu_gradient3<float>(gpu_grad, gpuI0, anisotropy, x, y, z);				//call the gradient function from the gpu.
f12505fb   Laila Saadatifard   upload the ivote ...
37
38
  	cudaFree(gpuI0);
  	
f12505fb   Laila Saadatifard   upload the ivote ...
39
40
  	float* gpu_vote;
  	cudaMalloc(&gpu_vote, bytes);
c9aba18a   Laila Saadatifard   ivote3 on the GPU...
41
  
f12505fb   Laila Saadatifard   upload the ivote ...
42
43
44
45
  	float cos_phi = cos(phi);
  
  	//call the vote function.
  	for (int i = 0; i < iter; i++){
a744d027   Laila Saadatifard   upload the ivote3...
46
47
  	
  		cudaMemset(gpu_vote, 0, bytes);
1f55a874   Laila Saadatifard   upload the fixed ...
48
  		gpu_vote3<float>(gpu_vote, gpu_grad, phi, cos_phi, r, x, y, z);
f12505fb   Laila Saadatifard   upload the ivote ...
49
  		cudaDeviceSynchronize();
a744d027   Laila Saadatifard   upload the ivote3...
50
51
  	
  		//if (phi >= d_phi){	
1f55a874   Laila Saadatifard   upload the fixed ...
52
  			gpu_update_dir3<float>(gpu_grad, gpu_vote, phi, cos_phi, r, x, y, z);
f12505fb   Laila Saadatifard   upload the ivote ...
53
54
55
  			cudaDeviceSynchronize();
  			phi = phi - d_phi;
  			cos_phi = cos(phi);
a744d027   Laila Saadatifard   upload the ivote3...
56
  		//}
f12505fb   Laila Saadatifard   upload the ivote ...
57
58
  	
  	}
c9aba18a   Laila Saadatifard   ivote3 on the GPU...
59
  
02fb26b3   Laila Saadatifard   change the vote a...
60
  	cudaFree(gpu_grad);	
46820b25   Laila Saadatifard   create an obj fil...
61
  	cudaMemcpy(img, gpu_vote, bytes, cudaMemcpyDeviceToHost);
f12505fb   Laila Saadatifard   upload the ivote ...
62
63
  
  	//allocate space on the gpu for the final detected cells.
46820b25   Laila Saadatifard   create an obj fil...
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
  	//float* gpu_output;
  	//cudaMalloc(&gpu_output, bytes);
  
  	////call the local max function
  	//gpu_local_max3<float>(gpu_output, gpu_vote, t, conn, x, y, z);
  
  	////copy the final result to the cpu.
  	//cudaMemcpy(center, gpu_output, bytes, cudaMemcpyDeviceToHost);
  	//	
  	//
  	cudaFree(gpu_vote);
  	//cudaFree(gpu_output);
  	
  }
  
  void lmax(float* out, float* in, float t, unsigned int conn[], unsigned int x, unsigned int y, unsigned int z){
  	unsigned int bytes = x * y * z * sizeof(float);
  
a744d027   Laila Saadatifard   upload the ivote3...
82
83
  	cudaSetDevice(1);
  
1f55a874   Laila Saadatifard   upload the fixed ...
84
85
  	
  	float* gpuV;					//assign memory on gpu for the input data.
46820b25   Laila Saadatifard   create an obj fil...
86
87
  	cudaMalloc(&gpuV, bytes);	
  
1f55a874   Laila Saadatifard   upload the fixed ...
88
  	cudaMemcpy(gpuV, in, bytes, cudaMemcpyHostToDevice);			//copy the image data to the GPU.
46820b25   Laila Saadatifard   create an obj fil...
89
90
91
  
  	float* gpuOut;
  	cudaMalloc(&gpuOut, bytes);
f12505fb   Laila Saadatifard   upload the ivote ...
92
  
1f55a874   Laila Saadatifard   upload the fixed ...
93
  	gpu_local_max3<float>(gpuOut, gpuV, t, conn, x, y, z);				//call the local max function
f12505fb   Laila Saadatifard   upload the ivote ...
94
  
1f55a874   Laila Saadatifard   upload the fixed ...
95
  	cudaMemcpy(out, gpuOut, bytes, cudaMemcpyDeviceToHost);				//copy the final result to the cpu.
f12505fb   Laila Saadatifard   upload the ivote ...
96
  	
46820b25   Laila Saadatifard   create an obj fil...
97
98
  	cudaFree(gpuV);
  	cudaFree(gpuOut);
a744d027   Laila Saadatifard   upload the ivote3...
99
  }