NVBIO
 All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Macros Groups Pages
primitives.h
Go to the documentation of this file.
1 /*
2  * nvbio
3  * Copyright (c) 2011-2014, NVIDIA CORPORATION. All rights reserved.
4  *
5  * Redistribution and use in source and binary forms, with or without
6  * modification, are permitted provided that the following conditions are met:
7  * * Redistributions of source code must retain the above copyright
8  * notice, this list of conditions and the following disclaimer.
9  * * Redistributions in binary form must reproduce the above copyright
10  * notice, this list of conditions and the following disclaimer in the
11  * documentation and/or other materials provided with the distribution.
12  * * Neither the name of the NVIDIA CORPORATION nor the
13  * names of its contributors may be used to endorse or promote products
14  * derived from this software without specific prior written permission.
15  *
16  * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" AND
17  * ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED
18  * WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE ARE
19  * DISCLAIMED. IN NO EVENT SHALL NVIDIA CORPORATION BE LIABLE FOR ANY
20  * DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES
21  * (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES;
22  * LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND
23  * ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT
24  * (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
25  * SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
26  */
27 
28 #pragma once
29 
30 #include <nvbio/basic/types.h>
31 #include <nvbio/basic/numbers.h>
32 #include <nvbio/basic/console.h>
33 #include <nvbio/basic/vector.h>
34 #include <nvbio/basic/algorithms.h>
35 #include <nvbio/basic/cuda/timer.h>
36 #include <thrust/reduce.h>
37 #include <thrust/scan.h>
38 #include <thrust/copy.h>
39 #include <thrust/sort.h>
40 #include <thrust/binary_search.h>
41 #include <thrust/merge.h>
42 #include <thrust/iterator/constant_iterator.h>
43 
44 #if defined(__CUDACC__)
45 #include <nvbio/basic/cuda/primitives.h>
46 #include <nvbio/basic/cuda/sort.h>
47 #endif
48 
49 #if defined (_OPENMP)
50 #include <omp.h>
51 #endif
52 
81 
82 namespace nvbio {
83 
86 
93 
96 template <typename system_tag, typename PredicateIterator>
97 bool any(
98  const uint32 n,
99  const PredicateIterator pred);
100 
103 template <typename system_tag, typename PredicateIterator>
104 bool all(
105  const uint32 n,
106  const PredicateIterator pred);
107 
110 template <typename system_tag, typename Iterator>
111 bool is_sorted(
112  const uint32 n,
113  const Iterator values);
114 
118 template <typename system_tag, typename Iterator, typename Headflags>
119 bool is_segment_sorted(
120  const uint32 n,
121  const Iterator values,
122  const Headflags flags);
123 
124 
127 template <typename system_tag, typename Iterator, typename Functor>
128 void for_each(
129  const uint64 n,
130  const Iterator in,
131  Functor functor);
132 
135 template <typename system_tag, typename Iterator, typename Output, typename Functor>
136 void transform(
137  const uint32 n,
138  const Iterator in,
139  const Output out,
140  const Functor functor);
141 
144 template <typename system_tag, typename Iterator1, typename Iterator2, typename Output, typename Functor>
145 void transform(
146  const uint32 n,
147  const Iterator1 in1,
148  const Iterator2 in2,
149  const Output out,
150  const Functor functor);
151 
159 template <typename system_tag, typename InputIterator, typename BinaryOp>
160 typename std::iterator_traits<InputIterator>::value_type reduce(
161  const uint32 n,
162  InputIterator in,
163  BinaryOp op,
164  nvbio::vector<system_tag,uint8>& temp_storage);
165 
174 template <typename system_tag, typename InputIterator, typename OutputIterator, typename BinaryOp>
175 void inclusive_scan(
176  const uint32 n,
177  InputIterator in,
178  OutputIterator out,
179  BinaryOp op,
180  nvbio::vector<system_tag,uint8>& temp_storage);
181 
191 template <typename system_tag, typename InputIterator, typename OutputIterator, typename BinaryOp, typename Identity>
192 void exclusive_scan(
193  const uint32 n,
194  InputIterator in,
195  OutputIterator out,
196  BinaryOp op,
197  Identity identity,
198  nvbio::vector<system_tag,uint8>& temp_storage);
199 
210 template <typename system_tag, typename InputIterator, typename FlagsIterator, typename OutputIterator>
211 uint32 copy_flagged(
212  const uint32 n,
213  InputIterator in,
214  FlagsIterator flags,
215  OutputIterator out,
216  nvbio::vector<system_tag,uint8>& temp_storage);
217 
228 template <typename system_tag, typename InputIterator, typename OutputIterator, typename Predicate>
229 uint32 copy_if(
230  const uint32 n,
231  InputIterator in,
232  OutputIterator out,
233  const Predicate pred,
234  nvbio::vector<system_tag,uint8>& temp_storage);
235 
246 template <typename system_tag, typename InputIterator, typename OutputIterator, typename CountIterator>
247 uint32 runlength_encode(
248  const uint32 n,
249  InputIterator in,
250  OutputIterator out,
251  CountIterator counts,
252  nvbio::vector<system_tag,uint8>& temp_storage);
253 
266 template <typename system_tag, typename KeyIterator, typename ValueIterator, typename OutputKeyIterator, typename OutputValueIterator, typename ReductionOp>
267 uint32 reduce_by_key(
268  const uint32 n,
269  KeyIterator keys_in,
270  ValueIterator values_in,
271  OutputKeyIterator keys_out,
272  OutputValueIterator values_out,
273  ReductionOp reduction_op,
274  nvbio::vector<system_tag,uint8>& temp_storage);
275 
284 template <typename system_tag, typename KeyIterator, typename ValueIterator, typename OutputIterator>
285 void lower_bound(
286  const uint32 n,
287  ValueIterator values,
288  const uint32 n_keys,
289  KeyIterator keys,
290  OutputIterator indices);
291 
300 template <typename system_tag, typename KeyIterator, typename ValueIterator, typename OutputIterator>
301 void upper_bound(
302  const uint32 n,
303  ValueIterator values,
304  const uint32 n_keys,
305  KeyIterator keys,
306  OutputIterator indices);
307 
314 template <typename system_tag, typename KeyIterator>
315 void radix_sort(
316  const uint32 n,
317  KeyIterator keys,
318  nvbio::vector<system_tag,uint8>& temp_storage);
319 
327 template <typename system_tag, typename KeyIterator, typename ValueIterator>
328 void radix_sort(
329  const uint32 n,
330  KeyIterator keys,
331  ValueIterator values,
332  nvbio::vector<system_tag,uint8>& temp_storage);
333 
346 template <
347  typename system_tag,
348  typename key_iterator1,
349  typename key_iterator2,
350  typename value_iterator1,
351  typename value_iterator2,
352  typename key_output,
353  typename value_output>
354 void merge_by_key(
355  const uint32 A_len,
356  const uint32 B_len,
357  const key_iterator1 A_keys,
358  const key_iterator2 B_keys,
359  const value_iterator1 A_values,
360  const value_iterator2 B_values,
361  key_output C_keys,
362  value_output C_values,
363  nvbio::vector<system_tag,uint8>& temp_storage);
364 
368 template <typename system_tag>
369 struct for_each_enactor
370 {
373  template <typename Iterator, typename Functor>
374  void operator () (
375  const uint64 n,
376  const Iterator in,
377  Functor functor)
378  {
379  for_each<system_tag>( n, in, functor );
380  }
381 
384  template <typename Functor>
385  void operator () (
386  const uint64 n,
387  Functor functor)
388  {
389  for_each<system_tag>( n, thrust::make_counting_iterator<uint64>(0), functor );
390  }
391 };
392 
396 template <>
397 struct for_each_enactor<device_tag>
398 {
401  for_each_enactor() :
402  m_blocks_lo( 0 ),
403  m_blocks_hi( 0 ),
404  m_speed_lo( 0.0f ),
405  m_speed_hi( 0.0f ) {}
406 
409  template <typename Iterator, typename Functor>
410  void operator () (
411  const uint64 n,
412  const Iterator in,
413  Functor functor);
414 
417  template <typename Functor>
418  void operator () (
419  const uint64 n,
420  Functor functor)
421  {
422  this->operator()( n, thrust::make_counting_iterator<uint64>(0), functor );
423  }
424 
425  private:
426 
429  template <typename KernelFunction>
430  uint32 suggested_blocks(KernelFunction kernel, const uint32 cta_size) const;
431 
434  void update(const uint32 n_blocks, const float speed);
435 
436  uint32 m_blocks_lo;
437  uint32 m_blocks_hi;
438  float m_speed_lo;
439  float m_speed_hi;
440 };
441 
444 
445 } // namespace nvbio
446 
447 #include <nvbio/basic/primitives_inl.h>
Generated on Wed Feb 25 2015 08:32:54 for NVBIO by   doxygen 1.8.4