CUDA Traction Reduction with Dual Band Arrays

Question

I have the following (compiled and executable) code using CUDA Thrust to perform abbreviations of float2 arrays. It works correctly

 using namespace std; // includes, system #include <stdlib.h> #include <stdio.h> #include <string.h> #include <math.h> #include <conio.h> #include <typeinfo> #include <iostream> // includes CUDA #include <cuda.h> #include <cuda_runtime.h> // includes Thrust #include <thrust/host_vector.h> #include <thrust/device_vector.h> #include <thrust/reduce.h> // float2 + struct struct add_float2 { __device__ float2 operator()(const float2& a, const float2& b) const { float2 r; rx = ax + bx; ry = ay + by; return r; } }; // double2 + struct struct add_double2 { __device__ double2 operator()(const double2& a, const double2& b) const { double2 r; rx = ax + bx; ry = ay + by; return r; } }; void main( int argc, char** argv) { int N = 20; // --- Host float2* ha; ha = (float2*) malloc(N*sizeof(float2)); for (unsigned i=0; i<N; ++i) { ha[i].x = 1; ha[i].y = 2; } // --- Device float2* da; cudaMalloc((void**)&da,N*sizeof(float2)); cudaMemcpy(da,ha,N*sizeof(float2),cudaMemcpyHostToDevice); thrust::device_ptr<float2> dev_ptr_1(da); thrust::device_ptr<float2> dev_ptr_2(da+N); float2 init; init.x = init.y = 0.0f; float2 sum = thrust::reduce(dev_ptr_1,dev_ptr_2,init,add_float2()); cout << " Real part = " << sum.x << "; Imaginary part = " << sum.y << endl; getch(); } 

However, when I change float2 to double2 in the main program, namely

 void main( int argc, char** argv) { int N = 20; // --- Host double2* ha; ha = (double2*) malloc(N*sizeof(double2)); for (unsigned i=0; i<N; ++i) { ha[i].x = 1; ha[i].y = 2; } // --- Device double2* da; cudaMalloc((void**)&da,N*sizeof(double2)); cudaMemcpy(da,ha,N*sizeof(double2),cudaMemcpyHostToDevice); thrust::device_ptr<double2> dev_ptr_1(da); thrust::device_ptr<double2> dev_ptr_2(da+N); double2 init; init.x = init.y = 0.0; double2 sum = thrust::reduce(dev_ptr_1,dev_ptr_2,init,add_double2()); cout << " Real part = " << sum.x << "; Imaginary part = " << sum.y << endl; getch(); } 

I get an exception in the reduce line. How can I use CUDA Thrust shortening with double2 arrays? Am I doing something wrong? Thanks in advance.

WORKING DECISION AFTER OPENING TALONMEN

using the std namespace;

 // includes, system #include <stdlib.h> #include <stdio.h> #include <string.h> #include <math.h> #include <conio.h> #include <typeinfo> #include <iostream> // includes CUDA #include <cuda.h> #include <cuda_runtime.h> // includes Thrust #include <thrust/host_vector.h> #include <thrust/device_vector.h> #include <thrust/reduce.h> struct my_double2 { double x, y; }; // double2 + struct struct add_my_double2 { __device__ my_double2 operator()(const my_double2& a, const my_double2& b) const { my_double2 r; rx = ax + bx; ry = ay + by; return r; } }; void main( int argc, char** argv) { int N = 20; // --- Host my_double2* ha; ha = (my_double2*) malloc(N*sizeof(my_double2)); for (unsigned i=0; i<N; ++i) { ha[i].x = 1; ha[i].y = 2; } // --- Device my_double2* da; cudaMalloc((void**)&da,N*sizeof(my_double2)); cudaMemcpy(da,ha,N*sizeof(my_double2),cudaMemcpyHostToDevice); thrust::device_ptr<my_double2> dev_ptr_1(da); thrust::device_ptr<my_double2> dev_ptr_2(da+N); my_double2 init; init.x = init.y = 0.0; cout << "here3\n"; my_double2 sum = thrust::reduce(dev_ptr_1,dev_ptr_2,init,add_my_double2()); cout << " Real part = " << sum.x << "; Imaginary part = " << sum.y << endl; getch(); } 

Answer 1

This is a known incompatibility with MSVC and nvcc. See here . The solution is to define your own version of double2 and use it instead.

Just for reference, I can compile and run your code correctly in a 64-bit Linux field with CUDA 5.5.