Replace CUB_RUNTIME_FUNCTION with THRUST_RUNTIME_FUNCTION.

Replace CUDA_VERSION (only available if <cuda.h> is included) with CUDART_VERSION.

Author: brycelelbach

dependencies/cub

@@ -119,7 +119,7 @@ void TestVectorFromSTLVector(void)
     ASSERT_EQUAL(v[2], 2);
 
     v = stl_vector;
-    
+
     ASSERT_EQUAL(v.size(), 3lu);
     ASSERT_EQUAL(v[0], 0);
     ASSERT_EQUAL(v[1], 1);
@@ -169,7 +169,7 @@ template <class Vector>
 void TestVectorFromBiDirectionalIterator(void)
 {
     typedef typename Vector::value_type T;
-    
+
     std::list<T> stl_list;
     stl_list.push_back(0);
     stl_list.push_back(1);
@@ -189,7 +189,7 @@ template <class Vector>
 void TestVectorAssignFromBiDirectionalIterator(void)
 {
     typedef typename Vector::value_type T;
-    
+
     std::list<T> stl_list;
     stl_list.push_back(0);
     stl_list.push_back(1);
@@ -246,7 +246,7 @@ void TestVectorToAndFromHostVector(void)
     v[1] = 11;
     v[2] = 12;
 
-    ASSERT_EQUAL(h[0], 0);  ASSERT_EQUAL(v[0], 10); 
+    ASSERT_EQUAL(h[0], 0);  ASSERT_EQUAL(v[0], 10);
     ASSERT_EQUAL(h[1], 1);  ASSERT_EQUAL(v[1], 11);
     ASSERT_EQUAL(h[2], 2);  ASSERT_EQUAL(v[2], 12);
 
@@ -303,7 +303,7 @@ void TestVectorToAndFromDeviceVector(void)
     v[1] = 11;
     v[2] = 12;
 
-    ASSERT_EQUAL(h[0], 0);  ASSERT_EQUAL(v[0], 10); 
+    ASSERT_EQUAL(h[0], 0);  ASSERT_EQUAL(v[0], 10);
     ASSERT_EQUAL(h[1], 1);  ASSERT_EQUAL(v[1], 11);
     ASSERT_EQUAL(h[2], 2);  ASSERT_EQUAL(v[2], 12);
 
@@ -348,7 +348,7 @@ void TestVectorSwap(void)
 
     v.swap(u);
 
-    ASSERT_EQUAL(v[0], 10); ASSERT_EQUAL(u[0], 0);  
+    ASSERT_EQUAL(v[0], 10); ASSERT_EQUAL(u[0], 0);
     ASSERT_EQUAL(v[1], 11); ASSERT_EQUAL(u[1], 1);
     ASSERT_EQUAL(v[2], 12); ASSERT_EQUAL(u[2], 2);
 }
@@ -363,33 +363,33 @@ void TestVectorErasePosition(void)
 
     v.erase(v.begin() + 2);
 
-    ASSERT_EQUAL(v.size(), 4lu); 
-    ASSERT_EQUAL(v[0], 0); 
-    ASSERT_EQUAL(v[1], 1); 
-    ASSERT_EQUAL(v[2], 3); 
-    ASSERT_EQUAL(v[3], 4); 
-    
+    ASSERT_EQUAL(v.size(), 4lu);
+    ASSERT_EQUAL(v[0], 0);
+    ASSERT_EQUAL(v[1], 1);
+    ASSERT_EQUAL(v[2], 3);
+    ASSERT_EQUAL(v[3], 4);
+
     v.erase(v.begin() + 0);
 
-    ASSERT_EQUAL(v.size(), 3lu); 
-    ASSERT_EQUAL(v[0], 1); 
-    ASSERT_EQUAL(v[1], 3); 
-    ASSERT_EQUAL(v[2], 4); 
-    
+    ASSERT_EQUAL(v.size(), 3lu);
+    ASSERT_EQUAL(v[0], 1);
+    ASSERT_EQUAL(v[1], 3);
+    ASSERT_EQUAL(v[2], 4);
+
     v.erase(v.begin() + 2);
 
-    ASSERT_EQUAL(v.size(), 2lu); 
-    ASSERT_EQUAL(v[0], 1); 
-    ASSERT_EQUAL(v[1], 3); 
-    
+    ASSERT_EQUAL(v.size(), 2lu);
+    ASSERT_EQUAL(v[0], 1);
+    ASSERT_EQUAL(v[1], 3);
+
     v.erase(v.begin() + 1);
 
-    ASSERT_EQUAL(v.size(), 1lu); 
-    ASSERT_EQUAL(v[0], 1); 
+    ASSERT_EQUAL(v.size(), 1lu);
+    ASSERT_EQUAL(v[0], 1);
 
     v.erase(v.begin() + 0);
 
-    ASSERT_EQUAL(v.size(), 0lu); 
+    ASSERT_EQUAL(v.size(), 0lu);
 }
 DECLARE_VECTOR_UNITTEST(TestVectorErasePosition);
 
