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

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

device_scan.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_SCAN_HPP_

31 #define HIPCUB_CUB_DEVICE_DEVICE_SCAN_HPP_

32 

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

34 

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

36 

37 BEGIN_HIPCUB_NAMESPACE

38 

39 class DeviceScan

40 {

41 public:

42  template <

43  typename InputIteratorT,

44  typename OutputIteratorT

45  >

46  HIPCUB_RUNTIME_FUNCTION static

47  hipError_t InclusiveSum(void *d_temp_storage,

48  size_t &temp_storage_bytes,

49  InputIteratorT d_in,

50  OutputIteratorT d_out,

51  int num_items,

52  hipStream_t stream = 0,

53  bool debug_synchronous = false)

54  {

55  return hipCUDAErrorTohipError(

56  ::cub::DeviceScan::InclusiveSum(

57  d_temp_storage, temp_storage_bytes,

58  d_in, d_out, num_items,

59  stream, debug_synchronous

60  )

61  );

62  }

63 

64  template <

65  typename InputIteratorT,

66  typename OutputIteratorT,

67  typename ScanOpT

68  >

69  HIPCUB_RUNTIME_FUNCTION static

70  hipError_t InclusiveScan(void *d_temp_storage,

71  size_t &temp_storage_bytes,

72  InputIteratorT d_in,

73  OutputIteratorT d_out,

74  ScanOpT scan_op,

75  int num_items,

76  hipStream_t stream = 0,

77  bool debug_synchronous = false)

78  {

79  return hipCUDAErrorTohipError(

80  ::cub::DeviceScan::InclusiveScan(

81  d_temp_storage, temp_storage_bytes,

82  d_in, d_out, scan_op, num_items,

83  stream, debug_synchronous

84  )

85  );

86  }

87 

88  template <

89  typename InputIteratorT,

90  typename OutputIteratorT

91  >

92  HIPCUB_RUNTIME_FUNCTION static

93  hipError_t ExclusiveSum(void *d_temp_storage,

94  size_t &temp_storage_bytes,

95  InputIteratorT d_in,

96  OutputIteratorT d_out,

97  int num_items,

98  hipStream_t stream = 0,

99  bool debug_synchronous = false)

100  {

101  return hipCUDAErrorTohipError(

102  ::cub::DeviceScan::ExclusiveSum(

103  d_temp_storage, temp_storage_bytes,

104  d_in, d_out, num_items,

105  stream, debug_synchronous

106  )

107  );

108  }

109 

110  template <

111  typename InputIteratorT,

112  typename OutputIteratorT,

113  typename ScanOpT,

114  typename InitValueT

115  >

116  HIPCUB_RUNTIME_FUNCTION static

117  hipError_t ExclusiveScan(void *d_temp_storage,

118  size_t &temp_storage_bytes,

119  InputIteratorT d_in,

120  OutputIteratorT d_out,

121  ScanOpT scan_op,

122  InitValueT init_value,

123  int num_items,

124  hipStream_t stream = 0,

125  bool debug_synchronous = false)

126  {

127  return hipCUDAErrorTohipError(

128  ::cub::DeviceScan::ExclusiveScan(

129  d_temp_storage, temp_storage_bytes,

130  d_in, d_out, scan_op, init_value, num_items,

131  stream, debug_synchronous

132  )

133  );

134  }

135 

136  template <

137  typename InputIteratorT,

138  typename OutputIteratorT,

139  typename ScanOpT,

140  typename InitValueT,

141  typename InitValueIterT = InitValueT*

142  >

143  HIPCUB_RUNTIME_FUNCTION static

144  hipError_t ExclusiveScan(void *d_temp_storage,

145  size_t &temp_storage_bytes,

146  InputIteratorT d_in,

147  OutputIteratorT d_out,

148  ScanOpT scan_op,

149  FutureValue<InitValueT, InitValueIterT> init_value,

150  int num_items,

151  hipStream_t stream = 0,

152  bool debug_synchronous = false)

153  {

154  return hipCUDAErrorTohipError(

155  ::cub::DeviceScan::ExclusiveScan(

156  d_temp_storage, temp_storage_bytes,

157  d_in, d_out, scan_op, init_value, num_items,

158  stream, debug_synchronous

159  )

160  );

161  }

162 

163  template <

164  typename KeysInputIteratorT,

165  typename ValuesInputIteratorT,

166  typename ValuesOutputIteratorT,

167  typename EqualityOpT = ::hipcub::Equality

168  >

169  HIPCUB_RUNTIME_FUNCTION static

170  hipError_t ExclusiveSumByKey(void *d_temp_storage,

171  size_t &temp_storage_bytes,

172  KeysInputIteratorT d_keys_in,

173  ValuesInputIteratorT d_values_in,

174  ValuesOutputIteratorT d_values_out,

175  int num_items,

176  EqualityOpT equality_op = EqualityOpT(),

177  hipStream_t stream = 0,

178  bool debug_synchronous = false)

179  {

180  return hipCUDAErrorTohipError(

181  ::cub::DeviceScan::ExclusiveSumByKey(

182  d_temp_storage, temp_storage_bytes,

183  d_keys_in, d_values_in, d_values_out, num_items,

184  equality_op, stream, debug_synchronous

185  )

186  );

187  }

188 

189  template <

190  typename KeysInputIteratorT,

191  typename ValuesInputIteratorT,

192  typename ValuesOutputIteratorT,

193  typename ScanOpT,

194  typename InitValueT,

195  typename EqualityOpT = ::hipcub::Equality

196  >

197  HIPCUB_RUNTIME_FUNCTION static

198  hipError_t ExclusiveScanByKey(void *d_temp_storage,

199  size_t &temp_storage_bytes,

200  KeysInputIteratorT d_keys_in,

201  ValuesInputIteratorT d_values_in,

202  ValuesOutputIteratorT d_values_out,

203  ScanOpT scan_op,

204  InitValueT init_value,

205  int num_items,

206  EqualityOpT equality_op = EqualityOpT(),

207  hipStream_t stream = 0,

208  bool debug_synchronous = false)

209  {

210  return hipCUDAErrorTohipError(

211  ::cub::DeviceScan::ExclusiveScanByKey(

212  d_temp_storage, temp_storage_bytes,

213  d_keys_in, d_values_in, d_values_out, scan_op, init_value,

214  num_items, equality_op, stream, debug_synchronous

215  )

216  );

217  }

218 

219  template <

220  typename KeysInputIteratorT,

221  typename ValuesInputIteratorT,

222  typename ValuesOutputIteratorT,

223  typename EqualityOpT = ::hipcub::Equality

224  >

225  HIPCUB_RUNTIME_FUNCTION static

226  hipError_t InclusiveSumByKey(void *d_temp_storage,

227  size_t &temp_storage_bytes,

228  KeysInputIteratorT d_keys_in,

229  ValuesInputIteratorT d_values_in,

230  ValuesOutputIteratorT d_values_out,

231  int num_items,

232  EqualityOpT equality_op = EqualityOpT(),

233  hipStream_t stream = 0,

234  bool debug_synchronous = false)

235  {

236  return hipCUDAErrorTohipError(

237  ::cub::DeviceScan::InclusiveSumByKey(

238  d_temp_storage, temp_storage_bytes,

239  d_keys_in, d_values_in, d_values_out, num_items,

240  equality_op, stream, debug_synchronous

241  )

242  );

243  }

244 

245  template <

246  typename KeysInputIteratorT,

247  typename ValuesInputIteratorT,

248  typename ValuesOutputIteratorT,

249  typename ScanOpT,

250  typename EqualityOpT = ::hipcub::Equality

251  >

252  HIPCUB_RUNTIME_FUNCTION static

253  hipError_t InclusiveScanByKey(void *d_temp_storage,

254  size_t &temp_storage_bytes,

255  KeysInputIteratorT d_keys_in,

256  ValuesInputIteratorT d_values_in,

257  ValuesOutputIteratorT d_values_out,

258  ScanOpT scan_op,

259  int num_items,

260  EqualityOpT equality_op = EqualityOpT(),

261  hipStream_t stream = 0,

262  bool debug_synchronous = false)

263  {

264  return hipCUDAErrorTohipError(

265  ::cub::DeviceScan::InclusiveScanByKey(

266  d_temp_storage, temp_storage_bytes,

267  d_keys_in, d_values_in, d_values_out, scan_op,

268  num_items, equality_op, stream, debug_synchronous

269  )

270  );

271  }

272 };

273 

274 END_HIPCUB_NAMESPACE

275 

276 #endif // HIPCUB_CUB_DEVICE_DEVICE_SCAN_HPP_

hipcub::DeviceScan

Definition: device_scan.hpp:40

hipcub::Equality

Definition: thread_operators.hpp:40