how are texture references threated with respect to cuda streams?

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi,

texture texRef;
two buffers, buffer1 and buffer2 created with cudaMalloc()

CUDA Stream #1:

  • async cudaMemCpy to buffer 1
  • bind linear memory buffer1 to texRef
  • run kernel using texRef for read access

CUDA Stream #2

  • async cudaMemCpy to buffer 2
  • bind linear memory buffer2 to texRef
  • run kernel using texRef for read access

Buffer 1 contains different input data than buffer 2.

Assuming overlapping kernel execution: What happens if stream 2 changes the texture reference TexRef to buffer 2, while Stream 1 is currently executing the kernel? Will the running kernel suddenly see different texture data - or does a texture binding remain constant while the kernel runs, independently from any cudaBindTexture() calls that may happen in the mean time?

s/threated/treated in thread title. Can’t edit that, it seems.

Christian

2

Hmm, no one seems to know for sure.

To be on the safe side I will probably use a kernel template argument to switch the texture reference accessed by the kernel based on the CUDA stream we launch. Because currently we have up to 4 streams, this is quite a bit of extra code: unfortunately texture reference arrays and dynamic indexing are not supported.

Christian