@@ -402,26 +402,26 @@ void TestVectorEraseRange(void)
 
     v.erase(v.begin() + 1, v.begin() + 3);
 
-    ASSERT_EQUAL(v.size(), 4lu); 
-    ASSERT_EQUAL(v[0], 0); 
-    ASSERT_EQUAL(v[1], 3); 
-    ASSERT_EQUAL(v[2], 4); 
-    ASSERT_EQUAL(v[3], 5); 
-    
+    ASSERT_EQUAL(v.size(), 4lu);
+    ASSERT_EQUAL(v[0], 0);
+    ASSERT_EQUAL(v[1], 3);
+    ASSERT_EQUAL(v[2], 4);
+    ASSERT_EQUAL(v[3], 5);
+
     v.erase(v.begin() + 2, v.end());
 
-    ASSERT_EQUAL(v.size(), 2lu); 
-    ASSERT_EQUAL(v[0], 0); 
-    ASSERT_EQUAL(v[1], 3); 
-    
+    ASSERT_EQUAL(v.size(), 2lu);
+    ASSERT_EQUAL(v[0], 0);
+    ASSERT_EQUAL(v[1], 3);
+
     v.erase(v.begin() + 0, v.begin() + 1);
 
-    ASSERT_EQUAL(v.size(), 1lu); 
-    ASSERT_EQUAL(v[0], 3); 
-    
+    ASSERT_EQUAL(v.size(), 1lu);
+    ASSERT_EQUAL(v[0], 3);
+
     v.erase(v.begin(), v.end());
 
-    ASSERT_EQUAL(v.size(), 0lu); 
+    ASSERT_EQUAL(v.size(), 0lu);
 }
 DECLARE_VECTOR_UNITTEST(TestVectorEraseRange);
 
@@ -449,37 +449,37 @@ void TestVectorEquality(void)
     s_b[0] = 0;    s_b[1] = 1;    s_b[2] = 3;
     s_b[0] = 0;    s_b[1] = 1;
 
-    ASSERT_EQUAL((h_a == h_a), true); ASSERT_EQUAL((h_a == d_a), true); ASSERT_EQUAL((d_a == h_a), true);  ASSERT_EQUAL((d_a == d_a), true); 
+    ASSERT_EQUAL((h_a == h_a), true); ASSERT_EQUAL((h_a == d_a), true); ASSERT_EQUAL((d_a == h_a), true);  ASSERT_EQUAL((d_a == d_a), true);
     ASSERT_EQUAL((h_b == h_b), true); ASSERT_EQUAL((h_b == d_b), true); ASSERT_EQUAL((d_b == h_b), true);  ASSERT_EQUAL((d_b == d_b), true);
     ASSERT_EQUAL((h_c == h_c), true); ASSERT_EQUAL((h_c == d_c), true); ASSERT_EQUAL((d_c == h_c), true);  ASSERT_EQUAL((d_c == d_c), true);
 
     // test vector vs device_vector
-    ASSERT_EQUAL((s_a == d_a), true); ASSERT_EQUAL((d_a == s_a), true); 
+    ASSERT_EQUAL((s_a == d_a), true); ASSERT_EQUAL((d_a == s_a), true);
     ASSERT_EQUAL((s_b == d_b), true); ASSERT_EQUAL((d_b == s_b), true);
     ASSERT_EQUAL((s_c == d_c), true); ASSERT_EQUAL((d_c == s_c), true);
 
     // test vector vs host_vector
-    ASSERT_EQUAL((s_a == h_a), true); ASSERT_EQUAL((h_a == s_a), true); 
+    ASSERT_EQUAL((s_a == h_a), true); ASSERT_EQUAL((h_a == s_a), true);
     ASSERT_EQUAL((s_b == h_b), true); ASSERT_EQUAL((h_b == s_b), true);
     ASSERT_EQUAL((s_c == h_c), true); ASSERT_EQUAL((h_c == s_c), true);
 
