aparler/c-llm
1
0
Fork 0
mirror of https://github.com/karpathy/llm.c.git synced 2026-07-28 20:35:09 -04:00
Code Issues Projects Releases Packages Wiki Activity Actions

uint and designated initializers non-standard for C++

Browse source 
uint isn't defined by the MS compiler and designated initializers aren't part of C++ until C++ 20.
This commit is contained in:
Ross Wheeler 2024-04-30 18:46:54 -07:00
parent 40da8cf233
commit 229d70b2d5
1 changed files with 7 additions and 7 deletions
Download patch file Download diff file Expand all files Collapse all files

14
train_gpt2.cu
View file

@ -409,11 +409,11 @@ MultiGpuConfig multi_gpu_config_init(int *argc, char ***argv) {
return result;
#else
printf("Multi-GPU support is disabled. Using a single GPU.\n");
return MultiGpuConfig{
.process_rank = 0,
.num_processes = 1,
.local_device_idx = 0,
};
MultiGpuConfig result;
result.process_rank = 0;
result.num_processes = 1;
result.local_device_idx = 0;
return result;
#endif
}
@ -765,7 +765,7 @@ __global__ void layernorm_backward_kernel6(floatX* dinp, floatX* dweight, floatX
dbias_shared[i] = 0.0f;
dweight_shared[i] = 0.0f;
}
uint *tmp_flag = (uint*)(shared + C*2);
unsigned int *tmp_flag = (unsigned int*)(shared + C*2);
__syncthreads();
int warps_in_grid = gridDim.x * warp.meta_group_size();
@ -818,7 +818,7 @@ __global__ void layernorm_backward_kernel6(floatX* dinp, floatX* dweight, floatX
__syncthreads();
float* scratch_dbias = scratch;
float* scratch_dweight = scratch + C;
uint* scratchFlag = (uint*)(scratch + (2 * C));
unsigned int* scratchFlag = (unsigned int*)(scratch + (2 * C));
for(int i = threadIdx.x; i < C; i+= blockDim.x) {
atomicAdd(&scratch_dbias[i], dbias_shared[i]);
atomicAdd(&scratch_dweight[i], dweight_shared[i]);