FM / native

/usr/local/lib64/python3.6/site-packages/torch/include/ATen/native/

/usr/local/lib64/python3.6/site-packages/torch/include/ATen/native

Name	Size	Mode	Actions
cpu/	-	0755	rm
cuda/	-	0755	rm
quantized/	-	0755	rm
Activation.h	3069	0644	edit dl rm
AdaptivePooling.h	1165	0644	edit dl rm
BatchLinearAlgebra.h	8246	0644	edit dl rm
batch_norm.h	1285	0644	edit dl rm
BinaryOps.h	4916	0644	edit dl rm
BucketizationUtils.h	4248	0644	edit dl rm
ComplexHelper.h	3797	0644	edit dl rm
CompositeRandomAccessor.h	888	0644	edit dl rm
CompositeRandomAccessorCommon.h	6713	0644	edit dl rm
ConvUtils.h	5350	0644	edit dl rm
Copy.h	356	0644	edit dl rm
CPUBlas.h	4199	0644	edit dl rm
CPUFallback.h	2404	0644	edit dl rm
Cross.h	262	0644	edit dl rm
DilatedConvolutionUtils.h	6416	0644	edit dl rm
DispatchStub.h	7672	0644	edit dl rm
Distance.h	732	0644	edit dl rm
Distributions.h	21654	0644	edit dl rm
DistributionTemplates.h	18623	0644	edit dl rm
EmbeddingBag.h	1320	0644	edit dl rm
Fill.h	384	0644	edit dl rm
ForeachUtils.h	5962	0644	edit dl rm
FunctionOfAMatrixUtils.h	436	0644	edit dl rm
GridSampler.h	10525	0644	edit dl rm
group_norm.h	896	0644	edit dl rm
Histogram.h	492	0644	edit dl rm
im2col.h	2838	0644	edit dl rm
im2col_shape_check.h	6181	0644	edit dl rm
IndexingUtils.h	5373	0644	edit dl rm
layer_norm.h	2892	0644	edit dl rm
Lerp.h	553	0644	edit dl rm
LinearAlgebra.h	603	0644	edit dl rm
LinearAlgebraUtils.h	25236	0644	edit dl rm
LossMulti.h	2197	0644	edit dl rm
Math.h	91356	0644	edit dl rm
MathBitFallThroughLists.h	4086	0644	edit dl rm
MathBitsFallback.h	7326	0644	edit dl rm
MaxPooling.h	1234	0644	edit dl rm
Normalization.h	302	0644	edit dl rm
PointwiseOps.h	749	0644	edit dl rm
Pool.h	10922	0644	edit dl rm
Pow.h	1694	0644	edit dl rm
ReduceAllOps.h	378	0644	edit dl rm
ReduceOps.h	1745	0644	edit dl rm
ReduceOpsUtils.h	12245	0644	edit dl rm
Repeat.h	1286	0644	edit dl rm
Resize.h	6501	0644	edit dl rm
ResizeCommon.h	1321	0644	edit dl rm
RNN.h	2467	0644	edit dl rm
ScatterGatherChecks.h	3641	0644	edit dl rm
SegmentReduce.h	685	0644	edit dl rm
SharedReduceOps.h	15785	0644	edit dl rm
SobolEngineOpsUtils.h	1723	0644	edit dl rm
Sorting.h	536	0644	edit dl rm
SortingUtils.h	5722	0644	edit dl rm
SpectralOpsUtils.h	3146	0644	edit dl rm
StridedRandomAccessor.h	6847	0644	edit dl rm
TensorAdvancedIndexing.h	3072	0644	edit dl rm
TensorCompare.h	1333	0644	edit dl rm
TensorDimApply.h	1832	0644	edit dl rm
TensorFactories.h	3382	0644	edit dl rm
TensorIterator.h	46	0644	edit dl rm
TensorIteratorDynamicCasting.h	2025	0644	edit dl rm
TensorShape.h	1049	0644	edit dl rm
TensorTransformations.h	938	0644	edit dl rm
TriangularOpsUtils.h	2000	0644	edit dl rm
TypeProperties.h	496	0644	edit dl rm
UnaryOps.h	4464	0644	edit dl rm
Unfold2d.h	551	0644	edit dl rm
Unfold3d.h	852	0644	edit dl rm
UnfoldBackward.h	5398	0644	edit dl rm
UpSample.h	13599	0644	edit dl rm
vol2col.h	3642	0644	edit dl rm

Edit: /usr/local/lib64/python3.6/site-packages/torch/include/ATen/native/ConvUtils.h (5350B)

#pragma once
#include <ATen/detail/CUDAHooksInterface.h>
#include <c10/util/env.h>

