/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-hipcub/checkouts/docs-5.4.3/hipcub/include/hipcub/backend/cub/device/device_segmented_radix_sort.hpp Source File

• home
• docs
• checkouts
• readthedocs.org
• user_builds
• advanced-micro-devices-hipcub
• checkouts
• docs-5.4.3
• hipcub
• include
• hipcub
• backend
• cub
• device

device_segmented_radix_sort.hpp

1 /******************************************************************************

2  * Copyright (c) 2010-2011, Duane Merrill. All rights reserved.

3  * Copyright (c) 2011-2018, NVIDIA CORPORATION. All rights reserved.

4  * Modifications Copyright (c) 2017-2020, Advanced Micro Devices, Inc. All rights reserved.

5  *

6  * Redistribution and use in source and binary forms, with or without

7  * modification, are permitted provided that the following conditions are met:

8  * * Redistributions of source code must retain the above copyright

9  * notice, this list of conditions and the following disclaimer.

10  * * Redistributions in binary form must reproduce the above copyright

11  * notice, this list of conditions and the following disclaimer in the

12  * documentation and/or other materials provided with the distribution.

13  * * Neither the name of the NVIDIA CORPORATION nor the

14  * names of its contributors may be used to endorse or promote products

15  * derived from this software without specific prior written permission.

16  *

17  * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" AND

18  * ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED

19  * WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE ARE

20  * DISCLAIMED. IN NO EVENT SHALL NVIDIA CORPORATION BE LIABLE FOR ANY

21  * DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES

22  * (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES;

23  * LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND

24  * ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT

25  * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS

26  * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.

27  *

28  ******************************************************************************/

29 

30 #ifndef HIPCUB_CUB_DEVICE_DEVICE_SEGMENTED_RADIX_SORT_HPP_

31 #define HIPCUB_CUB_DEVICE_DEVICE_SEGMENTED_RADIX_SORT_HPP_

32 

33 #include "../../../config.hpp"

34 

35 #include <cub/device/device_segmented_radix_sort.cuh>

36 

37 BEGIN_HIPCUB_NAMESPACE

38 

39 struct DeviceSegmentedRadixSort

