Skip to content

Commit d81a9b7

Browse files
ranapratap55ranapratap55
committed
update clang format
1 parent 46c833f commit d81a9b7

File tree

2 files changed

+58
-54
lines changed

2 files changed

+58
-54
lines changed

‎clang/lib/CodeGen/TargetBuiltins/AMDGPU.cpp‎

Lines changed: 43 additions & 41 deletions
Original file line numberDiff line numberDiff line change
@@ -217,8 +217,7 @@ static llvm::Value *LoadTextureDescPtorAsVec8I32(CodeGenFunction &CGF,
217217
if (RsrcPtr->getType()->isIntegerTy(32)) {
218218
unsigned AS = 8;
219219
llvm::PointerType *VecPtrTy = llvm::PointerType::get(VecTy, AS);
220-
llvm::Value *Ptr =
221-
B.CreateIntToPtr(RsrcPtr, VecPtrTy, "tex.rsrc.from.int");
220+
llvm::Value *Ptr = B.CreateIntToPtr(RsrcPtr, VecPtrTy, "tex.rsrc.from.int");
222221
return B.CreateAlignedLoad(VecTy, Ptr, llvm::Align(32), "tex.rsrc.val");
223222
}
224223

@@ -240,22 +239,25 @@ static llvm::Value *LoadTextureDescPtorAsVec8I32(CodeGenFunction &CGF,
240239
static unsigned GetTextureDescIndex(unsigned BuiltinID, const CallExpr *E) {
241240
unsigned N = E->getNumArgs();
242241
if (IsImageSampleBuiltIn(BuiltinID)) {
243-
if (N < 5) return (unsigned)-1;
242+
if (N < 5)
243+
return (unsigned)-1;
244244
return N - 5;
245245
}
246-
247-
if (N < 3) return (unsigned)-1;
246+
247+
if (N < 3)
248+
return (unsigned)-1;
248249
return N - 3;
249250
}
250251

251-
llvm::CallInst *EmitAMDGCNImageOverloadedReturnType(clang::CodeGen::CodeGenFunction &CGF,
252-
const clang::CallExpr *E,
253-
unsigned IntrinsicID,
254-
bool IsImageStore) {
252+
llvm::CallInst *
253+
EmitAMDGCNImageOverloadedReturnType(clang::CodeGen::CodeGenFunction &CGF,
254+
const clang::CallExpr *E,
255+
unsigned IntrinsicID,
256+
bool IsImageStore) {
255257
clang::SmallVector<llvm::Value *, 10> Args;
256258
unsigned RsrcIndex = GetTextureDescIndex(E->getBuiltinCallee(), E);
257259

258-
for (unsigned I = 0; I < E->getNumArgs(); ++I){
260+
for (unsigned I = 0; I < E->getNumArgs(); ++I) {
259261
llvm::Value *V = CGF.EmitScalarExpr(E->getArg(I));
260262
if (I == RsrcIndex)
261263
V = LoadTextureDescPtorAsVec8I32(CGF, V);
@@ -1028,133 +1030,133 @@ Value *CodeGenFunction::EmitAMDGPUBuiltinExpr(unsigned BuiltinID,
10281030
case AMDGPU::BI__builtin_amdgcn_image_load_1d_v4f32_i32:
10291031
case AMDGPU::BI__builtin_amdgcn_image_load_1d_v4f16_i32:
10301032
return EmitAMDGCNImageOverloadedReturnType(
1031-
*this, E, Intrinsic::amdgcn_image_load_1d, false);
1033+
*this, E, Intrinsic::amdgcn_image_load_1d, false);
10321034
case AMDGPU::BI__builtin_amdgcn_image_load_1darray_v4f32_i32:
10331035
case AMDGPU::BI__builtin_amdgcn_image_load_1darray_v4f16_i32:
10341036
return EmitAMDGCNImageOverloadedReturnType(
1035-
*this, E, Intrinsic::amdgcn_image_load_1darray, false);
1037+
*this, E, Intrinsic::amdgcn_image_load_1darray, false);
10361038
case AMDGPU::BI__builtin_amdgcn_image_load_2d_f32_i32:
10371039
case AMDGPU::BI__builtin_amdgcn_image_load_2d_v4f32_i32:
10381040
case AMDGPU::BI__builtin_amdgcn_image_load_2d_v4f16_i32:
10391041
return EmitAMDGCNImageOverloadedReturnType(
1040-
*this, E, Intrinsic::amdgcn_image_load_2d, false);
1042+
*this, E, Intrinsic::amdgcn_image_load_2d, false);
10411043
case AMDGPU::BI__builtin_amdgcn_image_load_2darray_f32_i32:
10421044
case AMDGPU::BI__builtin_amdgcn_image_load_2darray_v4f32_i32:
10431045
case AMDGPU::BI__builtin_amdgcn_image_load_2darray_v4f16_i32:
10441046
return EmitAMDGCNImageOverloadedReturnType(
1045-
*this, E, Intrinsic::amdgcn_image_load_2darray, false);
1047+
*this, E, Intrinsic::amdgcn_image_load_2darray, false);
10461048
case AMDGPU::BI__builtin_amdgcn_image_load_3d_v4f32_i32:
10471049
case AMDGPU::BI__builtin_amdgcn_image_load_3d_v4f16_i32:
10481050
return EmitAMDGCNImageOverloadedReturnType(
1049-
*this, E, Intrinsic::amdgcn_image_load_3d, false);
1051+
*this, E, Intrinsic::amdgcn_image_load_3d, false);
10501052
case AMDGPU::BI__builtin_amdgcn_image_load_cube_v4f32_i32:
10511053
case AMDGPU::BI__builtin_amdgcn_image_load_cube_v4f16_i32:
10521054
return EmitAMDGCNImageOverloadedReturnType(
1053-
*this, E, Intrinsic::amdgcn_image_load_cube, false);
1055+
*this, E, Intrinsic::amdgcn_image_load_cube, false);
10541056
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1d_v4f32_i32:
10551057
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1d_v4f16_i32:
10561058
return EmitAMDGCNImageOverloadedReturnType(
1057-
*this, E, Intrinsic::amdgcn_image_load_mip_1d, false);
1059+
*this, E, Intrinsic::amdgcn_image_load_mip_1d, false);
10581060
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1darray_v4f32_i32:
10591061
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1darray_v4f16_i32:
10601062
return EmitAMDGCNImageOverloadedReturnType(
1061-
*this, E, Intrinsic::amdgcn_image_load_mip_1darray, false);
1063+
*this, E, Intrinsic::amdgcn_image_load_mip_1darray, false);
10621064
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_f32_i32:
10631065
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_v4f32_i32:
10641066
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_v4f16_i32:
10651067
return EmitAMDGCNImageOverloadedReturnType(
1066-
*this, E, Intrinsic::amdgcn_image_load_mip_2d, false);
1068+
*this, E, Intrinsic::amdgcn_image_load_mip_2d, false);
10671069
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_f32_i32:
10681070
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_v4f32_i32:
10691071
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_v4f16_i32:
10701072
return EmitAMDGCNImageOverloadedReturnType(
1071-
*this, E, Intrinsic::amdgcn_image_load_mip_2darray, false);
1073+
*this, E, Intrinsic::amdgcn_image_load_mip_2darray, false);
10721074
case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f32_i32:
10731075
case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f16_i32:
10741076
return EmitAMDGCNImageOverloadedReturnType(
1075-
*this, E, Intrinsic::amdgcn_image_load_mip_3d, false);
1077+
*this, E, Intrinsic::amdgcn_image_load_mip_3d, false);
10761078
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f32_i32:
10771079
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f16_i32:
10781080
return EmitAMDGCNImageOverloadedReturnType(
1079-
*this, E, Intrinsic::amdgcn_image_load_mip_cube, false);
1081+
*this, E, Intrinsic::amdgcn_image_load_mip_cube, false);
10801082
case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f32_i32:
10811083
case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f16_i32:
10821084
return EmitAMDGCNImageOverloadedReturnType(
1083-
*this, E, Intrinsic::amdgcn_image_store_1d, true);
1085+
*this, E, Intrinsic::amdgcn_image_store_1d, true);
10841086
case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f32_i32:
10851087
case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f16_i32:
10861088
return EmitAMDGCNImageOverloadedReturnType(
1087-
*this, E, Intrinsic::amdgcn_image_store_1darray, true);
1089+
*this, E, Intrinsic::amdgcn_image_store_1darray, true);
10881090
case AMDGPU::BI__builtin_amdgcn_image_store_2d_f32_i32:
10891091
case AMDGPU::BI__builtin_amdgcn_image_store_2d_v4f32_i32:
10901092
case AMDGPU::BI__builtin_amdgcn_image_store_2d_v4f16_i32:
10911093
return EmitAMDGCNImageOverloadedReturnType(
1092-
*this, E, Intrinsic::amdgcn_image_store_2d, true);
1094+
*this, E, Intrinsic::amdgcn_image_store_2d, true);
10931095
case AMDGPU::BI__builtin_amdgcn_image_store_2darray_f32_i32:
10941096
case AMDGPU::BI__builtin_amdgcn_image_store_2darray_v4f32_i32:
10951097
case AMDGPU::BI__builtin_amdgcn_image_store_2darray_v4f16_i32:
10961098
return EmitAMDGCNImageOverloadedReturnType(
1097-
*this, E, Intrinsic::amdgcn_image_store_2darray, true);
1099+
*this, E, Intrinsic::amdgcn_image_store_2darray, true);
10981100
case AMDGPU::BI__builtin_amdgcn_image_store_3d_v4f32_i32:
10991101
case AMDGPU::BI__builtin_amdgcn_image_store_3d_v4f16_i32:
11001102
return EmitAMDGCNImageOverloadedReturnType(
1101-
*this, E, Intrinsic::amdgcn_image_store_3d, true);
1103+
*this, E, Intrinsic::amdgcn_image_store_3d, true);
11021104
case AMDGPU::BI__builtin_amdgcn_image_store_cube_v4f32_i32:
11031105
case AMDGPU::BI__builtin_amdgcn_image_store_cube_v4f16_i32:
11041106
return EmitAMDGCNImageOverloadedReturnType(
1105-
*this, E, Intrinsic::amdgcn_image_store_cube, true);
1107+
*this, E, Intrinsic::amdgcn_image_store_cube, true);
11061108
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1d_v4f32_i32:
11071109
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1d_v4f16_i32:
11081110
return EmitAMDGCNImageOverloadedReturnType(
1109-
*this, E, Intrinsic::amdgcn_image_store_mip_1d, true);
1111+
*this, E, Intrinsic::amdgcn_image_store_mip_1d, true);
11101112
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1darray_v4f32_i32:
11111113
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1darray_v4f16_i32:
11121114
return EmitAMDGCNImageOverloadedReturnType(
1113-
*this, E, Intrinsic::amdgcn_image_store_mip_1darray, true);
1115+
*this, E, Intrinsic::amdgcn_image_store_mip_1darray, true);
11141116
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_f32_i32:
11151117
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_v4f32_i32:
11161118
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_v4f16_i32:
11171119
return EmitAMDGCNImageOverloadedReturnType(
1118-
*this, E, Intrinsic::amdgcn_image_store_mip_2d, true);
1120+
*this, E, Intrinsic::amdgcn_image_store_mip_2d, true);
11191121
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_f32_i32:
11201122
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_v4f32_i32:
11211123
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_v4f16_i32:
11221124
return EmitAMDGCNImageOverloadedReturnType(
1123-
*this, E, Intrinsic::amdgcn_image_store_mip_2darray, true);
1125+
*this, E, Intrinsic::amdgcn_image_store_mip_2darray, true);
11241126
case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f32_i32:
11251127
case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f16_i32:
11261128
return EmitAMDGCNImageOverloadedReturnType(
1127-
*this, E, Intrinsic::amdgcn_image_store_mip_3d, true);
1129+
*this, E, Intrinsic::amdgcn_image_store_mip_3d, true);
11281130
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f32_i32:
1129-
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f16_i32:
1131+
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f16_i32:
11301132
return EmitAMDGCNImageOverloadedReturnType(
1131-
*this, E, Intrinsic::amdgcn_image_store_mip_cube, true);
1133+
*this, E, Intrinsic::amdgcn_image_store_mip_cube, true);
11321134
case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f32_f32:
11331135
case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f16_f32:
11341136
return EmitAMDGCNImageOverloadedReturnType(
1135-
*this, E, Intrinsic::amdgcn_image_sample_1d, false);
1137+
*this, E, Intrinsic::amdgcn_image_sample_1d, false);
11361138
case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f32_f32:
11371139
case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f16_f32:
11381140
return EmitAMDGCNImageOverloadedReturnType(
1139-
*this, E, Intrinsic::amdgcn_image_sample_1darray, false);
1141+
*this, E, Intrinsic::amdgcn_image_sample_1darray, false);
11401142
case AMDGPU::BI__builtin_amdgcn_image_sample_2d_f32_f32:
11411143
case AMDGPU::BI__builtin_amdgcn_image_sample_2d_v4f32_f32:
11421144
case AMDGPU::BI__builtin_amdgcn_image_sample_2d_v4f16_f32:
11431145
return EmitAMDGCNImageOverloadedReturnType(
1144-
*this, E, Intrinsic::amdgcn_image_sample_2d, false);
1146+
*this, E, Intrinsic::amdgcn_image_sample_2d, false);
11451147
case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_f32_f32:
11461148
case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_v4f32_f32:
11471149
case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_v4f16_f32:
11481150
return EmitAMDGCNImageOverloadedReturnType(
1149-
*this, E, Intrinsic::amdgcn_image_sample_2darray, false);
1151+
*this, E, Intrinsic::amdgcn_image_sample_2darray, false);
11501152
case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f32_f32:
11511153
case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f16_f32:
11521154
return EmitAMDGCNImageOverloadedReturnType(
1153-
*this, E, Intrinsic::amdgcn_image_sample_3d, false);
1155+
*this, E, Intrinsic::amdgcn_image_sample_3d, false);
11541156
case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f32_f32:
11551157
case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f16_f32:
11561158
return EmitAMDGCNImageOverloadedReturnType(
1157-
*this, E, Intrinsic::amdgcn_image_sample_cube, false);
1159+
*this, E, Intrinsic::amdgcn_image_sample_cube, false);
11581160
case AMDGPU::BI__builtin_amdgcn_mfma_scale_f32_16x16x128_f8f6f4:
11591161
case AMDGPU::BI__builtin_amdgcn_mfma_scale_f32_32x32x64_f8f6f4: {
11601162
llvm::FixedVectorType *VT = FixedVectorType::get(Builder.getInt32Ty(), 8);

‎clang/lib/Sema/SemaAMDGPU.cpp‎

Lines changed: 15 additions & 13 deletions
Original file line numberDiff line numberDiff line change
@@ -139,7 +139,7 @@ bool SemaAMDGPU::CheckAMDGCNBuiltinFunctionCall(unsigned BuiltinID,
139139
case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f32_i32:
140140
case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f16_i32:
141141
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f32_i32:
142-
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f16_i32:
142+
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f16_i32:
143143
case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f32_f32:
144144
case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f32_f32:
145145
case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f16_f32:
@@ -154,20 +154,21 @@ bool SemaAMDGPU::CheckAMDGCNBuiltinFunctionCall(unsigned BuiltinID,
154154
case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f16_f32:
155155
case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f32_f32:
156156
case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f16_f32: {
157-
StringRef FeatureList(getASTContext().BuiltinInfo.getRequiredFeatures(BuiltinID));
158-
if (!Builtin::evaluateRequiredTargetFeatures(
159-
FeatureList, CallerFeatureMap)){
157+
StringRef FeatureList(
158+
getASTContext().BuiltinInfo.getRequiredFeatures(BuiltinID));
159+
if (!Builtin::evaluateRequiredTargetFeatures(FeatureList,
160+
CallerFeatureMap)) {
160161
Diag(TheCall->getBeginLoc(), diag::err_builtin_needs_feature)
161-
<< FD->getDeclName() << FeatureList;
162+
<< FD->getDeclName() << FeatureList;
162163
return false;
163164
}
164165

165166
unsigned ArgCount = TheCall->getNumArgs() - 1;
166167
llvm::APSInt Result;
167168

168169
return ((SemaRef.BuiltinConstantArg(TheCall, 0, Result)) ||
169-
(SemaRef.BuiltinConstantArg(TheCall, ArgCount, Result)) ||
170-
(SemaRef.BuiltinConstantArg(TheCall, (ArgCount - 1), Result)));
170+
(SemaRef.BuiltinConstantArg(TheCall, ArgCount, Result)) ||
171+
(SemaRef.BuiltinConstantArg(TheCall, (ArgCount - 1), Result)));
171172
}
172173
case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f32_i32:
173174
case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f32_i32:
@@ -197,20 +198,21 @@ bool SemaAMDGPU::CheckAMDGCNBuiltinFunctionCall(unsigned BuiltinID,
197198
case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f16_i32:
198199
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f32_i32:
199200
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f16_i32: {
200-
StringRef FeatureList(getASTContext().BuiltinInfo.getRequiredFeatures(BuiltinID));
201-
if (!Builtin::evaluateRequiredTargetFeatures(
202-
FeatureList, CallerFeatureMap)){
201+
StringRef FeatureList(
202+
getASTContext().BuiltinInfo.getRequiredFeatures(BuiltinID));
203+
if (!Builtin::evaluateRequiredTargetFeatures(FeatureList,
204+
CallerFeatureMap)) {
203205
Diag(TheCall->getBeginLoc(), diag::err_builtin_needs_feature)
204-
<< FD->getDeclName() << FeatureList;
206+
<< FD->getDeclName() << FeatureList;
205207
return false;
206208
}
207209

208210
unsigned ArgCount = TheCall->getNumArgs() - 1;
209211
llvm::APSInt Result;
210212

211213
return ((SemaRef.BuiltinConstantArg(TheCall, 1, Result)) ||
212-
(SemaRef.BuiltinConstantArg(TheCall, ArgCount, Result)) ||
213-
(SemaRef.BuiltinConstantArg(TheCall, (ArgCount - 1), Result)));
214+
(SemaRef.BuiltinConstantArg(TheCall, ArgCount, Result)) ||
215+
(SemaRef.BuiltinConstantArg(TheCall, (ArgCount - 1), Result)));
214216
}
215217
default:
216218
return false;

0 commit comments

Comments
 (0)