title: CUDA Unified Memory: Difference in behaviour on Windows and Linux
body: <p>I am porting an application from Linux to Windows and discovered significant runtime differences of the same code on the same hardware between Windows and Linux.</p>
  <p>A minimal working example:</p>
  <pre class="lang-cpp prettyprint-override"><code>#include &lt;iostream&gt;
  #include &lt;chrono&gt;
  #include &lt;cuda.h&gt;
  
  constexpr unsigned int MB = 1000000;
  constexpr unsigned int num_bytes = 20 * MB;
  constexpr unsigned int repeats = 50;
  
  constexpr unsigned int the_answer = 42;
  constexpr unsigned int half_of_the_answer = the_answer / 2;
  constexpr unsigned int array_index = 100;
  
  
  __global__ void kernel(uint8_t* data){
      int i = blockIdx.x * blockDim.x + threadIdx.x;
      if(i&lt;num_bytes){
          data[i] = half_of_the_answer;
      }
  }
  
  void doSomethingOnGPU(uint8_t* data){
      cudaStream_t stream;
      cudaStreamCreate(&amp;stream);
      cudaStreamAttachMemAsync(stream, data, 0, cudaMemAttachSingle);
  
      kernel&lt;&lt;&lt;num_bytes/1000, 1000, 0, stream&gt;&gt;&gt;(data);
  
      cudaStreamSynchronize(stream);
      cudaStreamDestroy(stream);
      cudaDeviceSynchronize();
  }
  
  
  void doSomethingOnCPU(uint8_t* pic_unpacked){
      for(unsigned int i=0; i &lt; num_bytes; i++){
          pic_unpacked[i] = the_answer;
      }
  }
  
  int main() {
      uint8_t* data{};
      cudaMallocManaged(&amp;data, num_bytes, cudaMemAttachHost);
      for(unsigned int i=0;i&lt;repeats;i++){
          auto start_time_cpu = std::chrono::high_resolution_clock::now();
  
          doSomethingOnCPU(data);
  
          auto stop_time_cpu = std::chrono::high_resolution_clock::now();
          auto duration_cpu = std::chrono::duration_cast&lt;std::chrono::milliseconds&gt;(stop_time_cpu-start_time_cpu);
  
          std::cout &lt;&lt; &quot;CPU computation took &quot;&lt;&lt; duration_cpu.count() &lt;&lt; &quot;ms, data[&quot; &lt;&lt; array_index &lt;&lt; &quot;]=&quot;
                    &lt;&lt; static_cast&lt;unsigned int&gt;(data[array_index]) &lt;&lt; std::endl;
  
  
          auto start_time_gpu = std::chrono::high_resolution_clock::now();
  
          doSomethingOnGPU(data);
  
          auto stop_time_gpu = std::chrono::high_resolution_clock::now();
          auto duration_gpu = std::chrono::duration_cast&lt;std::chrono::milliseconds&gt;(stop_time_gpu-start_time_gpu);
  
          std::cout &lt;&lt; &quot;GPU computation took &quot;&lt;&lt; duration_gpu.count() &lt;&lt; &quot;ms, data[&quot; &lt;&lt; array_index &lt;&lt; &quot;]=&quot;
                    &lt;&lt; static_cast&lt;unsigned int&gt;(data[array_index]) &lt;&lt; std::endl &lt;&lt; std::endl;
      }
      cudaFree(data);
      return 0;
  }
  </code></pre>
  <p>This leads to the following output on Windows:</p>
  <pre><code>CPU computation took 216ms, data[100]=42
  GPU computation took 29ms, data[100]=21
  </code></pre>
  <p>and to the following output on Linux:</p>
  <pre><code>CPU computation took 20ms, data[100]=42
  GPU computation took 1ms, data[100]=21
  </code></pre>
  <p>Both are built in Release mode (Linux-&gt;GCC, Win-&gt;MSVC).</p>
  <p>It seems to me, that the automatic memory transfers do not work well under Windows.</p>
  <p>Explicit memory transfers with</p>
  <pre class="lang-cpp prettyprint-override"><code>cudaMallocHost(&amp;hostMem, size);
  cudaMalloc(&amp;cudaMem, size);
  cudaMemcpy(hostMem, cudaMem, size, cudaMemcpyDeviceToHost);
  cudaMemcpy(cudaMem, hostMem, size, cudaMemcpyHostToDevice);
  </code></pre>
  <p>work more or less with the same speed under Linux and Windows.</p>
  <p>Why is there this big runtime difference between Linux and Windows when working with unified memory?</p>