40 {

41  template<typename KeyT, typename ValueT, typename OffsetIteratorT>

42  HIPCUB_RUNTIME_FUNCTION static

43  hipError_t SortPairs(void * d_temp_storage,

44  size_t& temp_storage_bytes,

45  const KeyT * d_keys_in,

46  KeyT * d_keys_out,

47  const ValueT * d_values_in,

48  ValueT * d_values_out,

49  int num_items,

50  int num_segments,

51  OffsetIteratorT d_begin_offsets,

52  OffsetIteratorT d_end_offsets,

53  int begin_bit = 0,

54  int end_bit = sizeof(KeyT) * 8,

55  hipStream_t stream = 0,

56  bool debug_synchronous = false)

57  {

58  return hipCUDAErrorTohipError(

59  ::cub::DeviceSegmentedRadixSort::SortPairs(

60  d_temp_storage, temp_storage_bytes,

61  d_keys_in, d_keys_out,

62  d_values_in, d_values_out,

63  num_items, num_segments,

64  d_begin_offsets, d_end_offsets,

65  begin_bit, end_bit,

66  stream, debug_synchronous

67  )

68  );

69  }

70 

71  template<typename KeyT, typename ValueT, typename OffsetIteratorT>

72  HIPCUB_RUNTIME_FUNCTION static

73  hipError_t SortPairs(void * d_temp_storage,

74  size_t& temp_storage_bytes,

75  DoubleBuffer<KeyT>& d_keys,

76  DoubleBuffer<ValueT>& d_values,

77  int num_items,

78  int num_segments,

79  OffsetIteratorT d_begin_offsets,

80  OffsetIteratorT d_end_offsets,

81  int begin_bit = 0,

82  int end_bit = sizeof(KeyT) * 8,

83  hipStream_t stream = 0,

84  bool debug_synchronous = false)

85  {

86  return hipCUDAErrorTohipError(

87  ::cub::DeviceSegmentedRadixSort::SortPairs(

88  d_temp_storage, temp_storage_bytes,

89  d_keys, d_values,

90  num_items, num_segments,

91  d_begin_offsets, d_end_offsets,

92  begin_bit, end_bit,

93  stream, debug_synchronous

94  )

95  );

96  }

97 

98  template<typename KeyT, typename ValueT, typename OffsetIteratorT>

99  HIPCUB_RUNTIME_FUNCTION static

100  hipError_t SortPairsDescending(void * d_temp_storage,

101  size_t& temp_storage_bytes,

102  const KeyT * d_keys_in,

103  KeyT * d_keys_out,

104  const ValueT * d_values_in,

105  ValueT * d_values_out,

106  int num_items,

107  int num_segments,

108  OffsetIteratorT d_begin_offsets,

109  OffsetIteratorT d_end_offsets,

110  int begin_bit = 0,

111  int end_bit = sizeof(KeyT) * 8,

112  hipStream_t stream = 0,

113  bool debug_synchronous = false)

114  {

115  return hipCUDAErrorTohipError(

116  ::cub::DeviceSegmentedRadixSort::SortPairsDescending(

117  d_temp_storage, temp_storage_bytes,

118  d_keys_in, d_keys_out,

119  d_values_in, d_values_out,

120  num_items, num_segments,

121  d_begin_offsets, d_end_offsets,

122  begin_bit, end_bit,

123  stream, debug_synchronous

124  )

125  );

126  }

127 

128  template<typename KeyT, typename ValueT, typename OffsetIteratorT>

129  HIPCUB_RUNTIME_FUNCTION static

130  hipError_t SortPairsDescending(void * d_temp_storage,

131  size_t& temp_storage_bytes,

132  DoubleBuffer<KeyT>& d_keys,

133  DoubleBuffer<ValueT>& d_values,

134  int num_items,

135  int num_segments,

136  OffsetIteratorT d_begin_offsets,

137  OffsetIteratorT d_end_offsets,

138  int begin_bit = 0,

139  int end_bit = sizeof(KeyT) * 8,

140  hipStream_t stream = 0,

141  bool debug_synchronous = false)

142  {

143  return hipCUDAErrorTohipError(

144  ::cub::DeviceSegmentedRadixSort::SortPairsDescending(

145  d_temp_storage, temp_storage_bytes,

146  d_keys, d_values,

147  num_items, num_segments,

148  d_begin_offsets, d_end_offsets,

149  begin_bit, end_bit,

150  stream, debug_synchronous

151  )

152  );

153  }

154 

155  template<typename KeyT, typename OffsetIteratorT>

156  HIPCUB_RUNTIME_FUNCTION static

157  hipError_t SortKeys(void * d_temp_storage,

158  size_t& temp_storage_bytes,

159  const KeyT * d_keys_in,

160  KeyT * d_keys_out,

161  int num_items,

162  int num_segments,

163  OffsetIteratorT d_begin_offsets,

164  OffsetIteratorT d_end_offsets,

165  int begin_bit = 0,

166  int end_bit = sizeof(KeyT) * 8,

167  hipStream_t stream = 0,

168  bool debug_synchronous = false)

169  {

170  return hipCUDAErrorTohipError(

171  ::cub::DeviceSegmentedRadixSort::SortKeys(

172  d_temp_storage, temp_storage_bytes,

173  d_keys_in, d_keys_out,

174  num_items, num_segments,

175  d_begin_offsets, d_end_offsets,

176  begin_bit, end_bit,

177  stream, debug_synchronous

178  )

179  );

180  }

181 

182  template<typename KeyT, typename OffsetIteratorT>

183  HIPCUB_RUNTIME_FUNCTION static

184  hipError_t SortKeys(void * d_temp_storage,

185  size_t& temp_storage_bytes,

186  DoubleBuffer<KeyT>& d_keys,

187  int num_items,

188  int num_segments,

189  OffsetIteratorT d_begin_offsets,

190  OffsetIteratorT d_end_offsets,

191  int begin_bit = 0,

192  int end_bit = sizeof(KeyT) * 8,

193  hipStream_t stream = 0,

194  bool debug_synchronous = false)

195  {

196  return hipCUDAErrorTohipError(

197  ::cub::DeviceSegmentedRadixSort::SortKeys(

198  d_temp_storage, temp_storage_bytes,

199  d_keys,

200  num_items, num_segments,

201  d_begin_offsets, d_end_offsets,

202  begin_bit, end_bit,

203  stream, debug_synchronous

204  )

205  );

206  }

207 

208  template<typename KeyT, typename OffsetIteratorT>

209  HIPCUB_RUNTIME_FUNCTION static

210  hipError_t SortKeysDescending(void * d_temp_storage,

211  size_t& temp_storage_bytes,

212  const KeyT * d_keys_in,

213  KeyT * d_keys_out,

214  int num_items,

215  int num_segments,

216  OffsetIteratorT d_begin_offsets,

217  OffsetIteratorT d_end_offsets,

218  int begin_bit = 0,

219  int end_bit = sizeof(KeyT) * 8,

220  hipStream_t stream = 0,

221  bool debug_synchronous = false)

222  {

223  return hipCUDAErrorTohipError(

224  ::cub::DeviceSegmentedRadixSort::SortKeysDescending(

225  d_temp_storage, temp_storage_bytes,

226  d_keys_in, d_keys_out,

227  num_items, num_segments,

228  d_begin_offsets, d_end_offsets,

229  begin_bit, end_bit,

230  stream, debug_synchronous

231  )

232  );

233  }

234 

235  template<typename KeyT, typename OffsetIteratorT>

236  HIPCUB_RUNTIME_FUNCTION static

237  hipError_t SortKeysDescending(void * d_temp_storage,

238  size_t& temp_storage_bytes,

239  DoubleBuffer<KeyT>& d_keys,

240  int num_items,

241  int num_segments,

242  OffsetIteratorT d_begin_offsets,

243  OffsetIteratorT d_end_offsets,

244  int begin_bit = 0,

245  int end_bit = sizeof(KeyT) * 8,

246  hipStream_t stream = 0,

247  bool debug_synchronous = false)

248  {

249  return hipCUDAErrorTohipError(

250  ::cub::DeviceSegmentedRadixSort::SortKeysDescending(

251  d_temp_storage, temp_storage_bytes,

252  d_keys,

253  num_items, num_segments,

254  d_begin_offsets, d_end_offsets,

255  begin_bit, end_bit,

256  stream, debug_synchronous

257  )

258  );

259  }

260 };

261 

262 END_HIPCUB_NAMESPACE

263 

264 #endif // HIPCUB_CUB_DEVICE_DEVICE_SEGMENTED_RADIX_SORT_HPP_

hipcub::DeviceSegmentedRadixSort

Definition: device_segmented_radix_sort.hpp:40

hipcub::DoubleBuffer

Definition: util_type.hpp:108