Merge branch 'fix-pt1.5-bug' into 'master'

Fix compilation failure in PyTorch1.5

See merge request open-mmlab/mmdet.3d!81

mmdet3d/ops/ball_query/src/ball_query.cpp

@@ -33,7 +33,7 @@ int ball_query_wrapper(int b, int n, int m, float radius, int nsample,
   const float *xyz = xyz_tensor.data_ptr<float>();
   int *idx = idx_tensor.data_ptr<int>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   ball_query_kernel_launcher(b, n, m, radius, nsample, new_xyz, xyz, idx,
                              stream);
   return 1;

mmdet3d/ops/furthest_point_sample/src/furthest_point_sample.cpp

@@ -24,7 +24,7 @@ int furthest_point_sampling_wrapper(int b, int n, int m,
   float *temp = temp_tensor.data_ptr<float>();
   int *idx = idx_tensor.data_ptr<int>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   furthest_point_sampling_kernel_launcher(b, n, m, points, temp, idx, stream);
   return 1;
 }

mmdet3d/ops/gather_points/src/gather_points.cpp

@@ -32,7 +32,7 @@ int gather_points_wrapper(int b, int c, int n, int npoints,
   const int *idx = idx_tensor.data_ptr<int>();
   float *out = out_tensor.data_ptr<float>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   gather_points_kernel_launcher(b, c, n, npoints, points, idx, out, stream);
   return 1;
 }
@@ -45,7 +45,7 @@ int gather_points_grad_wrapper(int b, int c, int n, int npoints,
   const int *idx = idx_tensor.data_ptr<int>();
   float *grad_points = grad_points_tensor.data_ptr<float>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   gather_points_grad_kernel_launcher(b, c, n, npoints, grad_out, idx,
                                      grad_points, stream);
   return 1;

mmdet3d/ops/group_points/src/group_points.cpp

@@ -32,7 +32,7 @@ int group_points_grad_wrapper(int b, int c, int n, int npoints, int nsample,
   const int *idx = idx_tensor.data_ptr<int>();
   const float *grad_out = grad_out_tensor.data_ptr<float>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
 
   group_points_grad_kernel_launcher(b, c, n, npoints, nsample, grad_out, idx,
                                     grad_points, stream);
@@ -46,7 +46,7 @@ int group_points_wrapper(int b, int c, int n, int npoints, int nsample,
   const int *idx = idx_tensor.data_ptr<int>();
   float *out = out_tensor.data_ptr<float>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
 
   group_points_kernel_launcher(b, c, n, npoints, nsample, points, idx, out,
                                stream);

mmdet3d/ops/interpolate/src/interpolate.cpp

@@ -48,7 +48,7 @@ void three_nn_wrapper(int b, int n, int m, at::Tensor unknown_tensor,
   float *dist2 = dist2_tensor.data_ptr<float>();
   int *idx = idx_tensor.data_ptr<int>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   three_nn_kernel_launcher(b, n, m, unknown, known, dist2, idx, stream);
 }
 
@@ -61,7 +61,7 @@ void three_interpolate_wrapper(int b, int c, int m, int n,
   float *out = out_tensor.data_ptr<float>();
   const int *idx = idx_tensor.data_ptr<int>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   three_interpolate_kernel_launcher(b, c, m, n, points, idx, weight, out,
                                     stream);
 }
@@ -76,7 +76,7 @@ void three_interpolate_grad_wrapper(int b, int c, int n, int m,
   float *grad_points = grad_points_tensor.data_ptr<float>();
   const int *idx = idx_tensor.data_ptr<int>();
 
-  cudaStream_t stream = THCState_getCurrentStream(state);
+  cudaStream_t stream = at::cuda::getCurrentCUDAStream().stream();
   three_interpolate_grad_kernel_launcher(b, c, n, m, grad_out, idx, weight,
                                          grad_points, stream);
 }

setup.py

@@ -229,7 +229,11 @@ if __name__ == '__main__':
                 name='sparse_conv_ext',
                 module='mmdet3d.ops.spconv',
                 extra_include_path=[
-                    os.path.join(*'mmdet3d.ops.spconv'.split('.'), 'include/')
+                    # PyTorch 1.5 uses ninjia, which requires absolute path
+                    # of included files, relative path will cause failure.
+                    os.path.abspath(
+                        os.path.join(*'mmdet3d.ops.spconv'.split('.'),
+                                     'include/'))
                 ],
                 sources=[
                     'src/all.cc',