-
Notifications
You must be signed in to change notification settings - Fork 0
/
cuda.h
112 lines (84 loc) · 2.1 KB
/
cuda.h
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
#ifndef CUDA_H
#define CUDA_H
#include <thrust/device_vector.h>
#include <vector>
#include <iostream>
#include <opencv2/opencv.hpp>
using namespace std;
bool loadImage(string fileName, cv::Mat & image);
thrust::host_vector<int> doHistogramGPU();
std::vector<int> doHistogramCPU();
//#define IS_LOGGING 1
typedef thrust::tuple<int, int, int> Int3;
struct BinFinder
{
//This kernel assigns each element to a bin group
__host__ __device__ Int3 operator()(const Int3 & param1, const Int3 & param2) const
{
//return x + y;
//if (x >= 0 && x <=5)
//{
// return 1;
//}
//else if (x >=6 && x <= 10)
//{
// return 2;
//}
//else if (x >= 11 && x <= 15)
//{
// return 3;
//}
//else
//{
// return 4;
//}
//return (x >= 0 && x <= 5) * 1 +
// (x >=6 && x <= 10) * 2 +
// (x >= 11 && x <= 15) * 3 +
// (x >= 16 && x <=20) * 4;
//cout << x << " ";
int x = thrust::get<0>(param1);
int y = thrust::get<1>(param1);
int z = thrust::get<2>(param1);
int bin1 = (x >= 0 && x <= 63) * 1 +
(x >=64 && x <= 127) * 2 +
(x >= 128 && x <= 191) * 3 +
(x >= 192) * 4;
int bin2 = (y >= 0 && y <= 63) * 1 +
(y >=64 && y <= 127) * 2 +
(y >= 128 && y <= 191) * 3 +
(y >= 192) * 4;
int bin3 = (z >= 0 && z <= 63) * 1 +
(z >=64 && z <= 127) * 2 +
(z >= 128 && z <= 191) * 3 +
(z >= 192) * 4;
return thrust::make_tuple(bin1, bin2, bin3);
}
};
struct IndexFinder
{
//This kernel looks at a bin element and assigns the counting index if it is not 0 and assigns -1 if it is 0
__host__ __device__ int operator()(const int & x, const int & y) const
{
//if (x != 0)
//{
// return y;
//}
//else
//{
// return -1;
//}
return (x != 0) * y + (x == 0) * -1;
}
};
struct ZipComparator
{
__host__ __device__
inline bool operator() (const Int3 & a, const Int3 & b)
{
return thrust::get<0>(a) < thrust::get<0>(b) ||
(thrust::get<0>(a) == thrust::get<0>(b) && thrust::get<1>(a) < thrust::get<1>(b)) ||
(thrust::get<0>(a) == thrust::get<0>(b) && thrust::get<1>(a) == thrust::get<1>(b) && thrust::get<2>(a) < thrust::get<2>(b));
}
};
#endif