-    ASSERT_EQUAL((h_a == h_b), false); ASSERT_EQUAL((h_a == d_b), false); ASSERT_EQUAL((d_a == h_b), false); ASSERT_EQUAL((d_a == d_b), false); 
+    ASSERT_EQUAL((h_a == h_b), false); ASSERT_EQUAL((h_a == d_b), false); ASSERT_EQUAL((d_a == h_b), false); ASSERT_EQUAL((d_a == d_b), false);
     ASSERT_EQUAL((h_b == h_a), false); ASSERT_EQUAL((h_b == d_a), false); ASSERT_EQUAL((d_b == h_a), false); ASSERT_EQUAL((d_b == d_a), false);
     ASSERT_EQUAL((h_a == h_c), false); ASSERT_EQUAL((h_a == d_c), false); ASSERT_EQUAL((d_a == h_c), false); ASSERT_EQUAL((d_a == d_c), false);
     ASSERT_EQUAL((h_c == h_a), false); ASSERT_EQUAL((h_c == d_a), false); ASSERT_EQUAL((d_c == h_a), false); ASSERT_EQUAL((d_c == d_a), false);
     ASSERT_EQUAL((h_b == h_c), false); ASSERT_EQUAL((h_b == d_c), false); ASSERT_EQUAL((d_b == h_c), false); ASSERT_EQUAL((d_b == d_c), false);
     ASSERT_EQUAL((h_c == h_b), false); ASSERT_EQUAL((h_c == d_b), false); ASSERT_EQUAL((d_c == h_b), false); ASSERT_EQUAL((d_c == d_b), false);
 
     // test vector vs device_vector
-    ASSERT_EQUAL((s_a == d_b), false); ASSERT_EQUAL((d_a == s_b), false); 
+    ASSERT_EQUAL((s_a == d_b), false); ASSERT_EQUAL((d_a == s_b), false);
     ASSERT_EQUAL((s_b == d_a), false); ASSERT_EQUAL((d_b == s_a), false);
     ASSERT_EQUAL((s_a == d_c), false); ASSERT_EQUAL((d_a == s_c), false);
     ASSERT_EQUAL((s_c == d_a), false); ASSERT_EQUAL((d_c == s_a), false);
     ASSERT_EQUAL((s_b == d_c), false); ASSERT_EQUAL((d_b == s_c), false);
     ASSERT_EQUAL((s_c == d_b), false); ASSERT_EQUAL((d_c == s_b), false);
 
     // test vector vs host_vector
-    ASSERT_EQUAL((s_a == h_b), false); ASSERT_EQUAL((h_a == s_b), false); 
+    ASSERT_EQUAL((s_a == h_b), false); ASSERT_EQUAL((h_a == s_b), false);
     ASSERT_EQUAL((s_b == h_a), false); ASSERT_EQUAL((h_b == s_a), false);
     ASSERT_EQUAL((s_a == h_c), false); ASSERT_EQUAL((h_a == s_c), false);
     ASSERT_EQUAL((s_c == h_a), false); ASSERT_EQUAL((h_c == s_a), false);
@@ -511,37 +511,37 @@ void TestVectorInequality(void)
     s_b[0] = 0;    s_b[1] = 1;    s_b[2] = 3;
     s_b[0] = 0;    s_b[1] = 1;
 
-    ASSERT_EQUAL((h_a != h_a), false); ASSERT_EQUAL((h_a != d_a), false); ASSERT_EQUAL((d_a != h_a), false);  ASSERT_EQUAL((d_a != d_a), false); 
+    ASSERT_EQUAL((h_a != h_a), false); ASSERT_EQUAL((h_a != d_a), false); ASSERT_EQUAL((d_a != h_a), false);  ASSERT_EQUAL((d_a != d_a), false);
     ASSERT_EQUAL((h_b != h_b), false); ASSERT_EQUAL((h_b != d_b), false); ASSERT_EQUAL((d_b != h_b), false);  ASSERT_EQUAL((d_b != d_b), false);
     ASSERT_EQUAL((h_c != h_c), false); ASSERT_EQUAL((h_c != d_c), false); ASSERT_EQUAL((d_c != h_c), false);  ASSERT_EQUAL((d_c != d_c), false);
 
     // test vector vs device_vector
-    ASSERT_EQUAL((s_a != d_a), false); ASSERT_EQUAL((d_a != s_a), false); 
+    ASSERT_EQUAL((s_a != d_a), false); ASSERT_EQUAL((d_a != s_a), false);
     ASSERT_EQUAL((s_b != d_b), false); ASSERT_EQUAL((d_b != s_b), false);
     ASSERT_EQUAL((s_c != d_c), false); ASSERT_EQUAL((d_c != s_c), false);
 
     // test vector vs host_vector
