Use Tensor.data_ptr instead of .data (#1262)

* Use Tensor.data_ptr instead of .data

* use pytorch-nightly in CI

.travis.yml

@@ -26,7 +26,7 @@ before_install:
   # Useful for debugging any issues with conda
   - conda info -a
 
-  - conda create -q -n test-environment python=$TRAVIS_PYTHON_VERSION pytorch scipy -c pytorch
+  - conda create -q -n test-environment python=$TRAVIS_PYTHON_VERSION pytorch scipy -c pytorch-nightly
   - source activate test-environment
   - |
     if [[ "$IMAGE_BACKEND" == "Pillow-SIMD" ]]; then

torchvision/csrc/cpu/ROIAlign_cpu.cpp

@@ -406,7 +406,7 @@ at::Tensor ROIAlign_forward_cpu(
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(input.type(), "ROIAlign_forward", [&] {
     ROIAlignForward<scalar_t>(
         output_size,
-        input.contiguous().data<scalar_t>(),
+        input.contiguous().data_ptr<scalar_t>(),
         spatial_scale,
         channels,
         height,
@@ -414,8 +414,8 @@ at::Tensor ROIAlign_forward_cpu(
         pooled_height,
         pooled_width,
         sampling_ratio,
-        rois.contiguous().data<scalar_t>(),
-        output.data<scalar_t>());
+        rois.contiguous().data_ptr<scalar_t>(),
+        output.data_ptr<scalar_t>());
   });
   return output;
 }
@@ -456,7 +456,7 @@ at::Tensor ROIAlign_backward_cpu(
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(grad.type(), "ROIAlign_forward", [&] {
     ROIAlignBackward<scalar_t>(
         grad.numel(),
-        grad.data<scalar_t>(),
+        grad.data_ptr<scalar_t>(),
         spatial_scale,
         channels,
         height,
@@ -464,8 +464,8 @@ at::Tensor ROIAlign_backward_cpu(
         pooled_height,
         pooled_width,
         sampling_ratio,
-        grad_input.data<scalar_t>(),
-        rois.contiguous().data<scalar_t>(),
+        grad_input.data_ptr<scalar_t>(),
+        rois.contiguous().data_ptr<scalar_t>(),
         n_stride,
         c_stride,
         h_stride,

torchvision/csrc/cpu/ROIPool_cpu.cpp

@@ -151,17 +151,17 @@ std::tuple<at::Tensor, at::Tensor> ROIPool_forward_cpu(
 
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(input.type(), "ROIPool_forward", [&] {
     RoIPoolForward<scalar_t>(
-        input.contiguous().data<scalar_t>(),
+        input.contiguous().data_ptr<scalar_t>(),
         spatial_scale,
         channels,
         height,
         width,
         pooled_height,
         pooled_width,
-        rois.contiguous().data<scalar_t>(),
+        rois.contiguous().data_ptr<scalar_t>(),
         num_rois,
-        output.data<scalar_t>(),
-        argmax.data<int>());
+        output.data_ptr<scalar_t>(),
+        argmax.data_ptr<int>());
   });
   return std::make_tuple(output, argmax);
 }
@@ -205,16 +205,16 @@ at::Tensor ROIPool_backward_cpu(
 
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(grad.type(), "ROIPool_backward", [&] {
     RoIPoolBackward<scalar_t>(
-        grad.data<scalar_t>(),
-        argmax.data<int>(),
+        grad.data_ptr<scalar_t>(),
+        argmax.data_ptr<int>(),
         num_rois,
         channels,
         height,
         width,
         pooled_height,
         pooled_width,
-        grad_input.data<scalar_t>(),
-        rois.contiguous().data<scalar_t>(),
+        grad_input.data_ptr<scalar_t>(),
+        rois.contiguous().data_ptr<scalar_t>(),
         n_stride,
         c_stride,
         h_stride,

torchvision/csrc/cpu/nms_cpu.cpp

@@ -26,14 +26,14 @@ at::Tensor nms_cpu_kernel(
   at::Tensor suppressed_t = at::zeros({ndets}, dets.options().dtype(at::kByte));
   at::Tensor keep_t = at::zeros({ndets}, dets.options().dtype(at::kLong));
 
-  auto suppressed = suppressed_t.data<uint8_t>();
-  auto keep = keep_t.data<int64_t>();
-  auto order = order_t.data<int64_t>();
-  auto x1 = x1_t.data<scalar_t>();
-  auto y1 = y1_t.data<scalar_t>();
-  auto x2 = x2_t.data<scalar_t>();
-  auto y2 = y2_t.data<scalar_t>();
-  auto areas = areas_t.data<scalar_t>();
+  auto suppressed = suppressed_t.data_ptr<uint8_t>();
+  auto keep = keep_t.data_ptr<int64_t>();
+  auto order = order_t.data_ptr<int64_t>();
+  auto x1 = x1_t.data_ptr<scalar_t>();
+  auto y1 = y1_t.data_ptr<scalar_t>();
+  auto x2 = x2_t.data_ptr<scalar_t>();
+  auto y2 = y2_t.data_ptr<scalar_t>();
+  auto areas = areas_t.data_ptr<scalar_t>();
 
   int64_t num_to_keep = 0;
 

torchvision/csrc/cuda/ROIAlign_cuda.cu

@@ -340,7 +340,7 @@ at::Tensor ROIAlign_forward_cuda(
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(input.type(), "ROIAlign_forward", [&] {
     RoIAlignForward<scalar_t><<<grid, block, 0, stream>>>(
         output_size,
-        input.contiguous().data<scalar_t>(),
+        input.contiguous().data_ptr<scalar_t>(),
         spatial_scale,
         channels,
         height,
@@ -348,8 +348,8 @@ at::Tensor ROIAlign_forward_cuda(
         pooled_height,
         pooled_width,
         sampling_ratio,
-        rois.contiguous().data<scalar_t>(),
-        output.data<scalar_t>());
+        rois.contiguous().data_ptr<scalar_t>(),
+        output.data_ptr<scalar_t>());
   });
   AT_CUDA_CHECK(cudaGetLastError());
   return output;
@@ -402,7 +402,7 @@ at::Tensor ROIAlign_backward_cuda(
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(grad.type(), "ROIAlign_backward", [&] {
     RoIAlignBackward<scalar_t><<<grid, block, 0, stream>>>(
         grad.numel(),
-        grad.data<scalar_t>(),
+        grad.data_ptr<scalar_t>(),
         spatial_scale,
         channels,
         height,
@@ -410,8 +410,8 @@ at::Tensor ROIAlign_backward_cuda(
         pooled_height,
         pooled_width,
         sampling_ratio,
-        grad_input.data<scalar_t>(),
-        rois.contiguous().data<scalar_t>(),
+        grad_input.data_ptr<scalar_t>(),
+        rois.contiguous().data_ptr<scalar_t>(),
         n_stride,
         c_stride,
         h_stride,

torchvision/csrc/cuda/ROIPool_cuda.cu

@@ -160,16 +160,16 @@ std::tuple<at::Tensor, at::Tensor> ROIPool_forward_cuda(
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(input.type(), "ROIPool_forward", [&] {
     RoIPoolForward<scalar_t><<<grid, block, 0, stream>>>(
         output_size,
-        input.contiguous().data<scalar_t>(),
+        input.contiguous().data_ptr<scalar_t>(),
         spatial_scale,
         channels,
         height,
         width,
         pooled_height,
         pooled_width,
-        rois.contiguous().data<scalar_t>(),
-        output.data<scalar_t>(),
-        argmax.data<int>());
+        rois.contiguous().data_ptr<scalar_t>(),
+        output.data_ptr<scalar_t>(),
+        argmax.data_ptr<int>());
   });
   AT_CUDA_CHECK(cudaGetLastError());
   return std::make_tuple(output, argmax);
@@ -227,8 +227,8 @@ at::Tensor ROIPool_backward_cuda(
   AT_DISPATCH_FLOATING_TYPES_AND_HALF(grad.type(), "ROIPool_backward", [&] {
     RoIPoolBackward<scalar_t><<<grid, block, 0, stream>>>(
         grad.numel(),
-        grad.data<scalar_t>(),
-        argmax.contiguous().data<int>(),
+        grad.data_ptr<scalar_t>(),
+        argmax.contiguous().data_ptr<int>(),
         num_rois,
         spatial_scale,
         channels,
@@ -236,8 +236,8 @@ at::Tensor ROIPool_backward_cuda(
         width,
         pooled_height,
         pooled_width,
-        grad_input.data<scalar_t>(),
-        rois.contiguous().data<scalar_t>(),
+        grad_input.data_ptr<scalar_t>(),
+        rois.contiguous().data_ptr<scalar_t>(),
         n_stride,
         c_stride,
         h_stride,

torchvision/csrc/cuda/nms_cuda.cu

@@ -96,19 +96,19 @@ at::Tensor nms_cuda(const at::Tensor& dets,
         nms_kernel<scalar_t><<<blocks, threads, 0, stream>>>(
             dets_num,
             iou_threshold,
-            dets_sorted.data<scalar_t>(),
-            (unsigned long long*)mask.data<int64_t>());
+            dets_sorted.data_ptr<scalar_t>(),
+            (unsigned long long*)mask.data_ptr<int64_t>());
       });
 
   at::Tensor mask_cpu = mask.to(at::kCPU);
-  unsigned long long* mask_host = (unsigned long long*)mask_cpu.data<int64_t>();
+  unsigned long long* mask_host = (unsigned long long*)mask_cpu.data_ptr<int64_t>();
 
   std::vector<unsigned long long> remv(col_blocks);
   memset(&remv[0], 0, sizeof(unsigned long long) * col_blocks);
 
   at::Tensor keep =
       at::empty({dets_num}, dets.options().dtype(at::kLong).device(at::kCPU));
-  int64_t* keep_out = keep.data<int64_t>();
+  int64_t* keep_out = keep.data_ptr<int64_t>();
 
   int num_to_keep = 0;
   for (int i = 0; i < dets_num; i++) {