Fix non-nvcc Thrust compilation

doc/tutorial_thrust_cuda.qbk

@@ -52,7 +52,7 @@ Thrust follows a functional programming approach. If you want to perform a calcu
 struct add_two
 {
     template< class T >
-    __host__ __device__
+    _CCCL_HOST_DEVICE
     void operator()( T &t ) const
     {
         t += T( 2 );
@@ -74,7 +74,7 @@ The mean field is calculated in a class `mean_field_calculator`
 
 [thrust_phase_ensemble_mean_field_calculator] 
 
-Inside this class two member structures `sin_functor` and `cos_functor` are defined. They compute the sine and the cosine of a value and they are used within a transform iterator to calculate the sum of ['sin(__phi[subl k])] and ['cos(__phi[subl k])]. The classifiers `__host__` and `__device__` are CUDA specific and define a function or operator which can be executed on the GPU as well as on the CPU. The line 
+Inside this class two member structures `sin_functor` and `cos_functor` are defined. They compute the sine and the cosine of a value and they are used within a transform iterator to calculate the sum of ['sin(__phi[subl k])] and ['cos(__phi[subl k])]. The macro '_CCCL_HOST_DEVICE' defines a function or operator which can be executed on the GPU as well as on the CPU if compiled with nvcc. The line
 
 [thrust_phase_ensemble_sin_sum]
 

examples/thrust/lorenz_parameters.cu

@@ -48,7 +48,7 @@ struct lorenz_system
     struct lorenz_functor
     {
         template< class T >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( T t ) const
         {
             // unpack the parameter we want to vary and the Lorenz variables
@@ -98,7 +98,7 @@ struct lorenz_perturbation_system
     struct lorenz_perturbation_functor
     {
         template< class T >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( T t ) const
         {
             value_type R = thrust::get< 1 >( t );
@@ -171,7 +171,7 @@ struct lyap_observer
     struct lyap_functor
     {
         template< class T >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( T t ) const
         {
             value_type &dx = thrust::get< 0 >( t );

examples/thrust/phase_oscillator_chain.cu

@@ -64,7 +64,7 @@ public:
     struct sys_functor
     {
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t )  // this functor works on tuples of values
         {
             // first, unpack the tuple into value, neighbors and omega

examples/thrust/phase_oscillator_ensemble.cu

@@ -57,7 +57,7 @@ struct mean_field_calculator
 {
     struct sin_functor : public thrust::unary_function< value_type , value_type >
     {
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         value_type operator()( value_type x) const
         {
             return sin( x );
@@ -66,7 +66,7 @@ struct mean_field_calculator
 
     struct cos_functor : public thrust::unary_function< value_type , value_type >
     {
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         value_type operator()( value_type x) const
         {
             return cos( x );
@@ -111,7 +111,7 @@ public:
         : m_K( K ) , m_Theta( Theta ) , m_epsilon( epsilon ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t )
         {
             thrust::get<2>(t) = thrust::get<1>(t) + m_epsilon * m_K * sin( m_Theta - thrust::get<0>(t) );

examples/thrust/relaxation.cu

@@ -34,7 +34,7 @@ struct relaxation
     struct relaxation_functor
     {
         template< class T >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( T t ) const
         {
             // unpack the parameter we want to vary and the Lorenz variables

include/boost/numeric/odeint/external/thrust/thrust_algebra.hpp

@@ -37,7 +37,7 @@ namespace detail {
     struct maximum
     {
         template< class Fac1 , class Fac2 >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         Value operator()( const Fac1 t1 , const Fac2 t2 ) const
         {
             return ( abs( t1 ) < abs( t2 ) ) ? t2 : t1 ;

include/boost/numeric/odeint/external/thrust/thrust_operations.hpp

@@ -39,7 +39,7 @@ struct thrust_operations
         : m_alpha1( alpha1 ) , m_alpha2( alpha2 ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             thrust::get<0>(t) = m_alpha1 * thrust::get<1>(t) + m_alpha2 * thrust::get<2>(t);
@@ -56,7 +56,7 @@ struct thrust_operations
         : m_alpha1( alpha1 ) , m_alpha2( alpha2 ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             typename thrust::tuple_element<0,Tuple>::type tmp = thrust::get<0>(t);
@@ -76,7 +76,7 @@ struct thrust_operations
         : m_alpha1( alpha1 ) , m_alpha2( alpha2 ) , m_alpha3( alpha3 ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             thrust::get<0>(t) = m_alpha1 * thrust::get<1>(t) +
@@ -98,7 +98,7 @@ struct thrust_operations
         : m_alpha1( alpha1 ) , m_alpha2( alpha2 ) , m_alpha3( alpha3 ) , m_alpha4( alpha4 ){ }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             thrust::get<0>(t) = m_alpha1 * thrust::get<1>(t) +
@@ -125,7 +125,7 @@ struct thrust_operations
           m_alpha4( alpha4 ) , m_alpha5( alpha5 ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             thrust::get<0>(t) = m_alpha1 * thrust::get<1>(t) +
@@ -154,7 +154,7 @@ struct thrust_operations
           m_alpha4( alpha4 ) , m_alpha5( alpha5 ) , m_alpha6( alpha6 ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             thrust::get<0>(t) = m_alpha1 * thrust::get<1>(t) +
@@ -185,7 +185,7 @@ struct thrust_operations
           m_alpha4( alpha4 ) , m_alpha5( alpha5 ) , m_alpha6( alpha6 ) , m_alpha7( alpha7 ) { }
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             thrust::get<0>(t) = m_alpha1 * thrust::get<1>(t) +
@@ -211,7 +211,7 @@ struct thrust_operations
 
 
         template< class Tuple >
-        __host__ __device__
+        _CCCL_HOST_DEVICE
         void operator()( Tuple t ) const
         {
             using std::abs;