# streams not overlapping

**URL:** <https://forums.developer.nvidia.com/t/streams-not-overlapping/22529>\
**Category:** CUDA Programming and Performance\
**Created:** [May 19, 2011, 4:05pm UTC](https://forums.developer.nvidia.com/t/streams-not-overlapping/22529 "2011-05-19T16:05:26Z")\
**Posts on this page:** 2\
**Page:** 1

<div class="post-metadata">

**Author:** ![pmcr](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@pmcr](https://forums.developer.nvidia.com/u/pmcr)\
**Post date:** [May 19, 2011, 4:05pm UTC](https://forums.developer.nvidia.com/t/streams-not-overlapping/22529/1 "2011-05-19T16:05:26Z")

</div>

Hello, I have something very similar to the code:

```auto
int k, no_streams = 4;

cudaStream_t stream[no_streams];

for(k = 0; k < no_streams; k++) cudaStreamCreate(&stream[k]);

cudaMalloc(&g_in, size1*no_streams);

cudaMalloc(&g_out, size2*no_streams);

for (k = 0; k < no_streams; k++)

  cudaMemcpyAsync(g_in+k*size1/sizeof(float), h_ptr_in[k], size1, cudaMemcpyHostToDevice, stream[k]);

for (k = 0; k < no_streams; k++)

  mykernel<<<dimGrid, dimBlock, 0, stream[k]>>>(g_in+k*size1/sizeof(float), g_out+k*size2/sizeof(float));

for (k = 0; k < no_streams; k++)

  cudaMemcpyAsync(h_ptr_out[k], g_out+k*size2/sizeof(float), size2, cudaMemcpyDeviceToHost, stream[k]);

cudaThreadSynchronize();

cudaFree(g_in);

cudaFree(g_out);

```

‘h\_ptr\_in’ and ‘h\_ptr\_out’ are arrays of pointers allocated with cudaMallocHost (with no flags).

The problem is that the streams do not overlap.

In the visual profiler I can see the kernel execution from the first stream overlapping with the copy (H2D) from the second stream but nothing else overlaps.

I may not have resources to run 2 kernels (I think I do) but at least the kernel execution and copy should be overlaping, right?

And if I put all 3 (copy H2D, kernel execution, copy D2H) within the same for-loop none of them overlap…

Please HELP, what can be causing this?

I’m running on:

.Ubuntu 10.04 x64

.Device 0: “GeForce GTX 460”

CUDA Driver Version: 3.20

CUDA Runtime Version: 3.20

CUDA Capability Major/Minor version number: 2.1

Concurrent copy and execution: Yes

Concurrent kernel execution: Yes

---

<div class="post-metadata">

**Author:** ![fcs](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@fcs](https://forums.developer.nvidia.com/u/fcs)\
**Post date:** [May 23, 2011, 1:11pm UTC](https://forums.developer.nvidia.com/t/streams-not-overlapping/22529/2 "2011-05-23T13:11:49Z")

</div>

the kernel launch are blocking when profiling or debugging  
(with cuda 4, i saw an overlap in Computeprof )

If you want to test overlapping, try to time your calls with cudaEvents and compare the total time with the sum of partial time