cudaMemcpy transfer kinds: Default vs HostToDevice/DeviceToHost

Question

cudaMemcpy allows programmers to explicitly specify the direction of memory transfer.

Is there any advantage of manually specifying memory transfer direction (cudaMemcpyDeviceToHost/cudaMemcpyHostToDevice/cudaMemcpyDeviceToDevice) instead of letting cuda automatically infer (cudaMemcpyDefault) from the pointer values?

Answer 1

tl;dr: Almost certainly no advantage.

cudaMemcpyDefault was added IIRC when GPUs started becoming capable of easily identifying the memory space by inspecting the address ("Unified virtual addressing"). Before that, you had to specify the direction. See, for example, the CUDA 3 documentation, accessible here. Look for cudaMemcpyKind in the API reference - no Default, just H2H, H2D, D2H and H2H.

When this changed, I guess it made sense to nVIDIA not to overload the function or name it differently, but just add a different constant value for the new capability.

I'm not 100% certain there's no difference, it's just very reasonable; and speaking from anecdotal personal experience, I've not seen any advantage/difference. Certainly the copying is not faster.

Answer 2

From the docs of cudaMemcpy():

[...] Passing cudaMemcpyDefault is recommended, in which case the type of transfer is inferred from the pointer values. However, cudaMemcpyDefault is only allowed on systems that support unified virtual addressing. [...]

Therefore if you have a GPU that allows unified virtual addressing, use cudaMemcpyDefault, otherwise you got no option than to be explicit.

You can query if your system supports it with

cudaGetDeviceProperties() with the device property cudaDeviceProp::unifiedAddressing.