-    ASSERT_EQUAL((s_a != h_a), false); ASSERT_EQUAL((h_a != s_a), false); 
+    ASSERT_EQUAL((s_a != h_a), false); ASSERT_EQUAL((h_a != s_a), false);
     ASSERT_EQUAL((s_b != h_b), false); ASSERT_EQUAL((h_b != s_b), false);
     ASSERT_EQUAL((s_c != h_c), false); ASSERT_EQUAL((h_c != s_c), false);
 
-    ASSERT_EQUAL((h_a != h_b), true); ASSERT_EQUAL((h_a != d_b), true); ASSERT_EQUAL((d_a != h_b), true); ASSERT_EQUAL((d_a != d_b), true); 
+    ASSERT_EQUAL((h_a != h_b), true); ASSERT_EQUAL((h_a != d_b), true); ASSERT_EQUAL((d_a != h_b), true); ASSERT_EQUAL((d_a != d_b), true);
     ASSERT_EQUAL((h_b != h_a), true); ASSERT_EQUAL((h_b != d_a), true); ASSERT_EQUAL((d_b != h_a), true); ASSERT_EQUAL((d_b != d_a), true);
     ASSERT_EQUAL((h_a != h_c), true); ASSERT_EQUAL((h_a != d_c), true); ASSERT_EQUAL((d_a != h_c), true); ASSERT_EQUAL((d_a != d_c), true);
     ASSERT_EQUAL((h_c != h_a), true); ASSERT_EQUAL((h_c != d_a), true); ASSERT_EQUAL((d_c != h_a), true); ASSERT_EQUAL((d_c != d_a), true);
     ASSERT_EQUAL((h_b != h_c), true); ASSERT_EQUAL((h_b != d_c), true); ASSERT_EQUAL((d_b != h_c), true); ASSERT_EQUAL((d_b != d_c), true);
     ASSERT_EQUAL((h_c != h_b), true); ASSERT_EQUAL((h_c != d_b), true); ASSERT_EQUAL((d_c != h_b), true); ASSERT_EQUAL((d_c != d_b), true);
 
     // test vector vs device_vector
-    ASSERT_EQUAL((s_a != d_b), true); ASSERT_EQUAL((d_a != s_b), true); 
+    ASSERT_EQUAL((s_a != d_b), true); ASSERT_EQUAL((d_a != s_b), true);
     ASSERT_EQUAL((s_b != d_a), true); ASSERT_EQUAL((d_b != s_a), true);
     ASSERT_EQUAL((s_a != d_c), true); ASSERT_EQUAL((d_a != s_c), true);
     ASSERT_EQUAL((s_c != d_a), true); ASSERT_EQUAL((d_c != s_a), true);
     ASSERT_EQUAL((s_b != d_c), true); ASSERT_EQUAL((d_b != s_c), true);
     ASSERT_EQUAL((s_c != d_b), true); ASSERT_EQUAL((d_c != s_b), true);
 
     // test vector vs host_vector
-    ASSERT_EQUAL((s_a != h_b), true); ASSERT_EQUAL((h_a != s_b), true); 
+    ASSERT_EQUAL((s_a != h_b), true); ASSERT_EQUAL((h_a != s_b), true);
     ASSERT_EQUAL((s_b != h_a), true); ASSERT_EQUAL((h_b != s_a), true);
     ASSERT_EQUAL((s_a != h_c), true); ASSERT_EQUAL((h_a != s_c), true);
     ASSERT_EQUAL((s_c != h_a), true); ASSERT_EQUAL((h_c != s_a), true);
@@ -585,8 +585,8 @@ void TestVectorResizing(void)
 
     ASSERT_EQUAL(v.size(), 0lu);
 
-// TODO remove this WAR      
-#if defined(__CUDACC__) && CUDA_VERSION==3000
+// TODO remove this WAR
+#if defined(__CUDACC__) && CUDART_VERSION==3000
     // depending on sizeof(T), we will receive one
     // of two possible exceptions
     try
@@ -599,7 +599,7 @@ void TestVectorResizing(void)
       // reset the CUDA error
       cudaGetLastError();
     } // end catch
-#endif // defined(__CUDACC__) && CUDA_VERSION==3000
+#endif // defined(__CUDACC__) && CUDART_VERSION==3000
 
     ASSERT_EQUAL(v.size(), 0lu);
 }
@@ -622,15 +622,15 @@ void TestVectorReserving(void)
 
     ASSERT_EQUAL(v.capacity(), old_capacity);
 
