1 #include <unittest/unittest.h>
2 #include <thrust/sort.h>
3 #include <thrust/functional.h>
4
5 template <typename T>
6 struct less_div_10
7 {
operator ()less_div_108 __host__ __device__ bool operator()(const T &lhs, const T &rhs) const {return ((int) lhs) / 10 < ((int) rhs) / 10;}
9 };
10
11 template <typename T>
12 struct greater_div_10
13 {
operator ()greater_div_1014 __host__ __device__ bool operator()(const T &lhs, const T &rhs) const {return ((int) lhs) / 10 > ((int) rhs) / 10;}
15 };
16
17
18 template <typename T, unsigned int N>
_TestStableSortByKeyWithLargeKeys(void)19 void _TestStableSortByKeyWithLargeKeys(void)
20 {
21 size_t n = (128 * 1024) / sizeof(FixedVector<T,N>);
22
23 thrust::host_vector< FixedVector<T,N> > h_keys(n);
24 thrust::host_vector< unsigned int > h_vals(n);
25
26 for(size_t i = 0; i < n; i++)
27 {
28 h_keys[i] = FixedVector<T,N>(rand());
29 h_vals[i] = i;
30 }
31
32 thrust::device_vector< FixedVector<T,N> > d_keys = h_keys;
33 thrust::device_vector< unsigned int > d_vals = h_vals;
34
35 thrust::stable_sort_by_key(h_keys.begin(), h_keys.end(), h_vals.begin());
36 thrust::stable_sort_by_key(d_keys.begin(), d_keys.end(), d_vals.begin());
37
38 ASSERT_EQUAL_QUIET(h_keys, d_keys);
39 ASSERT_EQUAL_QUIET(h_vals, d_vals);
40 }
41
TestStableSortByKeyWithLargeKeys(void)42 void TestStableSortByKeyWithLargeKeys(void)
43 {
44 _TestStableSortByKeyWithLargeKeys<int, 4>();
45 _TestStableSortByKeyWithLargeKeys<int, 8>();
46 _TestStableSortByKeyWithLargeKeys<int, 16>();
47
48 // XXX these take too long to compile
49 // _TestStableSortByKeyWithLargeKeys<int, 32>();
50 // _TestStableSortByKeyWithLargeKeys<int, 64>();
51 // _TestStableSortByKeyWithLargeKeys<int, 128>();
52 // _TestStableSortByKeyWithLargeKeys<int, 256>();
53 // _TestStableSortByKeyWithLargeKeys<int, 512>();
54 // _TestStableSortByKeyWithLargeKeys<int, 1024>();
55 // _TestStableSortByKeyWithLargeKeys<int, 2048>();
56 // _TestStableSortByKeyWithLargeKeys<int, 4096>();
57 // _TestStableSortByKeyWithLargeKeys<int, 8192>();
58 }
59 DECLARE_UNITTEST(TestStableSortByKeyWithLargeKeys);
60
61
62 template <typename T, unsigned int N>
_TestStableSortByKeyWithLargeValues(void)63 void _TestStableSortByKeyWithLargeValues(void)
64 {
65 size_t n = (128 * 1024) / sizeof(FixedVector<T,N>);
66
67 thrust::host_vector< unsigned int > h_keys(n);
68 thrust::host_vector< FixedVector<T,N> > h_vals(n);
69
70 for(size_t i = 0; i < n; i++)
71 {
72 h_keys[i] = rand();
73 h_vals[i] = FixedVector<T,N>(i);
74 }
75
76 thrust::device_vector< unsigned int > d_keys = h_keys;
77 thrust::device_vector< FixedVector<T,N> > d_vals = h_vals;
78
79 thrust::stable_sort_by_key(h_keys.begin(), h_keys.end(), h_vals.begin());
80 thrust::stable_sort_by_key(d_keys.begin(), d_keys.end(), d_vals.begin());
81
82 ASSERT_EQUAL_QUIET(h_keys, d_keys);
83 ASSERT_EQUAL_QUIET(h_vals, d_vals);
84
85 // so cuda::stable_merge_sort_by_key() is called
86 thrust::stable_sort_by_key(h_keys.begin(), h_keys.end(), h_vals.begin(), greater_div_10<unsigned int>());
87 thrust::stable_sort_by_key(d_keys.begin(), d_keys.end(), d_vals.begin(), greater_div_10<unsigned int>());
88
89 ASSERT_EQUAL_QUIET(h_keys, d_keys);
90 ASSERT_EQUAL_QUIET(h_vals, d_vals);
91 }
92
TestStableSortByKeyWithLargeValues(void)93 void TestStableSortByKeyWithLargeValues(void)
94 {
95 _TestStableSortByKeyWithLargeValues<int, 4>();
96 _TestStableSortByKeyWithLargeValues<int, 8>();
97 _TestStableSortByKeyWithLargeValues<int, 16>();
98
99 // XXX these take too long to compile
100 // _TestStableSortByKeyWithLargeValues<int, 32>();
101 // _TestStableSortByKeyWithLargeValues<int, 64>();
102 // _TestStableSortByKeyWithLargeValues<int, 128>();
103 // _TestStableSortByKeyWithLargeValues<int, 256>();
104 // _TestStableSortByKeyWithLargeValues<int, 512>();
105 // _TestStableSortByKeyWithLargeValues<int, 1024>();
106 // _TestStableSortByKeyWithLargeValues<int, 2048>();
107 // _TestStableSortByKeyWithLargeValues<int, 4096>();
108 // _TestStableSortByKeyWithLargeValues<int, 8192>();
109 }
110 DECLARE_UNITTEST(TestStableSortByKeyWithLargeValues);
111
112
113 template <typename T, unsigned int N>
_TestStableSortByKeyWithLargeKeysAndValues(void)114 void _TestStableSortByKeyWithLargeKeysAndValues(void)
115 {
116 size_t n = (128 * 1024) / sizeof(FixedVector<T,N>);
117
118 thrust::host_vector< FixedVector<T,N> > h_keys(n);
119 thrust::host_vector< FixedVector<T,N> > h_vals(n);
120
121 for(size_t i = 0; i < n; i++)
122 {
123 h_keys[i] = FixedVector<T,N>(rand());
124 h_vals[i] = FixedVector<T,N>(i);
125 }
126
127 thrust::device_vector< FixedVector<T,N> > d_keys = h_keys;
128 thrust::device_vector< FixedVector<T,N> > d_vals = h_vals;
129
130 thrust::stable_sort_by_key(h_keys.begin(), h_keys.end(), h_vals.begin());
131 thrust::stable_sort_by_key(d_keys.begin(), d_keys.end(), d_vals.begin());
132
133 ASSERT_EQUAL_QUIET(h_keys, d_keys);
134 ASSERT_EQUAL_QUIET(h_vals, d_vals);
135 }
136
TestStableSortByKeyWithLargeKeysAndValues(void)137 void TestStableSortByKeyWithLargeKeysAndValues(void)
138 {
139 _TestStableSortByKeyWithLargeKeysAndValues<int, 4>();
140 _TestStableSortByKeyWithLargeKeysAndValues<int, 8>();
141 _TestStableSortByKeyWithLargeKeysAndValues<int, 16>();
142
143 // XXX these take too long to compile
144 // _TestStableSortByKeyWithLargeKeysAndValues<int, 32>();
145 // _TestStableSortByKeyWithLargeKeysAndValues<int, 64>();
146 // _TestStableSortByKeyWithLargeKeysAndValues<int, 128>();
147 // _TestStableSortByKeyWithLargeKeysAndValues<int, 256>();
148 // _TestStableSortByKeyWithLargeKeysAndValues<int, 512>();
149 // _TestStableSortByKeyWithLargeKeysAndValues<int, 1024>();
150 // _TestStableSortByKeyWithLargeKeysAndValues<int, 2048>();
151 // _TestStableSortByKeyWithLargeKeysAndValues<int, 4096>();
152 // _TestStableSortByKeyWithLargeKeysAndValues<int, 8192>();
153 }
154 DECLARE_UNITTEST(TestStableSortByKeyWithLargeKeysAndValues);
155
156