title: Application crashes in nvcuda.dll after running for a long time
body: <p>I have a strange issue when my application crashes in nvcuda.dll after running for about 2 hours. After spending a lot of time trying to debug the issue, I think I have an idea what’s going on but I’d like to know if anybody else has seen this problem.</p>
  
  <p>My application launches most of its kernels in non-default streams and this process can go on for hours before there is a need to use default stream. Everything was working fine until I upgraded the drivers from some 320 version to most recent 332.50 (for K40m) version. What happens now is if the app runs for about 2 hours and then makes any call which uses default stream then it crashes during the call somewhere inside nvcuda.dll. At first I thought something is wrong with my kernels but it happens even if I use some basic stuff like cudaMemcpy (which uses default stream). The crash does not happen when the app is running for, say, 1 hour or 1.5 hours. It took me a while to realize that might be an issue with the driver so I uninstalled the new driver and installed the old one (320.92) and the problem was gone! I repeated the same process (changing the driver, rebooting then running the app again) multiple times and had 100% repro.</p>
  
  <p>Unfortunately, I don’t have a small, self-contained repro but before I try to create one, has anybody seen something like that recently? The entry from Event Viewer at the time of the crash does not say much:</p>
  
  <pre><code>Faulting application name: &lt;app&gt;.exe, version: &lt;version&gt;, time stamp: 0x5316a970
  Faulting module name: nvcuda.dll, version: 8.17.13.3250, time stamp: 0x52e1fa40
  Exception code: 0xc00000fd
  Fault offset: 0x00000000002226e7
  Faulting process id: 0x1558
  Faulting application start time: 0x01cf3831a2f3b71b
  Faulting application path: &lt;app&gt;.exe
  Faulting module path: C:\windows\SYSTEM32\nvcuda.dll
  Report Id: aceb9a51-a433-11e3-9403-90b11c4725be
  Faulting package full name: 
  Faulting package-relative application ID:
  </code></pre>
  
  <p><strong>Update 1</strong>: 
  I now have a simple application which reproduces the crash both on K20m and K40m cards. 
  <strong>Update 2</strong>: 
  Updated sample app, was able to repro the crash. From the call stack it looks like there is a stack overflow somewhere in nvcuda.dll.</p>
  
  <p>Steps:</p>
  
  <ol>
  <li>Install latest version (332.50) of the drivers on the machine.</li>
  <li>In Visual Studio 2012 create a new CUDA 5.5 project.</li>
  <li>Replace the contents of the kernel.cu with the code below.</li>
  <li>Compile and run the code on the machine with K20m or K40m.</li>
  <li>After approximately 2 hours of execution the app will crash and the entry below will be written into event log.</li>
  <li>Uninstall the driver and install previous (e.g. 321.10) version of the driver.</li>
  <li>Run the app, it should still be running after 2, 3 and more hours.</li>
  </ol>
  
  <p>Log:</p>
  
  <pre><code>Faulting application name: CudaTests60.exe, version: 0.0.0.0, time stamp: 0x5317974f
  Faulting module name: nvcuda.dll, version: 8.17.13.3250, time stamp: 0x52e1fa40
  Exception code: 0xc00000fd
  Fault offset: 0x000000000004f5cb
  Faulting process id: 0x23d0
  Faulting application start time: 0x01cf38ba16961e74
  Faulting application path: d:\bin\test\CudaTests60.exe
  Faulting module path: C:\windows\system32\nvcuda.dll
  Report Id: 192506c4-a4be-11e3-9401-90b11c4b02c0
  Faulting package full name: 
  Faulting package-relative application ID: 
  </code></pre>
  
  <p>Code:</p>
  
  <pre><code>#include "cuda_runtime.h"
  #include "device_launch_parameters.h"
  #include &lt;vector&gt;
  #include &lt;stdio.h&gt;
  #include &lt;assert.h&gt;
  #include &lt;cublas_v2.h&gt;
  #include &lt;thrust/host_vector.h&gt;
  #include &lt;thrust/device_vector.h&gt;
  #include &lt;Windows.h&gt;
  
  int main()
  {
      cudaError_t cudaStatus;
      {
          int crow = 10000;
          int ccol = 10000;
          int cshared = 10000;
          int xLength = crow * cshared;
          int yLength = cshared * ccol;
          int matLength = crow * ccol;
  
          thrust::device_vector&lt;float&gt; x(xLength);
          thrust::device_vector&lt;float&gt; y(yLength);
          thrust::device_vector&lt;float&gt; mat(matLength);
  
          thrust::fill(x.begin(), x.end(), 1.0f);
          thrust::fill(y.begin(), y.end(), 1.0f);
          thrust::fill(mat.begin(), mat.end(), .0f);
  
          cudaStream_t ops;
          cudaStatus = cudaStreamCreate(&amp;ops);
          assert(0 == cudaStatus);
  
          cublasHandle_t cbh;
          cublasStatus_t cbstatus;
          cbstatus = cublasCreate(&amp;cbh);
          assert(0 == cbstatus);
  
          cbstatus = cublasSetStream(cbh, ops);
          assert(0 == cbstatus);
  
          float alpha = 1;
          float beta = 0;
          float* px = thrust::raw_pointer_cast(x.data());
          float* py = thrust::raw_pointer_cast(y.data());
          float* pmat = thrust::raw_pointer_cast(mat.data());
          ULONGLONG start = GetTickCount64();
          ULONGLONG iter = 0;
          while (true)
          {
              cbstatus = cublasSgemm(cbh, CUBLAS_OP_N, CUBLAS_OP_N, crow, ccol, cshared, &amp;alpha, px, crow, py, cshared, &amp;beta, pmat, crow);
              assert(0 == cbstatus);
              if (0 != cbstatus)
              {
                  printf("cublasSgemm failed: %d.\n", cbstatus);
                  break;
              }
              cudaStatus = cudaStreamSynchronize(ops);
              assert(0 == cudaStatus);
              if (0 != cudaStatus)
              {
                  printf("cudaStreamSynchronize failed: %d.\n", cudaStatus);
                  break;
              }
  
              ULONGLONG cur = GetTickCount64();
              // Exit after 2 hours.
              if (cur - start &gt; 2 * 3600 * 1000)
                  break;
              iter++;
          }
  
          // Crash will happen here.
          printf("Before cudaMemcpy.\n");
          float res = 0;
          cudaStatus = cudaMemcpy(&amp;res, px, sizeof(float), cudaMemcpyDeviceToHost);
          assert(0 == cudaStatus);
          if (0 == cudaStatus)
              printf("After cudaMemcpy: %f\n", res);
          else
              printf("cudaMemcpy failed: %d\n", cudaStatus);
      }
  
      return 0;
  }
  </code></pre>