diff --git a/train_gpt2.cu b/train_gpt2.cu index ac66d99..cde18c1 100644 --- a/train_gpt2.cu +++ b/train_gpt2.cu @@ -1854,7 +1854,7 @@ void gpt2_update(GPT2 *model, float learning_rate, float beta1, float beta2, flo printf("allocated %zu MiB for AdamW optimizer state m\n", (model->num_parameters * sizeof(float)) >> 20); printf("allocated %zu MiB for AdamW optimizer state v\n", (model->num_parameters * sizeof(float)) >> 20); } - + int block_size = 512; int num_blocks = CEIL_DIV(model->num_parameters, block_size); float beta1_correction = 1.0f - powf(beta1, t); @@ -2169,12 +2169,12 @@ int main(int argc, char *argv[]) { cublasCheck(cublasLtCreate(&cublaslt_handle)); cudaCheck(cudaMalloc(&cublaslt_workspace, cublaslt_workspace_size)); - // TF32 precision is equivalent to torch.set_float32_matmul_precision('high') + // TF32 precision is equivalent to torch.set_float32_matmul_precision('high') int enable_tf32 = cuda_arch_major >= 8 ? 1 : 0; cublas_compute_type = enable_tf32 ? CUBLAS_COMPUTE_32F_FAST_TF32 : CUBLAS_COMPUTE_32F; cublasMath_t cublas_math_mode = enable_tf32 ? CUBLAS_TF32_TENSOR_OP_MATH : CUBLAS_DEFAULT_MATH; cublasCheck(cublasSetMathMode(cublas_handle, cublas_math_mode)); - if(cublas_compute_type); // unused in BF16 mode, avoid warning + if(cublas_compute_type); // unused in BF16 mode, avoid warning printf0("| device | %-50s |\n", deviceProp.name); printf0("| TF32 | %-50s |\n", enable_tf32 ? "enabled" : "disabled"); @@ -2316,7 +2316,8 @@ int main(int argc, char *argv[]) { double time_elapsed_s = (end.tv_sec - start.tv_sec) + (end.tv_nsec - start.tv_nsec) / 1e9; total_sum_iteration_time_s += time_elapsed_s; int tokens_per_second = multi_gpu_config.num_processes * (B * T) / time_elapsed_s; - printf0("step %4d/%d: train loss %f (acc %f) (%f ms, %d tok/s)\n", step + 1, train_num_batches, model.mean_loss, model.accumulated_mean_loss, time_elapsed_s * 1000, tokens_per_second); + float accumulated_loss = multi_gpu_config.num_processes == 1 ? model.mean_loss : model.accumulated_mean_loss; + printf0("step %4d/%d: train loss %f (acc %f) (%f ms, %d tok/s)\n", step + 1, train_num_batches, model.mean_loss, accumulated_loss, time_elapsed_s * 1000, tokens_per_second); logger_log_train(&logger, step, model.mean_loss); } // add a total average, for optimizations that are only mild improvements