-// TODO remove this WAR      
-#if defined(__CUDACC__) && CUDA_VERSION==3000
+// TODO remove this WAR
+#if defined(__CUDACC__) && CUDART_VERSION==3000
     try
     {
       v.reserve(std::numeric_limits<size_t>::max());
     }
     catch(std::length_error e) {}
     catch(std::bad_alloc e) {}
-#endif // defined(__CUDACC__) && CUDA_VERSION==3000
+#endif // defined(__CUDACC__) && CUDART_VERSION==3000
 
     ASSERT_EQUAL(v.capacity(), old_capacity);
 }
@@ -680,7 +680,7 @@ struct LargeStruct
 
 void TestVectorContainingLargeType(void)
 {
-    // Thrust issue #5 
+    // Thrust issue #5
     // http://code.google.com/p/thrust/issues/detail?id=5
     const static int N = 100;
     typedef LargeStruct<N> T;
@@ -692,9 +692,9 @@ void TestVectorContainingLargeType(void)
 
     thrust::device_vector<T> dv2(20);
     thrust::host_vector<T>   hv2(20);
-    
+
     ASSERT_EQUAL_QUIET(dv2, hv2);
-    
+
     // initialize tofirst element to something nonzero
     T ls;
 
@@ -703,15 +703,15 @@ void TestVectorContainingLargeType(void)
 
     thrust::device_vector<T> dv3(20, ls);
     thrust::host_vector<T>   hv3(20, ls);
-    
+
     ASSERT_EQUAL_QUIET(dv3, hv3);
-    
+
     // change first element
     ls.data[0] = -13;
 
     dv3[2] = ls;
     hv3[2] = ls;
-    
+
     ASSERT_EQUAL_QUIET(dv3, hv3);
 }
 DECLARE_UNITTEST(TestVectorContainingLargeType);

testing/vector.cu

@@ -14,7 +14,7 @@
  *  See the License for the specific language governing permissions and
  *  limitations under the License.
  */
-#pragma once 
+#pragma once
 
 #include <cmath>
 #include <thrust/detail/complex/math_private.h>
@@ -27,7 +27,7 @@ namespace complex
 {
 
 // Define basic arithmetic functions so we can use them without explicit scope
-// keeping the code as close as possible to FreeBSDs for ease of maintenance. 
+// keeping the code as close as possible to FreeBSDs for ease of maintenance.
 // It also provides an easy way to support compilers with missing C99 functions.
 // When possible, just use the names in the global scope.
 // Some platforms define these as macros, others as free functions.
@@ -105,7 +105,7 @@ __host__ __device__ inline int isfinite(double x){
 // sometimes the CUDA toolkit provides these these names as macros,
 // sometimes functions in the global scope
 
-#    if (CUDA_VERSION >= 6500)
+#    if (CUDART_VERSION >= 6500)
 using ::isinf;
 using ::isnan;
 using ::signbit;
@@ -114,7 +114,7 @@ using ::isfinite;
 #    else
 // these names are macros, we don't need to define them
 
-#    endif // CUDA_VERSION
+#    endif // CUDART_VERSION
 
 #  else
 // Some compilers do not provide these in the global scope
@@ -128,7 +128,7 @@ using std::isfinite;
 
 using ::atanh;
 #endif // _MSC_VER
-  
+
 #if defined _MSC_VER
 
 __host__ __device__ inline double copysign(double x, double y){
@@ -159,7 +159,7 @@ inline double log1p(double x){
   }else{
     if(u > 2.0){
       // Use normal log for large arguments
-      return log(u); 
+      return log(u);
     }else{
       return log(u)*(x/(u-1.0));
     }
@@ -173,7 +173,7 @@ inline float log1pf(float x){
   }else{
     if(u > 2.0f){
       // Use normal log for large arguments
-      return logf(u); 
+      return logf(u);
     }else{
       return logf(u)*(x/(u-1.0f));
     }
@@ -202,4 +202,4 @@ inline double hypot(double x, double y){
 } // namespace detail
 
 } // namespace thrust
-      
+

thrust/detail/complex/c99math.h

@@ -20,7 +20,7 @@
 
 // XXX workaround gcc 4.8+'s complaints about unused local typedefs by silencing them globally
 #if defined(THRUST_GCC_VERSION) && (THRUST_GCC_VERSION >= 40800)
-#  if defined(__NVCC__) && (CUDA_VERSION >= 6000)
+#  if defined(__NVCC__) && (CUDART_VERSION >= 6000)
 #    pragma GCC diagnostic ignored "-Wunused-local-typedefs"
 #  endif // nvcc & cuda 6+
 #endif // gcc 4.8