title: CUDA thrust zip_iterator tuple transform_reduce
body: <p>I want to compute <a href="https://i.stack.imgur.com/gruWB.gif" rel="nofollow noreferrer"><img src="https://i.stack.imgur.com/gruWB.gif" alt="\left | \vec{a} - \vec{b} \right |"></a> for vectors <a href="https://i.stack.imgur.com/115MS.gif" rel="nofollow noreferrer"><img src="https://i.stack.imgur.com/115MS.gif" alt="\vec{a}"></a> and <a href="https://i.stack.imgur.com/nzrrp.gif" rel="nofollow noreferrer"><img src="https://i.stack.imgur.com/nzrrp.gif" alt="\vec{b}"></a>, where <a href="https://i.stack.imgur.com/qcQd4.gif" rel="nofollow noreferrer"><img src="https://i.stack.imgur.com/qcQd4.gif" alt="\left | \vec{x} \right |"></a> denotes the magnitude of the vector <a href="https://i.stack.imgur.com/NmSNC.gif" rel="nofollow noreferrer"><img src="https://i.stack.imgur.com/NmSNC.gif" alt="\vec{x}"></a>. Since this involves taking the square root of the sum of the squares of the differences between each corresponding component of the two vectors, it should be a highly parallelizable task. I am using Cuda and Thrust, through Cygwin, on Windows 10. Both Cuda and Thrust are in general working.</p>
  
  <p>The below code compiles and runs (with nvcc), but only because I have commented out three lines toward the bottom of <code>main</code>, each of which I think should work but does not. <code>func::operator()(tup t)</code> thinks that the arguments I'm passing it are not in fact of type tup. </p>
  
  <p>I have also commented out the actual body of the operator, in the interest of making it more likely to at least compile. The operator is supposed to find the squared difference between the elements of the input tup. The reduction <code>unary_op</code> from <code>transform_reduce</code> (which in this case is <code>func()</code>) would then add these, giving me the norm squared of the difference of vectors. </p>
  
  <pre><code>#include &lt;iostream&gt;
  #include &lt;stdlib.h&gt;
  #include &lt;thrust/device_vector.h&gt;
  #include &lt;thrust/transform.h&gt;
  #include &lt;thrust/tuple.h&gt;
  #include &lt;thrust/transform_reduce.h&gt;
  #include &lt;thrust/iterator/zip_iterator.h&gt;
  
  typedef thrust::device_vector&lt;float&gt; dvec;
  typedef dvec::iterator iter;
  typedef thrust::tuple&lt;iter, iter&gt; tup;
  
  struct func: public thrust::unary_function&lt;tup, float&gt;
  {
    __device__ float operator()(tup t) //difsq
    {
      // I've commented out these two lines for testing purposes:
      // float f = thrust::get&lt;0&gt;(t) - thrust::get&lt;1&gt;(t);
      // return f*f;
      return 3.14;
    }
  };
  
  int main()
  {
    dvec a(40, 4.f);
    dvec b(40, 3.f);
    auto begin = thrust::make_zip_iterator(thrust::make_tuple(a.begin(), b.begin()));
    auto end = thrust::make_zip_iterator(thrust::make_tuple(a.end(), b.end()));
  
    //these two lines work
    thrust::get&lt;0&gt;(begin[0]);
    std::cout &lt;&lt; thrust::get&lt;0&gt;(begin[0]) - thrust::get&lt;1&gt;(begin[0]);
  
  
    //these three lines do not
    //thrust::transform_reduce(begin, end, func(), 0.0f, thrust::plus&lt;float&gt;());
    //func()(begin[0]);
    //thrust::transform(begin, end, begin, func());
  
  
    std::cout &lt;&lt; "done" &lt;&lt; std::endl;
    return 0;
  }
  </code></pre>
  
  <p>I get this error: (my program is called <code>sandbox.cu</code>)</p>
  
  <pre><code>C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\bin/../include\thrust/detail/tuple.inl(310): error: no instance of constructor "thrust::detail::normal_iterator&lt;Pointer&gt;::normal_iterator [with Pointer=thrust::device_ptr&lt;float&gt;]" matches the argument list
          argument types are: (const thrust::device_reference&lt;float&gt;)
        detected during:
          instantiation of "thrust::detail::cons&lt;HT, TT&gt;::cons(const thrust::detail::cons&lt;HT2, TT2&gt; &amp;) [with HT=iter, TT=thrust::detail::cons&lt;iter, thrust::null_type&gt;, HT2=thrust::device_reference&lt;float&gt;, TT2=thrust::detail::cons&lt;thrust::device_reference&lt;float&gt;, thrust::null_type&gt;]"
  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\bin/../include\thrust/tuple.h(361): here
          instantiation of "thrust::tuple&lt;T0, T1, T2, T3, T4, T5, T6, T7, T8, T9&gt;::tuple(const thrust::detail::cons&lt;U1, U2&gt; &amp;) [with T0=iter, T1=iter, T2=thrust::null_type, T3=thrust::null_type, T4=thrust::null_type, T5=thrust::null_type, T6=thrust::null_type, T7=thrust::null_type, T8=thrust::null_type, T9=thrust::null_type, U1=thrust::device_reference&lt;float&gt;, U2=thrust::detail::cons&lt;thrust::device_reference&lt;float&gt;, thrust::null_type&gt;]"
  sandbox.cu(37): here
  
  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\bin/../include\thrust/detail/tuple.inl(411): error: no instance of constructor "thrust::detail::normal_iterator&lt;Pointer&gt;::normal_iterator [with Pointer=thrust::device_ptr&lt;float&gt;]" matches the argument list
              argument types are: (const thrust::device_reference&lt;float&gt;)
            detected during:
              instantiation of "thrust::detail::cons&lt;HT, thrust::null_type&gt;::cons(const thrust::detail::cons&lt;HT2, thrust::null_type&gt; &amp;) [with HT=iter, HT2=thrust::device_reference&lt;float&gt;]"
  (310): here
              instantiation of "thrust::detail::cons&lt;HT, TT&gt;::cons(const thrust::detail::cons&lt;HT2, TT2&gt; &amp;) [with HT=iter, TT=thrust::detail::cons&lt;iter, thrust::null_type&gt;, HT2=thrust::device_reference&lt;float&gt;, TT2=thrust::detail::cons&lt;thrust::device_reference&lt;float&gt;, thrust::null_type&gt;]"
  C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v7.5\bin/../include\thrust/tuple.h(361): here
              instantiation of "thrust::tuple&lt;T0, T1, T2, T3, T4, T5, T6, T7, T8, T9&gt;::tuple(const thrust::detail::cons&lt;U1, U2&gt; &amp;) [with T0=iter, T1=iter, T2=thrust::null_type, T3=thrust::null_type, T4=thrust::null_type, T5=thrust::null_type, T6=thrust::null_type, T7=thrust::null_type, T8=thrust::null_type, T9=thrust::null_type, U1=thrust::device_reference&lt;float&gt;, U2=thrust::detail::cons&lt;thrust::device_reference&lt;float&gt;, thrust::null_type&gt;]"
  sandbox.cu(37): here
  
  2 errors detected in the compilation of "C:/cygwin64/tmp/tmpxft_00001a90_00000000-10_sandbox.cpp1.ii".
  </code></pre>