prompt: You will be provided with a partial code base and an issue statement explaining a problem to resolve.
  <issue>
  NMS fails on any non-default GPU. 
  I'm creating an issue that corresponds to the problem that came up in #1603. I'm reposting the body of that issue here:
  
  I get an error when I try to run NMS code on any GPU except 0. 
  
  The issue is that I get RuntimeError: cuda runtime error (700) : an illegal memory access was encountered at mmdet/ops/nms/src/nms_kernel.cu:103 when I try to run NMS with a Tensor on any device except CPU or 0. The error happens on this line:
  
    THCudaCheck(cudaMemcpy(&mask_host[0],
                          mask_dev,
                          sizeof(unsigned long long) * boxes_num * col_blocks,
                          cudaMemcpyDeviceToHost));
  But I believe the issue is actually here:
  
    THCState *state = at::globalContext().lazyInitCUDA(); // TODO replace with getTHCState
  
    unsigned long long* mask_dev = NULL;
    //THCudaCheck(THCudaMalloc(state, (void**) &mask_dev,
    //                      boxes_num * col_blocks * sizeof(unsigned long long)));
  
    mask_dev = (unsigned long long*) THCudaMalloc(state, boxes_num * col_blocks * sizeof(unsigned long long));
  My guess is that THCudaMalloc is creating the mask_dev array on device 0 and not the device corresponding to the input at::Tensor boxes. It looks like state might encode which device a new cuda array is allocated on, so my intuition would be to try and grab the state from boxes. However, I'm not a CUDA expert, so I'm probably totally off base for how to use THCState objects. I was attempting to look through the pytorch docs / source to see if I could figure something out, but I'm not having any luck.
  
  Any pointers on how this issue might be handled would be appreciated. Note that if you have two GPUs you can reproduce the error by checking out this PR and running: xdoctest -m tests/test_nms.py test_nms_device_and_dtypes_gpu
  
  
  </issue>
  <code>
  [start of mmdet/ops/nms/nms_wrapper.py]
  1 import numpy as np
  2 import torch
  3 
  4 from . import nms_cpu, nms_cuda
  5 from .soft_nms_cpu import soft_nms_cpu
  6 
  7 
  8 def nms(dets, iou_thr, device_id=None):
  9     """Dispatch to either CPU or GPU NMS implementations.
  10 
  11     The input can be either a torch tensor or numpy array. GPU NMS will be used
  12     if the input is a gpu tensor or device_id is specified, otherwise CPU NMS
  13     will be used. The returned type will always be the same as inputs.
  14 
  15     Arguments:
  16         dets (torch.Tensor or np.ndarray): bboxes with scores.
  17         iou_thr (float): IoU threshold for NMS.
  18         device_id (int, optional): when `dets` is a numpy array, if `device_id`
  19             is None, then cpu nms is used, otherwise gpu_nms will be used.
  20 
  21     Returns:
  22         tuple: kept bboxes and indice, which is always the same data type as
  23             the input.
  24     """
  25     # convert dets (tensor or numpy array) to tensor
  26     if isinstance(dets, torch.Tensor):
  27         is_numpy = False
  28         dets_th = dets
  29     elif isinstance(dets, np.ndarray):
  30         is_numpy = True
  31         device = 'cpu' if device_id is None else 'cuda:{}'.format(device_id)
  32         dets_th = torch.from_numpy(dets).to(device)
  33     else:
  34         raise TypeError(
  35             'dets must be either a Tensor or numpy array, but got {}'.format(
  36                 type(dets)))
  37 
  38     # execute cpu or cuda nms
  39     if dets_th.shape[0] == 0:
  40         inds = dets_th.new_zeros(0, dtype=torch.long)
  41     else:
  42         if dets_th.is_cuda:
  43             inds = nms_cuda.nms(dets_th, iou_thr)
  44         else:
  45             inds = nms_cpu.nms(dets_th, iou_thr)
  46 
  47     if is_numpy:
  48         inds = inds.cpu().numpy()
  49     return dets[inds, :], inds
  50 
  51 
  52 def soft_nms(dets, iou_thr, method='linear', sigma=0.5, min_score=1e-3):
  53     if isinstance(dets, torch.Tensor):
  54         is_tensor = True
  55         dets_np = dets.detach().cpu().numpy()
  56     elif isinstance(dets, np.ndarray):
  57         is_tensor = False
  58         dets_np = dets
  59     else:
  60         raise TypeError(
  61             'dets must be either a Tensor or numpy array, but got {}'.format(
  62                 type(dets)))
  63 
  64     method_codes = {'linear': 1, 'gaussian': 2}
  65     if method not in method_codes:
  66         raise ValueError('Invalid method for SoftNMS: {}'.format(method))
  67     new_dets, inds = soft_nms_cpu(
  68         dets_np,
  69         iou_thr,
  70         method=method_codes[method],
  71         sigma=sigma,
  72         min_score=min_score)
  73 
  74     if is_tensor:
  75         return dets.new_tensor(new_dets), dets.new_tensor(
  76             inds, dtype=torch.long)
  77     else:
  78         return new_dets.astype(np.float32), inds.astype(np.int64)
  79
  [end of mmdet/ops/nms/nms_wrapper.py]
  </code>
  
  I need you to solve this issue by generating a single patch file that I can apply directly to this repository using git apply. Please respond with a single patch file in the following format.
  <patch>
  diff --git a/file.py b/file.py
  --- a/file.py
  +++ b/file.py
  @@ -1,27 +1,35 @@
   def euclidean(a, b):
  -    while b:
  -        a, b = b, a % b
  -    return a
  +    if b == 0:
  +        return a
  +    return euclidean(b, a % b)
   
   
   def bresenham(x0, y0, x1, y1):
       points = []
       dx = abs(x1 - x0)
       dy = abs(y1 - y0)
  -    sx = 1 if x0 < x1 else -1
  -    sy = 1 if y0 < y1 else -1
  -    err = dx - dy
  +    x, y = x0, y0
  +    sx = -1 if x0 > x1 else 1
  +    sy = -1 if y0 > y1 else 1
   
  -    while True:
  -        points.append((x0, y0))
  -        if x0 == x1 and y0 == y1:
  -            break
  -        e2 = 2 * err
  -        if e2 > -dy:
  -            err -= dy
  -            x0 += sx
  -        if e2 < dx:
  -            err += dx
  -            y0 += sy
  +    if dx > dy:
  +        err = dx / 2.0
  +        while x != x1:
  +            points.append((x, y))
  +            err -= dy
  +            if err < 0:
  +                y += sy
  +                err += dx
  +            x += sx
  +    else:
  +        err = dy / 2.0
  +        while y != y1:
  +            points.append((x, y))
  +            err -= dx
  +            if err < 0:
  +                x += sx
  +                err += dy
  +            y += sy
  +
  +    points.append((x, y))
       return points
  
  </patch>
