Skip to content

Commit 06d1be2

Browse files
gandalf2390facebook-github-bot
gandalf2390
authored and
facebook-github-bot
committed
[NOOP][clangformat][codemod] Enable CLANGFORMAT for caffe2/caffe2/* (pytorch#67624)
Summary: Pull Request resolved: pytorch#67624 Test Plan: Visual inspection. Sandcastle. Reviewed By: malfet Differential Revision: D31986628 fbshipit-source-id: c872bded7325997a2945dbf5d4d052628dcb3659
1 parent e86a5a3 commit 06d1be2

File tree

83 files changed

+1231
-947
lines changed

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

83 files changed

+1231
-947
lines changed

‎caffe2/cuda_rtc/common_rtc.h‎

Lines changed: 36 additions & 25 deletions
Original file line numberDiff line numberDiff line change
@@ -7,14 +7,14 @@
77
#include <cuda.h>
88
#include <nvrtc.h>
99

10-
#define NVRTC_CHECK(condition) \
11-
do { \
12-
nvrtcResult result = condition; \
13-
if (result != NVRTC_SUCCESS) { \
14-
LOG(FATAL) << "Error at: " << __FILE__ << ":" << __LINE__ << ": " \
15-
<< nvrtcGetErrorString(result); \
16-
} \
17-
} while(0)
10+
#define NVRTC_CHECK(condition) \
11+
do { \
12+
nvrtcResult result = condition; \
13+
if (result != NVRTC_SUCCESS) { \
14+
LOG(FATAL) << "Error at: " << __FILE__ << ":" << __LINE__ << ": " \
15+
<< nvrtcGetErrorString(result); \
16+
} \
17+
} while (0)
1818

1919
namespace caffe2 {
2020

@@ -39,15 +39,14 @@ class CudaRTCFunction {
3939
VLOG(1) << "function src:\n" << src;
4040
// Actually do the compiling.
4141
nvrtcProgram prog;
42-
NVRTC_CHECK(nvrtcCreateProgram(
43-
&prog, src.c_str(), nullptr, 0, nullptr, nullptr));
42+
NVRTC_CHECK(
43+
nvrtcCreateProgram(&prog, src.c_str(), nullptr, 0, nullptr, nullptr));
4444
// Compile the program.
4545
// TODO(Yangqing): how to find the current gpu architecture instead of hard
4646
// coding it?
47-
const char *nvrtc_opts[] = {"--gpu-architecture=compute_35",
48-
"--use_fast_math"};
49-
nvrtcResult compile_result = nvrtcCompileProgram(
50-
prog, 2, nvrtc_opts);
47+
const char* nvrtc_opts[] = {
48+
"--gpu-architecture=compute_35", "--use_fast_math"};
49+
nvrtcResult compile_result = nvrtcCompileProgram(prog, 2, nvrtc_opts);
5150
if (compile_result != NVRTC_SUCCESS) {
5251
size_t log_size;
5352
NVRTC_CHECK(nvrtcGetProgramLogSize(prog, &log_size));
@@ -74,21 +73,33 @@ class CudaRTCFunction {
7473
}
7574

7675
template <typename... Args>
77-
void Launch(unsigned int gx, unsigned int gy, unsigned int gz,
78-
unsigned int bx, unsigned int by, unsigned int bz,
79-
unsigned int shared_mem, cudaStream_t stream,
80-
Args... args) {
76+
void Launch(
77+
unsigned int gx,
78+
unsigned int gy,
79+
unsigned int gz,
80+
unsigned int bx,
81+
unsigned int by,
82+
unsigned int bz,
83+
unsigned int shared_mem,
84+
cudaStream_t stream,
85+
Args... args) {
8186
CAFFE_ENFORCE(
8287
module_loaded_, "Cannot call Launch before a module is loaded.");
83-
void * args_voidp[] = {&args...};
88+
void* args_voidp[] = {&args...};
8489
CUDA_DRIVERAPI_ENFORCE(cuLaunchKernel(
8590
kernel_, gx, gy, gz, bx, by, bz, shared_mem, stream, args_voidp, 0));
8691
}
8792

88-
void LaunchEx(unsigned int gx, unsigned int gy, unsigned int gz,
89-
unsigned int bx, unsigned int by, unsigned int bz,
90-
unsigned int shared_mem, cudaStream_t stream,
91-
void** extra) {
93+
void LaunchEx(
94+
unsigned int gx,
95+
unsigned int gy,
96+
unsigned int gz,
97+
unsigned int bx,
98+
unsigned int by,
99+
unsigned int bz,
100+
unsigned int shared_mem,
101+
cudaStream_t stream,
102+
void** extra) {
92103
CAFFE_ENFORCE(
93104
module_loaded_, "Cannot call Launch before a module is loaded.");
94105
CUDA_DRIVERAPI_ENFORCE(cuLaunchKernel(
@@ -115,6 +126,6 @@ inline std::string GetUniqueName() {
115126
return ss.str();
116127
}
117128

118-
} // namepsace caffe2
129+
} // namespace caffe2
119130

120-
#endif // CAFFE2_CUDA_RTC_COMMON_RTC_H_
131+
#endif // CAFFE2_CUDA_RTC_COMMON_RTC_H_

‎caffe2/cuda_rtc/elemenntwise_rtc_gpu.cc‎

Lines changed: 20 additions & 21 deletions
Original file line numberDiff line numberDiff line change
@@ -5,8 +5,7 @@
55

66
namespace caffe2 {
77
namespace {
8-
class ElementwiseRTCFunction
9-
: public CudaRTCFunction<ElementwiseRTCFunction> {
8+
class ElementwiseRTCFunction : public CudaRTCFunction<ElementwiseRTCFunction> {
109
public:
1110
ElementwiseRTCFunction() : CudaRTCFunction(), name_(GetUniqueName()) {}
1211

@@ -22,22 +21,21 @@ class ElementwiseRTCFunction
2221
string name_;
2322
};
2423

25-
template<>
24+
template <>
2625
string ElementwiseRTCFunction::GetSource(
27-
int input_size, int output_size,
26+
int input_size,
27+
int output_size,
2828
const string command_string) {
2929
std::stringstream ss;
30-
ss << "extern \"C\" __global__ void " << name_ <<
31-
"(const size_t nthreads, \n";
30+
ss << "extern \"C\" __global__ void " << name_
31+
<< "(const size_t nthreads, \n";
3232
// Insert the parameter list.
3333
int remain_params = input_size + output_size;
3434
for (int i = 0; i < input_size; ++i) {
35-
ss << "const float* in" << i
36-
<< ((remain_params--) ? ", \n" : "");
35+
ss << "const float* in" << i << ((remain_params--) ? ", \n" : "");
3736
}
3837
for (int i = 0; i < output_size; ++i) {
39-
ss << "float* out" << i
40-
<< ((remain_params--) ? ", \n" : "");
38+
ss << "float* out" << i << ((remain_params--) ? ", \n" : "");
4139
}
4240
ss << ") {\n"
4341
"for (int index = blockIdx.x * blockDim.x + threadIdx.x;\n"
@@ -46,7 +44,7 @@ string ElementwiseRTCFunction::GetSource(
4644
<< "}\n}";
4745
return ss.str();
4846
}
49-
} // namespace
47+
} // namespace
5048

5149
/**
5250
* A GPU operator that can generate limited elementwise operations.
@@ -75,17 +73,17 @@ class ElementwiseRTCOp final : public Operator<CUDAContext> {
7573
public:
7674
ElementwiseRTCOp(const OperatorDef& operator_def, Workspace* ws)
7775
: Operator<CUDAContext>(operator_def, ws) {
78-
const string src = OperatorBase::GetSingleArgument<string>(
79-
"rtc_src", "");
76+
const string src = OperatorBase::GetSingleArgument<string>("rtc_src", "");
8077
CAFFE_ENFORCE(src.size(), "Op should have a non-zero source code size.");
8178
func_.Compile(InputSize(), OutputSize(), src);
8279
}
8380
~ElementwiseRTCOp() override {}
8481

8582
bool RunOnDevice() override {
86-
static_assert(sizeof(void*) == sizeof(size_t),
87-
"The argbuffer relies on the assumption that void* and "
88-
"size_t have the same size.");
83+
static_assert(
84+
sizeof(void*) == sizeof(size_t),
85+
"The argbuffer relies on the assumption that void* and "
86+
"size_t have the same size.");
8987
vector<size_t> argBuffer_vec(InputSize() + OutputSize() + 1);
9088
size_t* argBuffer = argBuffer_vec.data();
9189
CAFFE_ENFORCE(
@@ -102,10 +100,11 @@ class ElementwiseRTCOp final : public Operator<CUDAContext> {
102100
}
103101
size_t argBufferSize = sizeof(argBuffer);
104102
void* config[] = {
105-
CU_LAUNCH_PARAM_BUFFER_POINTER, argBuffer,
106-
CU_LAUNCH_PARAM_BUFFER_SIZE, &argBufferSize,
107-
CU_LAUNCH_PARAM_END
108-
};
103+
CU_LAUNCH_PARAM_BUFFER_POINTER,
104+
argBuffer,
105+
CU_LAUNCH_PARAM_BUFFER_SIZE,
106+
&argBufferSize,
107+
CU_LAUNCH_PARAM_END};
109108
func_.LaunchEx(
110109
CAFFE_GET_BLOCKS(Input(0).numel()),
111110
1,
@@ -127,4 +126,4 @@ namespace {
127126
REGISTER_CUDA_OPERATOR_WITH_ENGINE(ElementwiseRTC, NVRTC, ElementwiseRTCOp);
128127
}
129128

130-
} // namespace caffe2
129+
} // namespace caffe2

‎caffe2/cuda_rtc/pool_op_rtc_gpu.cc‎

Lines changed: 44 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -2,14 +2,14 @@
22

33
#include "caffe2/core/common_gpu.h"
44
#include "caffe2/core/context_gpu.h"
5-
#include "caffe2/operators/pool_op.h"
65
#include "caffe2/cuda_rtc/common_rtc.h"
6+
#include "caffe2/operators/pool_op.h"
77

88
namespace caffe2 {
99
namespace {
1010
class AveragePool {};
1111
class MaxPool {};
12-
} // namespace
12+
} // namespace
1313

1414
namespace {
1515

@@ -98,7 +98,6 @@ __global__ void %s(
9898
}
9999
)";
100100

101-
102101
class MaxPoolRTCFunction : public CudaRTCFunction<MaxPoolRTCFunction> {
103102
public:
104103
MaxPoolRTCFunction() : CudaRTCFunction(), name_(GetUniqueName()) {}
@@ -132,7 +131,6 @@ class MaxPoolGradientRTCFunction
132131
string name_;
133132
};
134133

135-
136134
template <>
137135
string MaxPoolRTCFunction::GetSource(
138136
const int output_size,
@@ -149,9 +147,22 @@ string MaxPoolRTCFunction::GetSource(
149147
const int pad_l) {
150148
char buffer[65536];
151149
int nbytes = snprintf(
152-
buffer, 65536, kMaxPoolForwardNCHWSource, name_.c_str(), output_size,
153-
channels, height, width, pooled_height, pooled_width, kernel_h, kernel_w,
154-
stride_h, stride_w, pad_t, pad_l);
150+
buffer,
151+
65536,
152+
kMaxPoolForwardNCHWSource,
153+
name_.c_str(),
154+
output_size,
155+
channels,
156+
height,
157+
width,
158+
pooled_height,
159+
pooled_width,
160+
kernel_h,
161+
kernel_w,
162+
stride_h,
163+
stride_w,
164+
pad_t,
165+
pad_l);
155166
DCHECK_GE(nbytes, 0);
156167
DCHECK_LT(nbytes, 65536);
157168
return string(buffer);
@@ -174,16 +185,29 @@ string MaxPoolGradientRTCFunction::GetSource(
174185
const int pad_l) {
175186
char buffer[65536];
176187
int nbytes = snprintf(
177-
buffer, 65536, kMaxPoolBackwardNCHWSource, name_.c_str(), output_size,
178-
num, channels, height, width, pooled_height, pooled_width, kernel_h,
179-
kernel_w, stride_h, stride_w, pad_t, pad_l);
188+
buffer,
189+
65536,
190+
kMaxPoolBackwardNCHWSource,
191+
name_.c_str(),
192+
output_size,
193+
num,
194+
channels,
195+
height,
196+
width,
197+
pooled_height,
198+
pooled_width,
199+
kernel_h,
200+
kernel_w,
201+
stride_h,
202+
stride_w,
203+
pad_t,
204+
pad_l);
180205
DCHECK_GE(nbytes, 0);
181206
DCHECK_LT(nbytes, 65536);
182207
return string(buffer);
183208
}
184209

185-
} // namespace
186-
210+
} // namespace
187211

188212
class MaxPoolRTCOp final : public ConvPoolOpBase<CUDAContext> {
189213
public:
@@ -196,7 +220,8 @@ class MaxPoolRTCOp final : public ConvPoolOpBase<CUDAContext> {
196220

197221
bool RunOnDeviceWithOrderNCHW() override {
198222
auto& X = Input(0);
199-
auto output_sizes = ConvPoolOpBase<CUDAContext>::GetOutputSize(X, X.dim32(1));
223+
auto output_sizes =
224+
ConvPoolOpBase<CUDAContext>::GetOutputSize(X, X.dim32(1));
200225
auto* Y = Output(0, output_sizes, at::dtype<float>());
201226

202227
if (input_dims_ != X.sizes()) {
@@ -307,7 +332,9 @@ class MaxPoolGradientRTCOp final : public ConvPoolOpBase<CUDAContext> {
307332

308333
namespace {
309334
REGISTER_CUDA_OPERATOR_WITH_ENGINE(MaxPool, NVRTC, MaxPoolRTCOp);
310-
REGISTER_CUDA_OPERATOR_WITH_ENGINE(MaxPoolGradient, NVRTC,
311-
MaxPoolGradientRTCOp);
312-
} // namespace
313-
} // namespace caffe2
335+
REGISTER_CUDA_OPERATOR_WITH_ENGINE(
336+
MaxPoolGradient,
337+
NVRTC,
338+
MaxPoolGradientRTCOp);
339+
} // namespace
340+
} // namespace caffe2

‎caffe2/db/create_db_op.cc‎

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -6,4 +6,4 @@ REGISTER_CPU_OPERATOR(CreateDB, CreateDBOp<CPUContext>);
66
OPERATOR_SCHEMA(CreateDB).NumInputs(0).NumOutputs(1);
77

88
NO_GRADIENT(CreateDB);
9-
} // namespace caffe2
9+
} // namespace caffe2

‎caffe2/db/leveldb.cc‎

Lines changed: 32 additions & 14 deletions
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,6 @@
11
#include "caffe2/core/db.h"
2-
#include "caffe2/core/logging.h"
32
#include "caffe2/core/flags.h"
3+
#include "caffe2/core/logging.h"
44
#include "leveldb/db.h"
55
#include "leveldb/write_batch.h"
66

@@ -19,13 +19,27 @@ class LevelDBCursor : public Cursor {
1919
SeekToFirst();
2020
}
2121
~LevelDBCursor() override {}
22-
void Seek(const string& key) override { iter_->Seek(key); }
23-
bool SupportsSeek() override { return true; }
24-
void SeekToFirst() override { iter_->SeekToFirst(); }
25-
void Next() override { iter_->Next(); }
26-
string key() override { return iter_->key().ToString(); }
27-
string value() override { return iter_->value().ToString(); }
28-
bool Valid() override { return iter_->Valid(); }
22+
void Seek(const string& key) override {
23+
iter_->Seek(key);
24+
}
25+
bool SupportsSeek() override {
26+
return true;
27+
}
28+
void SeekToFirst() override {
29+
iter_->SeekToFirst();
30+
}
31+
void Next() override {
32+
iter_->Next();
33+
}
34+
string key() override {
35+
return iter_->key().ToString();
36+
}
37+
string value() override {
38+
return iter_->value().ToString();
39+
}
40+
bool Valid() override {
41+
return iter_->Valid();
42+
}
2943

3044
private:
3145
std::unique_ptr<leveldb::Iterator> iter_;
@@ -47,8 +61,7 @@ class LevelDBTransaction : public Transaction {
4761
leveldb::Status status = db_->Write(leveldb::WriteOptions(), batch_.get());
4862
batch_.reset(new leveldb::WriteBatch());
4963
CAFFE_ENFORCE(
50-
status.ok(),
51-
"Failed to write batch to leveldb. ", status.ToString());
64+
status.ok(), "Failed to write batch to leveldb. ", status.ToString());
5265
}
5366

5467
private:
@@ -71,12 +84,17 @@ class LevelDB : public DB {
7184
leveldb::Status status = leveldb::DB::Open(options, source, &db_temp);
7285
CAFFE_ENFORCE(
7386
status.ok(),
74-
"Failed to open leveldb ", source, ". ", status.ToString());
87+
"Failed to open leveldb ",
88+
source,
89+
". ",
90+
status.ToString());
7591
db_.reset(db_temp);
7692
VLOG(1) << "Opened leveldb " << source;
7793
}
7894

79-
void Close() override { db_.reset(); }
95+
void Close() override {
96+
db_.reset();
97+
}
8098
unique_ptr<Cursor> NewCursor() override {
8199
return make_unique<LevelDBCursor>(db_.get());
82100
}
@@ -92,5 +110,5 @@ REGISTER_CAFFE2_DB(LevelDB, LevelDB);
92110
// For lazy-minded, one can also call with lower-case name.
93111
REGISTER_CAFFE2_DB(leveldb, LevelDB);
94112

95-
} // namespace db
96-
} // namespace caffe2
113+
} // namespace db
114+
} // namespace caffe2

0 commit comments

Comments
 (0)