ceph/src/boost/boost/compute/algorithm/for_each.hpp

   1 //---------------------------------------------------------------------------//
   2 // Copyright (c) 2013 Kyle Lutz <kyle.r.lutz@gmail.com>
   3 //
   4 // Distributed under the Boost Software License, Version 1.0
   5 // See accompanying file LICENSE_1_0.txt or copy at
   6 // http://www.boost.org/LICENSE_1_0.txt
   7 //
   8 // See http://boostorg.github.com/compute for more information.
   9 //---------------------------------------------------------------------------//
  10
  11 #ifndef BOOST_COMPUTE_ALGORITHM_FOR_EACH_HPP
  12 #define BOOST_COMPUTE_ALGORITHM_FOR_EACH_HPP
  13
  14 #include <boost/static_assert.hpp>
  15
  16 #include <boost/compute/system.hpp>
  17 #include <boost/compute/command_queue.hpp>
  18 #include <boost/compute/detail/meta_kernel.hpp>
  19 #include <boost/compute/detail/iterator_range_size.hpp>
  20 #include <boost/compute/type_traits/is_device_iterator.hpp>
  21
  22 namespace boost {
  23 namespace compute {
  24 namespace detail {
  25
  26 template<class InputIterator, class Function>
  27 struct for_each_kernel : public meta_kernel
  28 {
  29     for_each_kernel(InputIterator first, InputIterator last, Function function)
  30         : meta_kernel("for_each")
  31     {
  32         // store range size
  33         m_count = detail::iterator_range_size(first, last);
  34
  35         // setup kernel source
  36         *this << function(first[get_global_id(0)]) << ";\n";
  37     }
  38
  39     void exec(command_queue &queue)
  40     {
  41         exec_1d(queue, 0, m_count);
  42     }
  43
  44     size_t m_count;
  45 };
  46
  47 } // end detail namespace
  48
  49 /// Calls \p function on each element in the range [\p first, \p last).
  50 ///
  51 /// Space complexity: \Omega(1)
  52 ///
  53 /// \see transform()
  54 template<class InputIterator, class UnaryFunction>
  55 inline UnaryFunction for_each(InputIterator first,
  56                               InputIterator last,
  57                               UnaryFunction function,
  58                               command_queue &queue = system::default_queue())
  59 {
  60     BOOST_STATIC_ASSERT(is_device_iterator<InputIterator>::value);
  61
  62     detail::for_each_kernel<InputIterator, UnaryFunction> kernel(first, last, function);
  63
  64     kernel.exec(queue);
  65
  66     return function;
  67 }
  68
  69 } // end compute namespace
  70 } // end boost namespace
  71
  72 #endif // BOOST_COMPUTE_ALGORITHM_FOR_EACH_HPP