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"
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
13
14
|
#include "update_dir3.cuh"
#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);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
22
|
// compute the number of bytes in the input data
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
23
|
unsigned int bytes = x * y * z * sizeof(float);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
24
|
|
a744d027
Laila Saadatifard
upload the ivote3...
|
25
|
//assign memory on gpu for the input data.z
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
26
|
float* gpuI0;
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
27
28
29
|
cudaMalloc(&gpuI0, bytes);
//copy the image data to the GPU.
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
30
31
|
cudaMemcpy(gpuI0, img, bytes, cudaMemcpyHostToDevice);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
32
|
//call the blurring function from the gpu.
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
33
|
gpu_gaussian_blur3<float>(gpuI0, sigma, x, y, z);
|
94d437dd
Laila Saadatifard
ivote3 code compi...
|
34
|
//cudaMemcpy(img, gpuI0, bytes, cudaMemcpyDeviceToHost);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
35
36
|
cudaDeviceSynchronize();
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
37
|
//assign memory on the gpu for the gradient along the X, y, z.
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
38
39
40
|
float* gpu_grad;
cudaMalloc(&gpu_grad, bytes*3);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
41
|
//call the gradient function from the gpu.
|
94d437dd
Laila Saadatifard
ivote3 code compi...
|
42
|
gpu_gradient3<float>(gpu_grad, gpuI0, anisotropy, x, y, z);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
43
44
|
cudaFree(gpuI0);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
45
46
|
float* gpu_vote;
cudaMalloc(&gpu_vote, bytes);
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
47
|
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
48
49
50
51
|
float cos_phi = cos(phi);
//call the vote function.
for (int i = 0; i < iter; i++){
|
a744d027
Laila Saadatifard
upload the ivote3...
|
52
53
|
cudaMemset(gpu_vote, 0, bytes);
|
02fb26b3
Laila Saadatifard
change the vote a...
|
54
|
gpu_vote3<float>(gpu_vote, gpu_grad, cos_phi, r, x, y, z);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
55
|
cudaDeviceSynchronize();
|
a744d027
Laila Saadatifard
upload the ivote3...
|
56
57
|
//if (phi >= d_phi){
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
58
59
60
61
|
gpu_update_dir3<float>(gpu_grad, gpu_vote, cos_phi, r, x, y, z);
cudaDeviceSynchronize();
phi = phi - d_phi;
cos_phi = cos(phi);
|
a744d027
Laila Saadatifard
upload the ivote3...
|
62
|
//}
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
63
64
|
}
|
c9aba18a
Laila Saadatifard
ivote3 on the GPU...
|
65
|
|
02fb26b3
Laila Saadatifard
change the vote a...
|
66
|
cudaFree(gpu_grad);
|
46820b25
Laila Saadatifard
create an obj fil...
|
67
|
cudaMemcpy(img, gpu_vote, bytes, cudaMemcpyDeviceToHost);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
68
69
|
//allocate space on the gpu for the final detected cells.
|
46820b25
Laila Saadatifard
create an obj fil...
|
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
|
//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...
|
88
89
|
cudaSetDevice(1);
|
46820b25
Laila Saadatifard
create an obj fil...
|
90
91
92
93
94
95
96
97
98
|
//assign memory on gpu for the input data.
float* gpuV;
cudaMalloc(&gpuV, bytes);
//copy the image data to the GPU.
cudaMemcpy(gpuV, in, bytes, cudaMemcpyHostToDevice);
float* gpuOut;
cudaMalloc(&gpuOut, bytes);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
99
100
|
//call the local max function
|
46820b25
Laila Saadatifard
create an obj fil...
|
101
|
gpu_local_max3<float>(gpuOut, gpuV, t, conn, x, y, z);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
102
103
|
//copy the final result to the cpu.
|
46820b25
Laila Saadatifard
create an obj fil...
|
104
|
cudaMemcpy(out, gpuOut, bytes, cudaMemcpyDeviceToHost);
|
f12505fb
Laila Saadatifard
upload the ivote ...
|
105
|
|
46820b25
Laila Saadatifard
create an obj fil...
|
106
107
|
cudaFree(gpuV);
cudaFree(gpuOut);
|
a744d027
Laila Saadatifard
upload the ivote3...
|
108
|
}
|