[WIP] Review fixes

rajeeja · Apr 2, 2020 · 6a2d004 · 6a2d004
1 parent 5ee7327
commit 6a2d004
Show file tree

Hide file tree

Showing 4 changed files with 17 additions and 19 deletions.
diff --git a/Src/Base/AMReX_CuptiTrace.H b/Src/Base/AMReX_CuptiTrace.H
@@ -6,6 +6,7 @@
 #include <vector>
 #include <cuda.h>
 #include <cupti.h>
+#include <memory>
 
 
 namespace amrex {
@@ -22,7 +23,6 @@ void cuptiTraceStop (unsigned boxUintID) noexcept;
 
 class CUpti_Activity_Userdata {
 private:
-    //CUpti_Activity* record_;
     unsigned uintID_;
     std::string charID_;
     unsigned long long startTime_;
@@ -32,15 +32,13 @@ private:
     std::string name_;
 
 public:
-    //void setRecord (CUpti_Activity*) noexcept;
     void setUintID (unsigned) noexcept;
     void setCharID (std::string) noexcept;
     void setStartTime (unsigned long long) noexcept;
     void setEndTime (unsigned long long) noexcept;
     void setTimeElapsed (unsigned long long) noexcept;
     void setStreamID (int) noexcept;
     void setName (std::string) noexcept;
-    CUpti_Activity* getRecord () noexcept;
     unsigned getUintID () noexcept;
     std::string getCharID () noexcept;
     unsigned long long getStartTime () noexcept;
@@ -60,7 +58,7 @@ public:
 };
 
 extern std::vector<std::unique_ptr<CUpti_Activity_Userdata>> activityRecordUserdata;
-double computeElapsedTimeUserdata(std::vector<std::unique_ptr<CUpti_Activity_Userdata>>
+double computeElapsedTimeUserdata(const std::vector<std::unique_ptr<CUpti_Activity_Userdata>>&
                                   activityRecordUserdata) noexcept;
 }
 

diff --git a/Src/Base/AMReX_CuptiTrace.cpp b/Src/Base/AMReX_CuptiTrace.cpp
@@ -2,9 +2,10 @@
 #ifdef AMREX_USE_CUPTI
 #include <AMReX_Print.H>
 #include <stdio.h>
-#include <map>
 #include <cuda.h>
 #include <cupti.h>
+#include <map>
+#include <memory>
 
 // CUPTI buffer size, enough for 4096 activity records in a single buffer;
 // `CUpti_Activity` objects are 8 bytes long
@@ -54,17 +55,15 @@ bfrCompleteCallback (CUcontext ctx, uint32_t streamId, uint8_t* bfr,
             if (status == CUPTI_SUCCESS) {
                 std::unique_ptr<CUpti_Activity_Userdata> recordUserData;
                 recordUserData.reset(new CUpti_Activity_Userdata());
-                std::unique_ptr<CUpti_Activity_Userdata4> kernel;
-                kernel.reset( (CUpti_ActivityKernel4*) record);
+                CUpti_ActivityKernel4* kernel = (CUpti_ActivityKernel4*) record;
 
                 // Save record data
-                //recordUserData->setRecord(record);
                 recordUserData->setStartTime(kernel->start);
                 recordUserData->setEndTime(kernel->end);
                 recordUserData->setTimeElapsed(kernel->end - kernel->start);
                 recordUserData->setStreamID(kernel->streamId);
                 recordUserData->setName((std::string)kernel->name);
-                activityRecordUserdata.push_back(recordUserData);
+                activityRecordUserdata.push_back( std::move(recordUserData) );
             }
             else if (status == CUPTI_ERROR_MAX_LIMIT_REACHED) {
                 // No more records in the buffer
@@ -84,8 +83,9 @@ bfrCompleteCallback (CUcontext ctx, uint32_t streamId, uint8_t* bfr,
                            << " activity records were dropped due to insufficient buffer space\n";
         }
     }
-    free(record);
     free(bfr);
+    record = NULL; // Cleanup dangling pointers
+    kernel = NULL; 
 }
 
 void
@@ -121,14 +121,14 @@ cuptiTraceStop (unsigned boxUintID) noexcept
 {
     cudaDeviceSynchronize();
     cuptiActivityFlushAll(0);
-    for (auto record : activityRecordUserdata) {
+    for (auto& record : activityRecordUserdata) {
         record->setUintID(boxUintID);
-        record->setCharID( ((std::string) "CharID_") + ((std::string) boxUintID) );
+        record->setCharID( ((std::string) "CharID_") + std::to_string(boxUintID) );
     }
 }
 
 double
-computeElapsedTimeUserdata (std::vector<std::unique_ptr<CUpti_Activity_Userdata>>
+computeElapsedTimeUserdata (const std::vector<std::unique_ptr<CUpti_Activity_Userdata>>&
                             activityRecordUserdata) noexcept
 {
     if (activityRecordUserdata.size() == 0) {
@@ -138,7 +138,7 @@ computeElapsedTimeUserdata (std::vector<std::unique_ptr<CUpti_Activity_Userdata>
     std::map<int, unsigned long long> streamIDToElapsedTimeMap;
 
     // Initialize tally of unique streams
-    for (auto record : activityRecordUserdata) {
+    for (auto& record : activityRecordUserdata) {
         if (streamIDToElapsedTimeMap.find(record->getStreamID())
             == streamIDToElapsedTimeMap.end()) {
             // Not found
@@ -149,7 +149,7 @@ computeElapsedTimeUserdata (std::vector<std::unique_ptr<CUpti_Activity_Userdata>
     }
 
     // Sum kernel times in each stream
-    for (auto record : activityRecordUserdata) {
+    for (auto& record : activityRecordUserdata) {
         streamIDToElapsedTimeMap[record->getStreamID()] += record->getTimeElapsed();
     }
 

diff --git a/Src/Base/AMReX_TinyProfiler.cpp b/Src/Base/AMReX_TinyProfiler.cpp
@@ -187,7 +187,7 @@ TinyProfiler::stop (unsigned boxUintID) noexcept
         t = computeElapsedTimeUserdata(activityRecordUserdata);
         int nKernelCalls = activityRecordUserdata.size();
 
-        for (auto record : activityRecordUserdata) 
+        for (auto& record : activityRecordUserdata) 
         {
             record->setUintID(boxUintID);
         }

diff --git a/Tests/GPU/CuptiTest/Exec/CUDA/myfunc.cpp b/Tests/GPU/CuptiTest/Exec/CUDA/myfunc.cpp
@@ -68,11 +68,11 @@ void doDeviceSleep (MultiFab& mf, int& n) {
 
         unsigned long long t_start = 0;
         unsigned long long t_stop = 0;
-        for (auto record : activityRecordUserdata) {
+        for (auto& record : activityRecordUserdata) {
             //std::unique_ptr<CUpti_ActivityKernel4> kernel; 
             //kernel.reset(record->getRecord());
-            t_start = (unsigned long long) record->getStartTime;//kernel->start;
-            t_stop = (unsigned long long) record->getEndTime;//kernel->end;
+            t_start = (unsigned long long) record->getStartTime();//kernel->start;
+            t_stop = (unsigned long long) record->getEndTime();   //kernel->end;
 
             unsigned long long dt = 0;
             dt = (((unsigned long long)t_stop) - ((unsigned long long)t_start));