namespace at { namespace native {

// ---------------------------------------------------------------------
//
// Math
//
// ---------------------------------------------------------------------

constexpr int input_batch_size_dim = 0;  // also grad_input
constexpr int input_channels_dim = 1;
constexpr int output_batch_size_dim = 0;  // also grad_output
constexpr int output_channels_dim = 1;
constexpr int weight_output_channels_dim = 0;
constexpr int weight_input_channels_dim = 1;

// Often written as 2 + max_dim (extra dims for batch size and channels)
constexpr int max_dim = 3;

// NB: conv_output_size and conv_input_size are not bijections,
// as conv_output_size loses information; this is why conv_input_size
// takes an extra output_padding argument to resolve the ambiguity.

static inline std::vector<int64_t> conv_output_size(
    IntArrayRef input_size, IntArrayRef weight_size,
    IntArrayRef padding, IntArrayRef stride, IntArrayRef dilation = IntArrayRef()
) {
  // ASSERT(input_size.size() > 2)
  // ASSERT(input_size.size() == weight_size.size())
  bool has_dilation = dilation.size() > 0;
  auto dim = input_size.size();
  std::vector<int64_t> output_size(dim);
  output_size[0] = input_size[input_batch_size_dim];
  output_size[1] = weight_size[weight_output_channels_dim];
  for (size_t d = 2; d < dim; ++d) {
    auto dilation_ = has_dilation ? dilation[d - 2] : 1;
    auto kernel = dilation_ * (weight_size[d] - 1) + 1;
    output_size[d] = (input_size[d] + (2 * padding[d - 2]) - kernel) / stride[d - 2] + 1;
  }
  return output_size;
}

static inline std::vector<int64_t> conv_input_size(
    IntArrayRef output_size, IntArrayRef weight_size,
    IntArrayRef padding, IntArrayRef output_padding, IntArrayRef stride, IntArrayRef dilation, int64_t groups
) {
  // ASSERT(output_size.size() > 2)
  // ASSERT(output_size.size() == weight_size.size())
  auto dim = output_size.size();
  std::vector<int64_t> input_size(dim);
  input_size[0] = output_size[output_batch_size_dim];
  input_size[1] = weight_size[weight_input_channels_dim] * groups;
  for (size_t d = 2; d < dim; ++d) {
    int kernel = dilation[d - 2] * (weight_size[d] - 1) + 1;
    input_size[d] = (output_size[d] - 1) * stride[d - 2] - (2 * padding[d - 2]) +
                     kernel + output_padding[d - 2];
  }
  return input_size;
}

static inline std::vector<int64_t> conv_weight_size(
    IntArrayRef input_size, IntArrayRef output_size,
    IntArrayRef padding, IntArrayRef output_padding, IntArrayRef stride, IntArrayRef dilation, int64_t groups
) {
  auto dim = input_size.size();
  std::vector<int64_t> weight_size(dim);
  weight_size[0] = output_size[1];
  weight_size[1] = input_size[1] / groups;
  for (size_t d = 2; d < dim; ++d) {
    int kernel = input_size[d] - (output_size[d] - 1) * stride[d - 2]
               + 2 * padding[d - 2] - output_padding[d - 2];
    weight_size[d] = (kernel - 1) / dilation[d - 2] + 1;
  }
  return weight_size;
}

static inline Tensor reshape_bias(int64_t dim, const Tensor& bias) {
  std::vector<int64_t> shape(dim, 1);
  shape[1] = -1;
  return bias.reshape(shape);
}

static inline bool cudnn_conv_use_channels_last(const at::Tensor& input, const at::Tensor& weight) {
  // disable NHWC for float64 input.
  if (!at::detail::getCUDAHooks().compiledWithCuDNN() ||
      input.scalar_type() == at::kDouble ||
      weight.scalar_type() == at::kDouble) {
    return false;
  }
  long cudnn_version = at::detail::getCUDAHooks().versionCuDNN();
  auto input_memory_format = input.suggest_memory_format();
  auto weight_memory_format = weight.suggest_memory_format();

bool can_use_cudnn_channels_last_2d = (cudnn_version >= 7603) && (
    (input_memory_format  == at::MemoryFormat::ChannelsLast) ||
    (weight_memory_format == at::MemoryFormat::ChannelsLast)
  );

bool can_use_cudnn_channels_last_3d = (cudnn_version >= 8005) && (
    (input_memory_format  == at::MemoryFormat::ChannelsLast3d) ||
    (weight_memory_format == at::MemoryFormat::ChannelsLast3d)
  );

return can_use_cudnn_channels_last_2d || can_use_cudnn_channels_last_3d;
}

static inline bool miopen_conv_use_channels_last(const at::Tensor& input, const at::Tensor& weight) {

// disable NHWC for float64 input.
  if (!at::detail::getCUDAHooks().compiledWithMIOpen() ||
      input.scalar_type() == at::kDouble ||
      weight.scalar_type() == at::kDouble) {
    return false;
  }

bool can_use_miopen_channels_last_2d = false;
#if defined(USE_ROCM) && (ROCM_VERSION >= 40300)
  // TODO: Remove PYTORCH_MIOPEN_SUGGEST_NHWC once ROCm officially supports NHWC in MIOpen
  // See #64427
  static c10::optional<bool> PYTORCH_MIOPEN_SUGGEST_NHWC = c10::utils::check_env("PYTORCH_MIOPEN_SUGGEST_NHWC");

auto input_memory_format = input.suggest_memory_format();
  auto weight_memory_format = weight.suggest_memory_format();

can_use_miopen_channels_last_2d = PYTORCH_MIOPEN_SUGGEST_NHWC &&  *PYTORCH_MIOPEN_SUGGEST_NHWC && (
            ( (input_memory_format  == at::MemoryFormat::ChannelsLast) ||
            (weight_memory_format == at::MemoryFormat::ChannelsLast) )
        );
#endif

bool can_use_miopen_channels_last_3d = false;

return can_use_miopen_channels_last_2d || can_use_miopen_channels_last_3d;
}

}} // namespace at::native

cmd: