@@ -84,7 +84,11 @@ the functionality of the selected device:
 
    .. attribute:: name
 
-      The name of the device (e.g. "GeForce GTX 970")
+      The name of the device (e.g. "GeForce GTX 970").
+
+   .. attribute:: uuid
+
+      The UUID of the device (e.g. "GPU-e6489c45-5b68-3b03-bab7-0e7c8e809643").
 
    .. method:: reset
 
 
@@ -74,3 +74,19 @@ which the current GPU context can also be retrieved:
     :members: current
     :noindex:
 
+
+Device UUIDs
+============
+
+The UUID of a device (equal to that returned by ``nvidia-smi -L``) is available
+in the :attr:`uuid <numba.cuda.cudadrv.driver.Device.uuid>` attribute of a CUDA
+device object.
+
+For example, to obtain the UUID of the current device:
+
+.. code-block:: python
+
+   dev = cuda.current_context().device
+   # prints e.g. "GPU-e6489c45-5b68-3b03-bab7-0e7c8e809643"
+   print(dev.uuid)
+
@@ -33,7 +33,7 @@
 from numba.core import utils, serialize, config
 from .error import CudaSupportError, CudaDriverError
 from .drvapi import API_PROTOTYPES
-from .drvapi import cu_occupancy_b2d_size, cu_stream_callback_pyobj
+from .drvapi import cu_occupancy_b2d_size, cu_stream_callback_pyobj, cu_uuid
 from numba.cuda.cudadrv import enums, drvapi, _extras
 from numba.cuda.envvars import get_numba_envvar
 
@@ -462,17 +462,30 @@ def __init__(self, devnum):
         assert devnum == got_devnum.value, "Driver returned another device"
         self.id = got_devnum.value
         self.attributes = {}
+
         # Read compute capability
         cc_major = c_int()
         cc_minor = c_int()
         driver.cuDeviceComputeCapability(byref(cc_major), byref(cc_minor),
                                          self.id)
         self.compute_capability = (cc_major.value, cc_minor.value)
+
         # Read name
         bufsz = 128
         buf = (c_char * bufsz)()
         driver.cuDeviceGetName(buf, bufsz, self.id)
         self.name = buf.value
+
+        # Read UUID
+        uuid = cu_uuid()
+        driver.cuDeviceGetUuid(byref(uuid), self.id)
+        b = '%02x'
+        b2 = b * 2
+        b4 = b * 4
+        b6 = b * 6
+        fmt = f'GPU-{b4}-{b2}-{b2}-{b2}-{b6}'
+        self.uuid = fmt % tuple(bytes(uuid))
+
         self.primary_context = None
 
     def get_device_identity(self):
 
@@ -16,6 +16,7 @@
 cu_link_state = c_void_p
 cu_function_attribute = c_int
 cu_ipc_mem_handle = (c_byte * _extras.CUDA_IPC_HANDLE_SIZE)   # 64 bytes wide
+cu_uuid = (c_byte * 16)         # Device UUID
 
 cu_stream_callback_pyobj = CFUNCTYPE(None, cu_stream, c_int, py_object)
 
@@ -388,4 +389,6 @@
     'cuDeviceCanAccessPeer': (c_int,
                               POINTER(c_int), cu_device, cu_device),
 
+    # CUresult cuDeviceGetUuid ( CUuuid* uuid, CUdevice dev )
+    'cuDeviceGetUuid': (c_int, POINTER(cu_uuid), cu_device),
 }
@@ -4,13 +4,19 @@
 _MemoryInfo = namedtuple("_MemoryInfo", "free,total")
 
 
-class FakeCUDAContext(object):
+class FakeCUDADevice:
+    def __init__(self):
+        self.uuid = 'GPU-00000000-0000-0000-0000-000000000000'
+
+
+class FakeCUDAContext:
     '''
     This stub implements functionality only for simulating a single GPU
     at the moment.
     '''
-    def __init__(self, device):
-        self._device = device
+    def __init__(self, device_id):
+        self._device_id = device_id
+        self._device = FakeCUDADevice()
 
     def __enter__(self):
         pass
@@ -23,6 +29,10 @@ def __str__(self):
 
     @property
     def id(self):
+        return self._device_id
+
+    @property
+    def device(self):
         return self._device
 
     @property
@@ -53,7 +63,7 @@ def memhostalloc(self, sz, mapped=False, portable=False, wc=False):
         return self.memalloc(sz)
 
 
-class FakeDeviceList(object):
+class FakeDeviceList:
     '''
     This stub implements a device list containing a single GPU. It also
     keeps track of the GPU status, i.e. whether the context is closed or not,
 
@@ -5,6 +5,7 @@
 from numba.cuda.testing import unittest, CUDATestCase
 from numba.cuda.testing import skip_on_cudasim
 
+
 ptx1 = '''
     .version 1.4
     .target sm_10, map_f64_to_f32
@@ -190,5 +191,28 @@ def b2d(bs):
         self.assertTrue(block > 0)
 
 
+class TestDevice(CUDATestCase):
+    def test_device_get_uuid(self):
+        # A device UUID looks like:
+        #
+        #     GPU-e6489c45-5b68-3b03-bab7-0e7c8e809643
+        #
+        # To test, we construct an RE that matches this form and verify that
+        # the returned UUID matches.
+        #
+        # Device UUIDs may not conform to parts of the UUID specification (RFC
+        # 4122) pertaining to versions and variants, so we do not extract and
+        # validate the values of these bits.
+
+        h = '[0-9a-f]{%d}'
+        h4 = h % 4
+        h8 = h % 8
+        h12 = h % 12
+        uuid_format = f'^GPU-{h8}-{h4}-{h4}-{h4}-{h12}$'
+
+        dev = devices.get_context().device
+        self.assertRegex(dev.uuid, uuid_format)
+
+
 if __name__ == '__main__':
     unittest.main()