Part 4: Using Thrust's Implementation

Author: Zixin Zhang

src/main.cpp

@@ -137,23 +137,20 @@ int main(int argc, char* argv[]) {
     printCmpResult(NPOT, b, c);
 
 
-#if 0
     zeroArray(SIZE, c);
     printDesc("thrust scan, power-of-two");
     StreamCompaction::Thrust::scan(SIZE, c, a);
     printElapsedTime(StreamCompaction::Thrust::timer().getGpuElapsedTimeForPreviousOperation(), "(CUDA Measured)");
-    //printArray(SIZE, c, true);
+    printArray(SIZE, c, true);
     printCmpResult(SIZE, b, c);
 
     zeroArray(SIZE, c);
     printDesc("thrust scan, non-power-of-two");
     StreamCompaction::Thrust::scan(NPOT, c, a);
     printElapsedTime(StreamCompaction::Thrust::timer().getGpuElapsedTimeForPreviousOperation(), "(CUDA Measured)");
-    //printArray(NPOT, c, true);
+    printArray(NPOT, c, true);
     printCmpResult(NPOT, b, c);
 
-#endif
-
 
     printf("\n");
     printf("*****************************\n");

stream_compaction/thrust.cu

@@ -18,11 +18,17 @@ namespace StreamCompaction {
          * Performs prefix-sum (aka scan) on idata, storing the result into odata.
          */
         void scan(int n, int *odata, const int *idata) {
+
+            thrust::device_vector<int> d_idata(idata, idata + n);
+            thrust::device_vector<int> d_result(n);
+
             timer().startGpuTimer();
             // TODO use `thrust::exclusive_scan`
             // example: for device_vectors dv_in and dv_out:
             // thrust::exclusive_scan(dv_in.begin(), dv_in.end(), dv_out.begin());
+            thrust::exclusive_scan(d_idata.begin(), d_idata.end(), d_result.begin());
             timer().endGpuTimer();
+            thrust::copy(d_result.begin(), d_result.end(), odata);
         }
     }
 }