Skip to content

Commit 92d4a35

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

File tree

2 files changed

+58
-55
lines changed

2 files changed

+58
-55
lines changed

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

Lines changed: 43 additions & 42 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,24 @@ 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) {
255-
clang::SmallVector<llvm::Value *, 10> Args;
252+
llvm::CallInst *
253+
EmitAMDGCNImageOverloadedReturnType(clang::CodeGen::CodeGenFunction &CGF,
254+
const clang::CallExpr *E,
255+
unsigned IntrinsicID, bool IsImageStore) {
256+
clang::SmallVector<llvm::Value *, 10> Args;
256257
unsigned RsrcIndex = GetTextureDescIndex(E->getBuiltinCallee(), E);
257258

258-
for (unsigned I = 0; I < E->getNumArgs(); ++I){
259+
for (unsigned I = 0; I < E->getNumArgs(); ++I) {
259260
llvm::Value *V = CGF.EmitScalarExpr(E->getArg(I));
260261
if (I == RsrcIndex)
261262
V = LoadTextureDescPtorAsVec8I32(CGF, V);
@@ -1028,133 +1029,133 @@ Value *CodeGenFunction::EmitAMDGPUBuiltinExpr(unsigned BuiltinID,
10281029
case AMDGPU::BI__builtin_amdgcn_image_load_1d_v4f32_i32:
10291030
case AMDGPU::BI__builtin_amdgcn_image_load_1d_v4f16_i32:
10301031
return EmitAMDGCNImageOverloadedReturnType(
1031-
*this, E, Intrinsic::amdgcn_image_load_1d, false);
1032+
*this, E, Intrinsic::amdgcn_image_load_1d, false);
10321033
case AMDGPU::BI__builtin_amdgcn_image_load_1darray_v4f32_i32:
10331034
case AMDGPU::BI__builtin_amdgcn_image_load_1darray_v4f16_i32:
10341035
return EmitAMDGCNImageOverloadedReturnType(
1035-
*this, E, Intrinsic::amdgcn_image_load_1darray, false);
1036+
*this, E, Intrinsic::amdgcn_image_load_1darray, false);
10361037
case AMDGPU::BI__builtin_amdgcn_image_load_2d_f32_i32:
10371038
case AMDGPU::BI__builtin_amdgcn_image_load_2d_v4f32_i32:
10381039
case AMDGPU::BI__builtin_amdgcn_image_load_2d_v4f16_i32:
10391040
return EmitAMDGCNImageOverloadedReturnType(
1040-
*this, E, Intrinsic::amdgcn_image_load_2d, false);
1041+
*this, E, Intrinsic::amdgcn_image_load_2d, false);
10411042
case AMDGPU::BI__builtin_amdgcn_image_load_2darray_f32_i32:
10421043
case AMDGPU::BI__builtin_amdgcn_image_load_2darray_v4f32_i32:
10431044
case AMDGPU::BI__builtin_amdgcn_image_load_2darray_v4f16_i32:
10441045
return EmitAMDGCNImageOverloadedReturnType(
1045-
*this, E, Intrinsic::amdgcn_image_load_2darray, false);
1046+
*this, E, Intrinsic::amdgcn_image_load_2darray, false);
10461047
case AMDGPU::BI__builtin_amdgcn_image_load_3d_v4f32_i32:
10471048
case AMDGPU::BI__builtin_amdgcn_image_load_3d_v4f16_i32:
10481049
return EmitAMDGCNImageOverloadedReturnType(
1049-
*this, E, Intrinsic::amdgcn_image_load_3d, false);
1050+
*this, E, Intrinsic::amdgcn_image_load_3d, false);
10501051
case AMDGPU::BI__builtin_amdgcn_image_load_cube_v4f32_i32:
10511052
case AMDGPU::BI__builtin_amdgcn_image_load_cube_v4f16_i32:
10521053
return EmitAMDGCNImageOverloadedReturnType(
1053-
*this, E, Intrinsic::amdgcn_image_load_cube, false);
1054+
*this, E, Intrinsic::amdgcn_image_load_cube, false);
10541055
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1d_v4f32_i32:
10551056
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1d_v4f16_i32:
10561057
return EmitAMDGCNImageOverloadedReturnType(
1057-
*this, E, Intrinsic::amdgcn_image_load_mip_1d, false);
1058+
*this, E, Intrinsic::amdgcn_image_load_mip_1d, false);
10581059
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1darray_v4f32_i32:
10591060
case AMDGPU::BI__builtin_amdgcn_image_load_mip_1darray_v4f16_i32:
10601061
return EmitAMDGCNImageOverloadedReturnType(
1061-
*this, E, Intrinsic::amdgcn_image_load_mip_1darray, false);
1062+
*this, E, Intrinsic::amdgcn_image_load_mip_1darray, false);
10621063
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_f32_i32:
10631064
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_v4f32_i32:
10641065
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2d_v4f16_i32:
10651066
return EmitAMDGCNImageOverloadedReturnType(
1066-
*this, E, Intrinsic::amdgcn_image_load_mip_2d, false);
1067+
*this, E, Intrinsic::amdgcn_image_load_mip_2d, false);
10671068
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_f32_i32:
10681069
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_v4f32_i32:
10691070
case AMDGPU::BI__builtin_amdgcn_image_load_mip_2darray_v4f16_i32:
10701071
return EmitAMDGCNImageOverloadedReturnType(
1071-
*this, E, Intrinsic::amdgcn_image_load_mip_2darray, false);
1072+
*this, E, Intrinsic::amdgcn_image_load_mip_2darray, false);
10721073
case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f32_i32:
10731074
case AMDGPU::BI__builtin_amdgcn_image_load_mip_3d_v4f16_i32:
10741075
return EmitAMDGCNImageOverloadedReturnType(
1075-
*this, E, Intrinsic::amdgcn_image_load_mip_3d, false);
1076+
*this, E, Intrinsic::amdgcn_image_load_mip_3d, false);
10761077
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f32_i32:
10771078
case AMDGPU::BI__builtin_amdgcn_image_load_mip_cube_v4f16_i32:
10781079
return EmitAMDGCNImageOverloadedReturnType(
1079-
*this, E, Intrinsic::amdgcn_image_load_mip_cube, false);
1080+
*this, E, Intrinsic::amdgcn_image_load_mip_cube, false);
10801081
case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f32_i32:
10811082
case AMDGPU::BI__builtin_amdgcn_image_store_1d_v4f16_i32:
10821083
return EmitAMDGCNImageOverloadedReturnType(
1083-
*this, E, Intrinsic::amdgcn_image_store_1d, true);
1084+
*this, E, Intrinsic::amdgcn_image_store_1d, true);
10841085
case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f32_i32:
10851086
case AMDGPU::BI__builtin_amdgcn_image_store_1darray_v4f16_i32:
10861087
return EmitAMDGCNImageOverloadedReturnType(
1087-
*this, E, Intrinsic::amdgcn_image_store_1darray, true);
1088+
*this, E, Intrinsic::amdgcn_image_store_1darray, true);
10881089
case AMDGPU::BI__builtin_amdgcn_image_store_2d_f32_i32:
10891090
case AMDGPU::BI__builtin_amdgcn_image_store_2d_v4f32_i32:
10901091
case AMDGPU::BI__builtin_amdgcn_image_store_2d_v4f16_i32:
10911092
return EmitAMDGCNImageOverloadedReturnType(
1092-
*this, E, Intrinsic::amdgcn_image_store_2d, true);
1093+
*this, E, Intrinsic::amdgcn_image_store_2d, true);
10931094
case AMDGPU::BI__builtin_amdgcn_image_store_2darray_f32_i32:
10941095
case AMDGPU::BI__builtin_amdgcn_image_store_2darray_v4f32_i32:
10951096
case AMDGPU::BI__builtin_amdgcn_image_store_2darray_v4f16_i32:
10961097
return EmitAMDGCNImageOverloadedReturnType(
1097-
*this, E, Intrinsic::amdgcn_image_store_2darray, true);
1098+
*this, E, Intrinsic::amdgcn_image_store_2darray, true);
10981099
case AMDGPU::BI__builtin_amdgcn_image_store_3d_v4f32_i32:
10991100
case AMDGPU::BI__builtin_amdgcn_image_store_3d_v4f16_i32:
11001101
return EmitAMDGCNImageOverloadedReturnType(
1101-
*this, E, Intrinsic::amdgcn_image_store_3d, true);
1102+
*this, E, Intrinsic::amdgcn_image_store_3d, true);
11021103
case AMDGPU::BI__builtin_amdgcn_image_store_cube_v4f32_i32:
11031104
case AMDGPU::BI__builtin_amdgcn_image_store_cube_v4f16_i32:
11041105
return EmitAMDGCNImageOverloadedReturnType(
1105-
*this, E, Intrinsic::amdgcn_image_store_cube, true);
1106+
*this, E, Intrinsic::amdgcn_image_store_cube, true);
11061107
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1d_v4f32_i32:
11071108
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1d_v4f16_i32:
11081109
return EmitAMDGCNImageOverloadedReturnType(
1109-
*this, E, Intrinsic::amdgcn_image_store_mip_1d, true);
1110+
*this, E, Intrinsic::amdgcn_image_store_mip_1d, true);
11101111
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1darray_v4f32_i32:
11111112
case AMDGPU::BI__builtin_amdgcn_image_store_mip_1darray_v4f16_i32:
11121113
return EmitAMDGCNImageOverloadedReturnType(
1113-
*this, E, Intrinsic::amdgcn_image_store_mip_1darray, true);
1114+
*this, E, Intrinsic::amdgcn_image_store_mip_1darray, true);
11141115
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_f32_i32:
11151116
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_v4f32_i32:
11161117
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2d_v4f16_i32:
11171118
return EmitAMDGCNImageOverloadedReturnType(
1118-
*this, E, Intrinsic::amdgcn_image_store_mip_2d, true);
1119+
*this, E, Intrinsic::amdgcn_image_store_mip_2d, true);
11191120
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_f32_i32:
11201121
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_v4f32_i32:
11211122
case AMDGPU::BI__builtin_amdgcn_image_store_mip_2darray_v4f16_i32:
11221123
return EmitAMDGCNImageOverloadedReturnType(
1123-
*this, E, Intrinsic::amdgcn_image_store_mip_2darray, true);
1124+
*this, E, Intrinsic::amdgcn_image_store_mip_2darray, true);
11241125
case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f32_i32:
11251126
case AMDGPU::BI__builtin_amdgcn_image_store_mip_3d_v4f16_i32:
11261127
return EmitAMDGCNImageOverloadedReturnType(
1127-
*this, E, Intrinsic::amdgcn_image_store_mip_3d, true);
1128+
*this, E, Intrinsic::amdgcn_image_store_mip_3d, true);
11281129
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f32_i32:
1129-
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f16_i32:
1130+
case AMDGPU::BI__builtin_amdgcn_image_store_mip_cube_v4f16_i32:
11301131
return EmitAMDGCNImageOverloadedReturnType(
1131-
*this, E, Intrinsic::amdgcn_image_store_mip_cube, true);
1132+
*this, E, Intrinsic::amdgcn_image_store_mip_cube, true);
11321133
case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f32_f32:
11331134
case AMDGPU::BI__builtin_amdgcn_image_sample_1d_v4f16_f32:
11341135
return EmitAMDGCNImageOverloadedReturnType(
1135-
*this, E, Intrinsic::amdgcn_image_sample_1d, false);
1136+
*this, E, Intrinsic::amdgcn_image_sample_1d, false);
11361137
case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f32_f32:
11371138
case AMDGPU::BI__builtin_amdgcn_image_sample_1darray_v4f16_f32:
11381139
return EmitAMDGCNImageOverloadedReturnType(
1139-
*this, E, Intrinsic::amdgcn_image_sample_1darray, false);
1140+
*this, E, Intrinsic::amdgcn_image_sample_1darray, false);
11401141
case AMDGPU::BI__builtin_amdgcn_image_sample_2d_f32_f32:
11411142
case AMDGPU::BI__builtin_amdgcn_image_sample_2d_v4f32_f32:
11421143
case AMDGPU::BI__builtin_amdgcn_image_sample_2d_v4f16_f32:
11431144
return EmitAMDGCNImageOverloadedReturnType(
1144-
*this, E, Intrinsic::amdgcn_image_sample_2d, false);
1145+
*this, E, Intrinsic::amdgcn_image_sample_2d, false);
11451146
case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_f32_f32:
11461147
case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_v4f32_f32:
11471148
case AMDGPU::BI__builtin_amdgcn_image_sample_2darray_v4f16_f32:
11481149
return EmitAMDGCNImageOverloadedReturnType(
1149-
*this, E, Intrinsic::amdgcn_image_sample_2darray, false);
1150+
*this, E, Intrinsic::amdgcn_image_sample_2darray, false);
11501151
case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f32_f32:
11511152
case AMDGPU::BI__builtin_amdgcn_image_sample_3d_v4f16_f32:
11521153
return EmitAMDGCNImageOverloadedReturnType(
1153-
*this, E, Intrinsic::amdgcn_image_sample_3d, false);
1154+
*this, E, Intrinsic::amdgcn_image_sample_3d, false);
11541155
case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f32_f32:
11551156
case AMDGPU::BI__builtin_amdgcn_image_sample_cube_v4f16_f32:
11561157
return EmitAMDGCNImageOverloadedReturnType(
1157-
*this, E, Intrinsic::amdgcn_image_sample_cube, false);
1158+
*this, E, Intrinsic::amdgcn_image_sample_cube, false);
11581159
case AMDGPU::BI__builtin_amdgcn_mfma_scale_f32_16x16x128_f8f6f4:
11591160
case AMDGPU::BI__builtin_amdgcn_mfma_scale_f32_32x32x64_f8f6f4: {
11601161
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)