Make sure ThrustPatches is included before thrust.

vtkm/cont/cuda/internal/ArrayManagerExecutionCuda.h

@@ -41,6 +41,7 @@
 VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/copy.h>
 #include <thrust/device_ptr.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 #include <limits>

vtkm/cont/cuda/internal/DeviceAdapterAlgorithmCuda.h

@@ -43,12 +43,8 @@
 #include <vtkm/cont/cuda/internal/ThrustExceptionHandler.h>
 #include <vtkm/exec/cuda/internal/WrappedOperators.h>
 
-#include <vtkm/exec/internal/ErrorMessageBuffer.h>
-
-
 #include <vtkm/exec/cuda/internal/TaskStrided.h>
-
-#include <cuda.h>
+#include <vtkm/exec/internal/ErrorMessageBuffer.h>
 
 // #define PARAMETER_SWEEP_VTKM_SCHEDULER_1D
 // #define PARAMETER_SWEEP_VTKM_SCHEDULER_3D
@@ -58,7 +54,9 @@
 
 // Disable warnings we check vtkm for but Thrust does not.
 VTKM_THIRDPARTY_PRE_INCLUDE
-//our own custom thrust execution policy
+//This is required to be first so that we get patches for thrust included
+//in the correct order
+#include <cuda.h>
 #include <thrust/advance.h>
 #include <thrust/binary_search.h>
 #include <thrust/copy.h>
@@ -70,6 +68,7 @@ VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/system/cuda/vector.h>
 #include <thrust/unique.h>
 #include <vtkm/exec/cuda/internal/ExecutionPolicy.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/cont/cuda/internal/DeviceAdapterAtomicArrayImplementationCuda.h

@@ -33,6 +33,7 @@
 // Disable warnings we check vtkm for but Thrust does not.
 VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/device_ptr.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/cont/cuda/internal/ThrustExceptionHandler.h

@@ -26,6 +26,7 @@
 
 VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/system_error.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/exec/cuda/internal/ArrayPortalFromThrust.h

@@ -28,6 +28,7 @@
 
 VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/system/cuda/memory.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/exec/cuda/internal/ExecutionPolicy.h

@@ -29,6 +29,7 @@ VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/execution_policy.h>
 #include <thrust/system/cuda/execution_policy.h>
 #include <thrust/system/cuda/memory.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 #define ThrustCudaPolicyPerThread ::thrust::cuda::par.on(cudaStreamPerThread)

vtkm/exec/cuda/internal/IteratorFromArrayPortal.h

@@ -30,6 +30,7 @@ VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/functional.h>
 #include <thrust/iterator/iterator_facade.h>
 #include <thrust/system/cuda/execution_policy.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/exec/cuda/internal/WrappedOperators.h

@@ -29,6 +29,7 @@
 // Disable warnings we check vtkm for but Thrust does not.
 VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/system/cuda/memory.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/interop/cuda/internal/TransferToOpenGL.h

@@ -33,6 +33,7 @@ VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/device_ptr.h>
 #include <thrust/system/cuda/execution_policy.h>
 #include <vtkm/exec/cuda/internal/ExecutionPolicy.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 
 namespace vtkm

vtkm/worklet/Clip.h

@@ -49,6 +49,7 @@
 // wrong results
 VTKM_THIRDPARTY_PRE_INCLUDE
 #include <thrust/detail/type_traits.h>
+#include <vtkm/exec/cuda/internal/ThrustPatches.h>
 VTKM_THIRDPARTY_POST_INCLUDE
 #define THRUST_SCAN_WORKAROUND
 #endif