Why is cudaPointerGetAttributes() returning invalid argument for host pointer?

Question

I want to write a function that tells me if a pointer is a host or device pointer. This is essentially a wrapper around cudaPointerGetAttributes() that returns either 1 or 0 if the pointer is for the device or not.

What I can't understand is why cudaPointerGetAttributes fails my error checking by returning invalid argument when I'm testing a host pointer. An example is provided below.

#include <stdio.h>
#include <stdlib.h>

#define CUDA_ERROR_CHECK(fun)                                                                   \
do{                                                                                             \
    cudaError_t err = fun;                                                                      \
    if(err != cudaSuccess)                                                                      \
    {                                                                                           \
      fprintf(stderr, "Cuda error %d %s:: %s\n", __LINE__, __func__, cudaGetErrorString(err));  \
      exit(EXIT_FAILURE);                                                                       \
    }                                                                                           \
}while(0);

int is_device_pointer(const void *ptr)
{
  int is_device_ptr = 0;
  cudaPointerAttributes attributes;

  CUDA_ERROR_CHECK(cudaPointerGetAttributes(&attributes, ptr));

  if(attributes.devicePointer != NULL)
  {
    is_device_ptr = 1;
  }

  return is_device_ptr;
}

int main()
{
  int *host_ptr, x = 0;
  int is_dev_ptr;

  host_ptr = &x;

  int *dev_ptr;
  cudaMalloc((void **)&dev_ptr, 16);

  //is_dev_ptr = is_device_pointer((const void *)host_ptr); //Causes invalid argument
  is_dev_ptr = is_device_pointer((const void *)dev_ptr);  //Works

  if(is_dev_ptr == 1)
  {
    fprintf(stdout, "Device pointer\n");
  }
  else
  {
    fprintf(stdout, "Not device Pointer\n");
  }

  CUDA_ERROR_CHECK(cudaFree((void *)dev_ptr));
  CUDA_ERROR_CHECK(cudaDeviceReset());

  return EXIT_SUCCESS;
}

Answer 1

This is expected behavior. cudaPointerGetAttributes can only introspect pointers that have been recorded in some fashion with the CUDA runtime API. Refer to the documentation:

If pointer was not allocated in, mapped by or registered with context supporting unified addressing cudaErrorInvalidValue is returned.

What this is saying is that the pointer must have been returned or passed through an API such as cudaMalloc, cudaMallocManaged, cudaHostRegister, etc. for it to be "recognized" by cudaPointerGetAttributes. You must be in a UVA regime, and you must have acquired the pointer using an appropriate method.

In your case, passing a bare host pointer this way doesn't meet the requirements spelled out in the documentation, so the error return is expected.

This particular error return code is a "non-sticky" CUDA error, meaning it can be cleared out via cudaGetLastError(). In my view, it should be safe to interpret this error return code as "this is an ordinary host pointer". But of course, if you pass a garbage value, or an unallocated pointer, you will get the same error code.