/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-hipcub/checkouts/docs-5.7.0/hipcub/include/hipcub/backend/rocprim/block/block_store_func.hpp Source File

/home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-hipcub/checkouts/docs-5.7.0/hipcub/include/hipcub/backend/rocprim/block/block_store_func.hpp Source File#

hipCUB: /home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-hipcub/checkouts/docs-5.7.0/hipcub/include/hipcub/backend/rocprim/block/block_store_func.hpp Source File
block_store_func.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_ROCPRIM_BLOCK_BLOCK_STORE_FUNC_HPP_
31 #define HIPCUB_ROCPRIM_BLOCK_BLOCK_STORE_FUNC_HPP_
32 
33 #include "../../../config.hpp"
34 
35 #include <rocprim/block/block_store_func.hpp>
36 
37 BEGIN_HIPCUB_NAMESPACE
38 
39 template<
40  typename T,
41  int ITEMS_PER_THREAD,
42  typename OutputIteratorT
43 >
44 HIPCUB_DEVICE inline
45 void StoreDirectBlocked(int linear_id,
46  OutputIteratorT block_iter,
47  T (&items)[ITEMS_PER_THREAD])
48 {
49  ::rocprim::block_store_direct_blocked(
50  linear_id, block_iter, items
51  );
52 }
53 
54 template<
55  typename T,
56  int ITEMS_PER_THREAD,
57  typename OutputIteratorT
58 >
59 HIPCUB_DEVICE inline
60 void StoreDirectBlocked(int linear_id,
61  OutputIteratorT block_iter,
62  T (&items)[ITEMS_PER_THREAD],
63  int valid_items)
64 {
65  ::rocprim::block_store_direct_blocked(
66  linear_id, block_iter, items, valid_items
67  );
68 }
69 
70 template <
71  typename T,
72  int ITEMS_PER_THREAD
73 >
74 HIPCUB_DEVICE inline
75 void StoreDirectBlockedVectorized(int linear_id,
76  T* block_iter,
77  T (&items)[ITEMS_PER_THREAD])
78 {
79  ::rocprim::block_store_direct_blocked_vectorized(
80  linear_id, block_iter, items
81  );
82 }
83 
84 template<
85  int BLOCK_THREADS,
86  typename T,
87  int ITEMS_PER_THREAD,
88  typename OutputIteratorT
89 >
90 HIPCUB_DEVICE inline
91 void StoreDirectStriped(int linear_id,
92  OutputIteratorT block_iter,
93  T (&items)[ITEMS_PER_THREAD])
94 {
95  ::rocprim::block_store_direct_striped<BLOCK_THREADS>(
96  linear_id, block_iter, items
97  );
98 }
99 
100 template<
101  int BLOCK_THREADS,
102  typename T,
103  int ITEMS_PER_THREAD,
104  typename OutputIteratorT
105 >
106 HIPCUB_DEVICE inline
107 void StoreDirectStriped(int linear_id,
108  OutputIteratorT block_iter,
109  T (&items)[ITEMS_PER_THREAD],
110  int valid_items)
111 {
112  ::rocprim::block_store_direct_striped<BLOCK_THREADS>(
113  linear_id, block_iter, items, valid_items
114  );
115 }
116 
117 template<
118  typename T,
119  int ITEMS_PER_THREAD,
120  typename OutputIteratorT
121 >
122 HIPCUB_DEVICE inline
123 void StoreDirectWarpStriped(int linear_id,
124  OutputIteratorT block_iter,
125  T (&items)[ITEMS_PER_THREAD])
126 {
127  ::rocprim::block_store_direct_warp_striped(
128  linear_id, block_iter, items
129  );
130 }
131 
132 template<
133  typename T,
134  int ITEMS_PER_THREAD,
135  typename OutputIteratorT
136 >
137 HIPCUB_DEVICE inline
138 void StoreDirectWarpStriped(int linear_id,
139  OutputIteratorT block_iter,
140  T (&items)[ITEMS_PER_THREAD],
141  int valid_items)
142 {
143  ::rocprim::block_store_direct_warp_striped(
144  linear_id, block_iter, items, valid_items
145  );
146 }
147 
148 END_HIPCUB_NAMESPACE
149 
150 #endif // HIPCUB_ROCPRIM_BLOCK_BLOCK_STORE_FUNC_HPP_