golden_diff: diff --git a/mmdet/ops/nms/nms_wrapper.py b/mmdet/ops/nms/nms_wrapper.py
  --- a/mmdet/ops/nms/nms_wrapper.py
  +++ b/mmdet/ops/nms/nms_wrapper.py
  @@ -21,6 +21,18 @@
       Returns:
           tuple: kept bboxes and indice, which is always the same data type as
               the input.
  +
  +    Example:
  +        >>> dets = np.array([[49.1, 32.4, 51.0, 35.9, 0.9],
  +        >>>                  [49.3, 32.9, 51.0, 35.3, 0.9],
  +        >>>                  [49.2, 31.8, 51.0, 35.4, 0.5],
  +        >>>                  [35.1, 11.5, 39.1, 15.7, 0.5],
  +        >>>                  [35.6, 11.8, 39.3, 14.2, 0.5],
  +        >>>                  [35.3, 11.5, 39.9, 14.5, 0.4],
  +        >>>                  [35.2, 11.7, 39.7, 15.7, 0.3]], dtype=np.float32)
  +        >>> iou_thr = 0.7
  +        >>> supressed, inds = nms(dets, iou_thr)
  +        >>> assert len(inds) == len(supressed) == 3
       """
       # convert dets (tensor or numpy array) to tensor
       if isinstance(dets, torch.Tensor):
  @@ -50,6 +62,18 @@
   
   
   def soft_nms(dets, iou_thr, method='linear', sigma=0.5, min_score=1e-3):
  +    """
  +    Example:
  +        >>> dets = np.array([[4., 3., 5., 3., 0.9],
  +        >>>                  [4., 3., 5., 4., 0.9],
  +        >>>                  [3., 1., 3., 1., 0.5],
  +        >>>                  [3., 1., 3., 1., 0.5],
  +        >>>                  [3., 1., 3., 1., 0.4],
  +        >>>                  [3., 1., 3., 1., 0.0]], dtype=np.float32)
  +        >>> iou_thr = 0.7
  +        >>> supressed, inds = soft_nms(dets, iou_thr, sigma=0.5)
  +        >>> assert len(inds) == len(supressed) == 3
  +    """
       if isinstance(dets, torch.Tensor):
           is_tensor = True
           dets_np = dets.detach().cpu().numpy()
  
verification_info: {"golden_diff": "diff --git a/mmdet/ops/nms/nms_wrapper.py b/mmdet/ops/nms/nms_wrapper.py\n--- a/mmdet/ops/nms/nms_wrapper.py\n+++ b/mmdet/ops/nms/nms_wrapper.py\n@@ -21,6 +21,18 @@\n     Returns:\n         tuple: kept bboxes and indice, which is always the same data type as\n             the input.\n+\n+    Example:\n+        >>> dets = np.array([[49.1, 32.4, 51.0, 35.9, 0.9],\n+        >>>                  [49.3, 32.9, 51.0, 35.3, 0.9],\n+        >>>                  [49.2, 31.8, 51.0, 35.4, 0.5],\n+        >>>                  [35.1, 11.5, 39.1, 15.7, 0.5],\n+        >>>                  [35.6, 11.8, 39.3, 14.2, 0.5],\n+        >>>                  [35.3, 11.5, 39.9, 14.5, 0.4],\n+        >>>                  [35.2, 11.7, 39.7, 15.7, 0.3]], dtype=np.float32)\n+        >>> iou_thr = 0.7\n+        >>> supressed, inds = nms(dets, iou_thr)\n+        >>> assert len(inds) == len(supressed) == 3\n     \"\"\"\n     # convert dets (tensor or numpy array) to tensor\n     if isinstance(dets, torch.Tensor):\n@@ -50,6 +62,18 @@\n \n \n def soft_nms(dets, iou_thr, method='linear', sigma=0.5, min_score=1e-3):\n+    \"\"\"\n+    Example:\n+        >>> dets = np.array([[4., 3., 5., 3., 0.9],\n+        >>>                  [4., 3., 5., 4., 0.9],\n+        >>>                  [3., 1., 3., 1., 0.5],\n+        >>>                  [3., 1., 3., 1., 0.5],\n+        >>>                  [3., 1., 3., 1., 0.4],\n+        >>>                  [3., 1., 3., 1., 0.0]], dtype=np.float32)\n+        >>> iou_thr = 0.7\n+        >>> supressed, inds = soft_nms(dets, iou_thr, sigma=0.5)\n+        >>> assert len(inds) == len(supressed) == 3\n+    \"\"\"\n     if isinstance(dets, torch.Tensor):\n         is_tensor = True\n         dets_np = dets.detach().cpu().numpy()\n", "issue": "NMS fails on any non-default GPU. \nI'm creating an issue that corresponds to the problem that came up in #1603. I'm reposting the body of that issue here:\r\n\r\nI get an error when I try to run NMS code on any GPU except 0. \r\n\r\nThe issue is that I get RuntimeError: cuda runtime error (700) : an illegal memory access was encountered at mmdet/ops/nms/src/nms_kernel.cu:103 when I try to run NMS with a Tensor on any device except CPU or 0. The error happens on this line:\r\n\r\n  THCudaCheck(cudaMemcpy(&mask_host[0],\r\n                        mask_dev,\r\n                        sizeof(unsigned long long) * boxes_num * col_blocks,\r\n                        cudaMemcpyDeviceToHost));\r\nBut I believe the issue is actually here:\r\n\r\n  THCState *state = at::globalContext().lazyInitCUDA(); // TODO replace with getTHCState\r\n\r\n  unsigned long long* mask_dev = NULL;\r\n  //THCudaCheck(THCudaMalloc(state, (void**) &mask_dev,\r\n  //                      boxes_num * col_blocks * sizeof(unsigned long long)));\r\n\r\n  mask_dev = (unsigned long long*) THCudaMalloc(state, boxes_num * col_blocks * sizeof(unsigned long long));\r\nMy guess is that THCudaMalloc is creating the mask_dev array on device 0 and not the device corresponding to the input at::Tensor boxes. It looks like state might encode which device a new cuda array is allocated on, so my intuition would be to try and grab the state from boxes. However, I'm not a CUDA expert, so I'm probably totally off base for how to use THCState objects. I was attempting to look through the pytorch docs / source to see if I could figure something out, but I'm not having any luck.\r\n\r\nAny pointers on how this issue might be handled would be appreciated. Note that if you have two GPUs you can reproduce the error by checking out this PR and running: xdoctest -m tests/test_nms.py test_nms_device_and_dtypes_gpu\r\n\n", "before_files": [{"content": "import numpy as np\nimport torch\n\nfrom . import nms_cpu, nms_cuda\nfrom .soft_nms_cpu import soft_nms_cpu\n\n\ndef nms(dets, iou_thr, device_id=None):\n    \"\"\"Dispatch to either CPU or GPU NMS implementations.\n\n    The input can be either a torch tensor or numpy array. GPU NMS will be used\n    if the input is a gpu tensor or device_id is specified, otherwise CPU NMS\n    will be used. The returned type will always be the same as inputs.\n\n    Arguments:\n        dets (torch.Tensor or np.ndarray): bboxes with scores.\n        iou_thr (float): IoU threshold for NMS.\n        device_id (int, optional): when `dets` is a numpy array, if `device_id`\n            is None, then cpu nms is used, otherwise gpu_nms will be used.\n\n    Returns:\n        tuple: kept bboxes and indice, which is always the same data type as\n            the input.\n    \"\"\"\n    # convert dets (tensor or numpy array) to tensor\n    if isinstance(dets, torch.Tensor):\n        is_numpy = False\n        dets_th = dets\n    elif isinstance(dets, np.ndarray):\n        is_numpy = True\n        device = 'cpu' if device_id is None else 'cuda:{}'.format(device_id)\n        dets_th = torch.from_numpy(dets).to(device)\n    else:\n        raise TypeError(\n            'dets must be either a Tensor or numpy array, but got {}'.format(\n                type(dets)))\n\n    # execute cpu or cuda nms\n    if dets_th.shape[0] == 0:\n        inds = dets_th.new_zeros(0, dtype=torch.long)\n    else:\n        if dets_th.is_cuda:\n            inds = nms_cuda.nms(dets_th, iou_thr)\n        else:\n            inds = nms_cpu.nms(dets_th, iou_thr)\n\n    if is_numpy:\n        inds = inds.cpu().numpy()\n    return dets[inds, :], inds\n\n\ndef soft_nms(dets, iou_thr, method='linear', sigma=0.5, min_score=1e-3):\n    if isinstance(dets, torch.Tensor):\n        is_tensor = True\n        dets_np = dets.detach().cpu().numpy()\n    elif isinstance(dets, np.ndarray):\n        is_tensor = False\n        dets_np = dets\n    else:\n        raise TypeError(\n            'dets must be either a Tensor or numpy array, but got {}'.format(\n                type(dets)))\n\n    method_codes = {'linear': 1, 'gaussian': 2}\n    if method not in method_codes:\n        raise ValueError('Invalid method for SoftNMS: {}'.format(method))\n    new_dets, inds = soft_nms_cpu(\n        dets_np,\n        iou_thr,\n        method=method_codes[method],\n        sigma=sigma,\n        min_score=min_score)\n\n    if is_tensor:\n        return dets.new_tensor(new_dets), dets.new_tensor(\n            inds, dtype=torch.long)\n    else:\n        return new_dets.astype(np.float32), inds.astype(np.int64)\n", "path": "mmdet/ops/nms/nms_wrapper.py"}]}