Storage size fixes + docker updates + cupti fixes

docker/config/apt.sh

@@ -56,7 +56,18 @@ fi
 
 DISPLAY_PACKAGES="xserver-xorg freeglut3-dev libx11-dev libx11-xcb-dev libxpm-dev libxft-dev libxmu-dev libxv-dev libxrandr-dev \
     libglew-dev libftgl-dev libxkbcommon-x11-dev libxrender-dev libxxf86vm-dev libxinerama-dev qt5-default \
-    qtcreator emacs-nox vim-nox"
+    emacs-nox vim-nox"
+CUDA_VER=$(dpkg --get-selections | grep cuda-cudart- | awk '{print $1}' | head -n 1 | sed 's/cuda-cudart-//g')
+
+#-----------------------------------------------------------------------------#
+#
+#   CUDA nsight tools
+#
+#-----------------------------------------------------------------------------#
+
+if [ -n "${CUDA_VER}" ]; then
+    run-verbose apt-get install -y cuda-nsight-{compute,systems}-${CUDA_VER}
+fi
 
 #-----------------------------------------------------------------------------#
 #

docker/config/etc/profile.d/cuda.sh

@@ -1,4 +1,12 @@
 #!/bin/sh
 
-PATH=/usr/local/cuda/bin:${PATH}
+_CUDA=/usr/local/cuda/bin
+_NSIGHT_SYSTEMS=/usr/local/cuda/NsightSystems-2019.3/Target-x86_64/x86_64
+_NSIGHT_COMPUTE=/usr/local/cuda/NsightCompute-2019.3
+
+PATH=${_CUDA}:${_NSIGHT_SYSTEMS}:${_NSIGHT_COMPUTE}:${PATH}
 export PATH
+
+unset _CUDA
+unset _NSIGHT_SYSTEMS
+unset _NSIGHT_COMPUTE

source/timemory/backends/cupti.hpp

@@ -127,7 +127,7 @@ static void CUPTIAPI
 //--------------------------------------------------------------------------------------//
 
 static void CUPTIAPI
-            buffer_completed(CUcontext ctx, uint32_t streamId, uint8_t* buffer, size_t size,
+            buffer_completed(CUcontext /* ctx */, uint32_t streamId, uint8_t* buffer, size_t /* size */,
                              size_t validSize)
 {
     CUpti_Activity* record = nullptr;
@@ -427,9 +427,9 @@ static void CUPTIAPI
         k_data.m_name        = corr_data;
         auto& pass_data      = k_data.m_pass_data;
 
-        for(int j = 0; j < pass_data.size(); ++j)
+        for(size_t j = 0; j < pass_data.size(); ++j)
         {
-            for(int i = 0; i < pass_data[j].event_groups->numEventGroups; i++)
+            for(uint32_t i = 0; i < pass_data[j].event_groups->numEventGroups; i++)
             {
                 _LOG("  Enabling group %d", i);
                 uint32_t all = 1;
@@ -446,12 +446,12 @@ static void CUPTIAPI
     else if(cbInfo->callbackSite == CUPTI_API_EXIT)
     {
         auto& current_kernel = (*kernel_data)[corr_data];
-        for(int current_pass = 0; current_pass < current_kernel.m_pass_data.size();
+        for(size_t current_pass = 0; current_pass < current_kernel.m_pass_data.size();
             ++current_pass)
         {
             auto& pass_data = current_kernel.m_pass_data[current_pass];
 
-            for(int i = 0; i < pass_data.event_groups->numEventGroups; i++)
+            for(uint32_t i = 0; i < pass_data.event_groups->numEventGroups; i++)
             {
                 CUpti_EventGroup    group = pass_data.event_groups->eventGroups[i];
                 CUpti_EventDomainID group_domain;
@@ -712,7 +712,7 @@ struct profiler
             _LOG("[metric] Looking at set (pass) %d", i);
             uint32_t num_events      = 0;
             size_t   num_events_size = sizeof(num_events);
-            for(int j = 0; j < m_metric_pass_data->sets[i].numEventGroups; ++j)
+            for(uint32_t j = 0; j < m_metric_pass_data->sets[i].numEventGroups; ++j)
             {
                 CUPTI_CALL(cuptiEventGroupGetAttribute(
                     m_metric_pass_data->sets[i].eventGroups[j],
@@ -730,7 +730,7 @@ struct profiler
             _LOG("[event] Looking at set (pass) %d", i);
             uint32_t num_events      = 0;
             size_t   num_events_size = sizeof(num_events);
-            for(int j = 0; j < m_event_pass_data->sets[i].numEventGroups; ++j)
+            for(uint32_t j = 0; j < m_event_pass_data->sets[i].numEventGroups; ++j)
             {
                 CUPTI_CALL(cuptiEventGroupGetAttribute(
                     m_event_pass_data->sets[i].eventGroups[j],
@@ -744,7 +744,9 @@ struct profiler
 
         m_kernel_data[impl::dummy_kernel_id] = dummy_data;
         cuptiEnableKernelReplayMode(m_context);
+#if defined(__NVCC__)
         impl::warmup<<<1, 1>>>();
+#endif
     }
 
     ~profiler()

source/timemory/components/cupti_event.hpp

@@ -174,9 +174,9 @@ struct cupti_event : public base<cupti_event, cupti::profiler::results_t>
         {
             switch(itr.index)
             {
-                case 0: values.push_back(std::get<0>(obj.data)); break;
-                case 1: values.push_back(std::get<1>(obj.data)); break;
-                case 2: values.push_back(std::get<2>(obj.data)); break;
+                case 0: values.push_back(std::get<0>(itr.data)); break;
+                case 1: values.push_back(std::get<1>(itr.data)); break;
+                case 2: values.push_back(std::get<2>(itr.data)); break;
             }
         }
         return values;

source/timemory/utility/storage.hpp

@@ -394,7 +394,10 @@ class storage
     }
 
     void print();
-    bool empty() const { return (m_data.graph().size() == 0); }
+
+    // there is always a head node that should not be counted
+    bool          empty() const { return (m_data.graph().size() <= 1); }
+    inline size_t size() const { return m_data.graph().size() - 1; }
 
     const graph_data& data() const { return m_data; }
     const graph_t&    graph() const { return m_data.graph(); }
@@ -404,8 +407,6 @@ class storage
     iterator&   current() { return m_data.current(); }
     graph_t&    graph() { return m_data.graph(); }
 
-    inline size_t size() const { return m_data.graph().size(); }
-
 public:
     //----------------------------------------------------------------------------------//
     //