编译流程、make命令与nccl-test中的Makefile解析
将源代码转换成计算机可以执行的程序,是一个看似简单实则精妙的过程。对于涉及多种技术(如 CUDA、MPI、NCCL)和复杂依赖的项目,这个过程可能会变得相当曲折,充满了各种意想不到的“坑”。就像我们之前调试 MPI+NCCL CUDA 程序时遇到的那样,从找不到头文件到链接器报错,再到最终的可执行文件格式错误,每一步都可能隐藏着问题。
为了更好地理解和解决这些问题,我们需要深入了解编译的完整流程,以及像 make
这样的构建自动化工具是如何在其中发挥作用的。这篇博客将详细拆解从 .cu
或 .c
源代码文件到最终可执行文件的完整旅程,结合我们实际遇到的错误案例进行分析,并详细介绍 make
命令的原理和常用方法。
一、 编译的完整流程
将人类可读的源代码(如 C++, CUDA C++)转换成机器可执行的二进制代码,大致经历以下几个主要阶段:
1. 预处理 (Preprocessing)
- 任务: 处理源代码中以
#
开头的预处理指令。 - 执行者: 预处理器 (通常是 C Preprocessor -
cpp
,会被编译器自动调用)。 - 具体操作:
- 宏展开: 将
#define
定义的宏替换为其对应的值。 - 头文件包含: 将
#include
指令指定的文件内容(如stdio.h
,mpi.h
,cuda_runtime.h
,nccl.h
)直接插入到源代码中。 - 条件编译: 处理
#ifdef
,#ifndef
,#if
,#else
,#elif
,#endif
等指令,根据条件保留或移除部分代码块。 - 移除注释: 删除源代码中的所有注释 (
// ...
和/* ... */
)。
- 宏展开: 将
- 输出: 生成一个扩展后的、仍然是文本格式的源代码文件(通常带有
.i
或.ii
后缀)。 - 遇到的问题关联:
fatal error: mpi.h: No such file or directory
: 这个错误就发生在预处理阶段!预处理器在处理#include <mpi.h>
时,无法在其配置的头文件搜索路径中找到mpi.h
文件。- 解决方法: 通过编译器的
-I/path/to/include
选项,或者修改 Makefile 中的CFLAGS
/CXXFLAGS
/NVCCFLAGS
或INCLUDES
变量,明确告知预处理器去哪里查找所需的头文件。
2. 编译 (Compilation)
- 任务: 将预处理后的源代码翻译成特定 CPU 或 GPU 架构的汇编代码 (Assembly Code)。
- 执行者: 编译器核心 (如
gcc
的cc1
,g++
的cc1plus
,nvcc
的cicc
等)。 - 具体操作:
- 词法分析: 将代码分解成一个个“单词”(token)。
- 语法分析: 检查代码是否符合语言的语法规则,构建抽象语法树 (AST)。
- 语义分析: 检查代码的语义是否正确(如类型匹配、变量声明等)。
- 优化: 进行各种代码优化(如循环展开、常量折叠、死代码消除等),目的是提高最终程序的运行效率。优化级别由
-O0
,-O1
,-O2
,-O3
等选项控制。 - 代码生成: 将优化后的中间表示转换成目标架构的汇编指令。
- 输出: 生成汇编代码文件(通常带有
.s
后缀)。 nvcc
的特殊性: 对于.cu
文件,nvcc
会:- 区分主机代码 (Host Code) 和设备代码 (Device Code) (用
__global__
,__device__
,__host__
等关键字标记)。 - 设备代码编译: 将设备代码编译成 PTX (Parallel Thread Execution - 一种中间汇编语言) 或者直接编译成特定 GPU 架构的 SASS (机器指令)。这一步由
nvcc
内部的设备编译器完成,并受-arch=compute_XX
和-code=sm_XX
(或-gencode
) 标志控制。 - 主机代码编译: 通常调用系统配置的 C++ 编译器(如
g++
)来编译主机代码部分。
- 区分主机代码 (Host Code) 和设备代码 (Device Code) (用
3. 汇编 (Assembly)
- 任务: 将汇编代码转换成机器可以直接理解的二进制目标代码 (Object Code)。
- 执行者: 汇编器 (如
as
,通常由编译器自动调用)。 - 具体操作: 将汇编指令和伪指令翻译成对应的机器码,并生成目标文件的结构(包含代码段、数据段、符号表、重定位信息等)。
- 输出: 生成目标文件(通常带有
.o
后缀)。每个源文件(.c
,.cpp
,.cu
)通常会生成一个对应的.o
文件。
4. 链接 (Linking)
- 任务: 这是生成最终可执行文件或共享库的最后一步。它将多个目标文件 (
.o
) 以及程序所需的库文件(静态库.a
或共享库.so
)合并在一起。 - 执行者: 链接器 (Linker,通常是
ld
,会被编译器驱动程序如gcc
,g++
,nvcc
自动调用)。 - 具体操作:
- 符号解析 (Symbol Resolution): 检查每个目标文件引用的函数或变量(符号),并在其他目标文件或库文件中找到它们的定义。如果找不到定义,就会报 “undefined reference to …” 链接错误。
- 重定位 (Relocation): 将代码和数据安排到最终程序的内存地址空间中,并修正代码中对这些地址的引用。
- 库链接:
- 静态链接: 将静态库 (
.a
) 中的代码完全复制到最终的可执行文件中。优点是程序不依赖外部库文件,缺点是文件体积大,库更新需要重新链接。 - 动态链接: 只在最终可执行文件中记录对共享库 (
.so
) 中函数/变量的引用。程序运行时,操作系统的动态链接器 (ld.so
或dyld
) 负责将共享库加载到内存,并完成最终的地址解析。优点是文件体积小,多个程序可以共享库,库更新方便。缺点是程序运行时需要能找到对应的.so
文件。
- 静态链接: 将静态库 (
- 输出: 生成最终的可执行文件(无后缀或
.exe
)或共享库 (.so
或.dll
)。 - 我们遇到的问题关联:
skipping incompatible libnccl.so
: 链接器在搜索库文件时(根据-L
路径和默认路径),找到了我们指定的libnccl.so
,但发现它的格式或架构不兼容(比如是 ARM 库而我们需要 x86_64),于是跳过它,继续查找,结果可能链接到了系统中的旧版本。Unknown option '-pthread'
或-Wl,-rpath'
:nvcc
作为编译器驱动,在调用底层链接器ld
之前,需要正确解析传递给它的链接选项。它不认识-pthread
,对于-Wl,...
的处理方式也有特定要求,需要使用-Xlinker
来安全传递。Error: Exec format error
: 这通常发生在运行时尝试执行程序时,但根源可能在链接阶段。如果链接了错误的库(架构不匹配),或者链接过程本身出错导致可执行文件结构损坏,操作系统就无法识别其为有效的可执行格式。- 解决方法:
- 确保提供给链接器的库文件架构正确。
- 使用
-L/path/to/lib
告诉链接器去哪里搜索库。 - 使用
-lmylib
告诉链接器具体要链接哪个库。 - 使用
-Xlinker
通过nvcc
安全地传递特定选项给ld
(如-rpath
)。
二、 make
命令
面对包含多个源文件、复杂依赖关系和编译选项的项目,手动执行上面所有的编译链接命令会非常繁琐且容易出错。这时,make
就派上用场了。
- 作用:
make
是一个构建自动化工具。它读取一个名为Makefile
(或makefile
)的特殊文件,根据其中定义的规则,自动地、智能地执行必要的编译和链接命令,来生成最终的目标文件(通常是可执行程序)。 - 核心概念:
Makefile
: 指示make
如何构建项目的**“蓝图”或“食谱”**。- 目标 (Target): 你想要生成的文件名(如
mpi_nccl_example
,all_reduce_perf
,common.o
)。也可以是“伪目标 (Phony Target)”(如all
,clean
),它们不对应实际文件,而是代表一组动作。 - 依赖 (Dependency): 构建一个目标所需要的其他文件或目标。例如,生成
mpi_nccl_example
需要mpi_nccl_example.o
和其他.o
文件以及库文件;生成mpi_nccl_example.o
需要mpi_nccl_example.cu
和相关的头文件。 - 规则 (Rule) / 命令 (Command): 如何从依赖生成目标的具体 Shell 命令。规则必须以 Tab 键 开头(这是 Makefile 语法的一个怪癖)。
# 这是一个规则示例 target: dependency1 dependency2 <Tab>command1 <Tab>command2
- 工作原理:
make
读取Makefile
。- 你指定一个目标(或使用默认目标,通常是 Makefile 中的第一个目标)。
make
检查这个目标的依赖是否存在以及是否比目标更新(根据文件的最后修改时间)。- 如果任何依赖不存在或者比目标更新,
make
就会认为目标需要重新构建。 - 在重新构建目标之前,
make
会递归地检查所有依赖,确保它们也是最新的。如果依赖也需要更新,会先执行更新依赖的规则。 - 一旦所有依赖都满足且是更新的,
make
就执行该目标对应的规则中的命令来生成或更新目标文件。 - 如果目标及其所有依赖都已经是最新的,
make
就什么也不做,直接报告 “target is up to date.”。
- 智能之处:
make
的核心价值在于其增量构建能力。它只重新编译那些其依赖发生变化的文件,大大节省了大型项目的构建时间。
make
的常用方法:
make
: 执行Makefile
中的第一个目标(通常约定为构建整个项目的默认目标,比如all
)。make <target_name>
: 执行名为<target_name>
的目标及其依赖的规则。例如make mpi_nccl_example
。make clean
: 执行名为clean
的伪目标对应的规则,通常用于删除所有编译生成的文件(.o
文件、可执行文件等),以便进行全新的编译。这是一个非常常用的命令。make VAR=value <target>
: 在执行make
命令时,临时覆盖Makefile
中定义的变量VAR
的值。这非常有用!- 我们遇到的应用:
make MPI=1 CUDA_HOME=... NCCL_HOME=... MPI_HOME=...
MPI=1
: 覆盖了 Makefile 中可能存在的MPI ?= 0
的默认值,使得ifeq ($(MPI), 1)
条件为真,从而启用了 MPI 相关的编译和链接标志。CUDA_HOME=...
/NCCL_HOME=...
/MPI_HOME=...
: 覆盖了 Makefile 中对这些路径变量的默认设置(如CUDA_HOME ?= /usr/local/cuda
),让make
使用我们指定的本地安装路径。
- 我们遇到的应用:
好的,我们来详细解读这个 nccl-tests
的 src/Makefile
文件,并解答你关于 make
命令如何工作的疑问。
三、 nccl-test 中 Makefile 文件内容解读
一个典型的 Makefile 主要包含以下几个部分:
1. 变量定义 (Variables):
- 设置默认值:
CUDA_HOME ?= /usr/local/cuda PREFIX ?= /usr/local VERBOSE ?= 0 DEBUG ?= 0 BUILDDIR ?= ../build
?=
的意思是:如果这个变量还没有被定义(例如,没有通过命令行make VAR=value
传递进来,也没有在之前的 Makefile 行中定义),那么就使用这个默认值。否则,使用已经定义的值。- 这就是为什么你在命令行
make CUDA_HOME=... NCCL_HOME=...
可以覆盖这些默认值的原因! 命令行传入的变量定义优先级更高。
- 基于其他变量推导:
CUDA_LIB ?= $(CUDA_HOME)/lib64 CUDA_INC ?= $(CUDA_HOME)/include NVCC ?= $(CUDA_HOME)/bin/nvcc
$(VAR_NAME)
或${VAR_NAME}
用于引用变量的值。这里根据CUDA_HOME
推导出了库路径、包含路径和nvcc
命令的路径。
- 执行 Shell 命令获取值:
CUDA_VERSION = $(strip $(shell which $(NVCC) >/dev/null && $(NVCC) --version | ... )) CUDA_MAJOR = $(shell echo $(CUDA_VERSION) | cut -d "." -f 1) CUDA_MINOR = $(shell echo $(CUDA_VERSION) | cut -d "." -f 2)
$(shell ...)
: 这个函数会执行括号里的 Shell 命令,并将其标准输出作为变量的值。这里用来自动检测 CUDA 版本。
- 编译器和链接器标志:
NVCUFLAGS := -ccbin $(CXX) $(NVCC_GENCODE) -std=c++11 CXXFLAGS := -std=c++11 LDFLAGS := -L${CUDA_LIB} -lcudart -lrt NVLDFLAGS := -L${CUDA_LIB} -l${CUDARTLIB} -lrt
:=
是一种赋值方式,它会立即展开右侧变量的值。NVCUFLAGS
: 传递给nvcc
的编译标志。-ccbin $(CXX)
指定nvcc
使用哪个主机 C++ 编译器 (通常是g++
)。$(NVCC_GENCODE)
是下面条件判断生成的 GPU 架构标志。CXXFLAGS
: 传递给主机 C++ 编译器 ($(CXX)
) 的标志。LDFLAGS
: 通用的链接器标志 (这个 Makefile 里好像没直接用到)。NVLDFLAGS
: 传递给nvcc
用于链接阶段的标志。这里包含了 CUDA 运行时库的路径 (-L
) 和库名 (-l
)。
2. 条件语句 (Conditional Statements):
- 根据 CUDA 版本选择 GPU 架构:
ifeq ($(shell test ...),0) # 判断 CUDA 版本条件 NVCC_GENCODE ?= ... # 如果满足条件,设置默认架构 else ifeq (...) ... else ... endif
- 根据检测到的
CUDA_MAJOR
和CUDA_MINOR
版本,自动选择合适的-gencode
参数,确保生成的代码能在对应版本的 CUDA 支持的 GPU 上运行。?=
同样允许你在命令行覆盖这个默认选择。
- 根据检测到的
- 根据
DEBUG
变量设置优化级别:ifeq ($(DEBUG), 0) NVCUFLAGS += -O3 -g # 非 Debug 模式,使用 O3 优化 CXXFLAGS += -O3 -g else NVCUFLAGS += -O0 -G -g # Debug 模式,关闭优化,加调试信息 CXXFLAGS += -O0 -g -ggdb3 endif
+=
是追加赋值。根据你是否在make
命令中设置DEBUG=1
,会自动调整编译优化级别和调试信息。
- 根据
VERBOSE
变量设置警告:ifneq ($(VERBOSE), 0) # 如果 VERBOSE 不等于 0 NVCUFLAGS += -Xcompiler -Wall,-Wextra,-Wno-unused-parameter # 添加更多警告 else .SILENT: # 否则,让 make 安静执行命令,不打印命令本身 endif
- 根据
NCCL_HOME
添加 NCCL 路径:ifneq ($(NCCL_HOME), "") # 如果 NCCL_HOME 不为空 (通过命令行传入了) NVCUFLAGS += -I$(NCCL_HOME)/include/ # 添加 NCCL 头文件路径 NVLDFLAGS += -L$(NCCL_HOME)/lib # 添加 NCCL 库文件路径 endif
- 根据
MPI
变量添加 MPI 路径和标志:
这就是为什么设置ifeq ($(MPI), 1) # 如果 MPI=1 (通过命令行传入了) NVCUFLAGS += -DMPI_SUPPORT -I$(MPI_HOME)/include # 添加 MPI 宏和头文件路径 NVLDFLAGS += -L$(MPI_HOME)/lib -L$(MPI_HOME)/lib64 -lmpi # 添加 MPI 库路径和库名 endif
MPI=1
和MPI_HOME
很重要的原因,Makefile 会自动处理路径。 - 添加要链接的库:
LIBRARIES += nccl # 将 nccl 添加到 LIBRARIES 变量 NVLDFLAGS += $(LIBRARIES:%=-l%) # 将 LIBRARIES 中的每个库名加上 -l 前缀,追加到 NVLDFLAGS
$(LIBRARIES:%=-l%)
是一个替换引用,它将LIBRARIES
变量中的每个词(这里是nccl
)前面加上-l
,得到-lnccl
。
3. 规则定义 (Rules):
- 变量推导 (文件列表):
DST_DIR := $(BUILDDIR) # 目标目录 SRC_FILES := $(wildcard *.cu) # 查找当前目录下所有 .cu 文件 OBJ_FILES := $(SRC_FILES:%.cu=${DST_DIR}/%.o) # 将 .cu 文件列表转换成对应的 .o 文件列表 BIN_FILES_LIST := all_reduce all_gather ... # 定义要生成的可执行文件前缀 BIN_FILES := $(BIN_FILES_LIST:%=${DST_DIR}/%_perf) # 生成最终可执行文件的完整路径列表
- 默认目标 (
build
):build: ${BIN_FILES}
- 当你只运行
make
时,它会尝试构建build
这个目标。 build
目标的依赖是${BIN_FILES}
变量中列出的所有可执行文件(如../build/all_reduce_perf
,../build/all_gather_perf
等)。make
会依次尝试构建这些依赖的可执行文件。
- 当你只运行
- 伪目标 (
clean
):.PHONY: build clean # 告诉 make 这些是伪目标,不对应实际文件 clean:rm -rf ${DST_DIR} # 执行删除目标目录的命令
- 运行
make clean
会执行rm -rf ../build
命令。
- 运行
- 包含其他 Makefile (
include
):include ../verifiable/verifiable.mk
- 将另一个 Makefile 的内容包含进来,通常用于模块化管理。
- 模式规则 (Pattern Rules): 这是 Makefile 的精髓之一,定义了如何从一类文件生成另一类文件。
- 编译
.cu
文件生成.o
文件:${DST_DIR}/%.o: %.cu common.h $(TEST_VERIFIABLE_HDRS)@printf "Compiling %-35s > %s\n" $< $@@mkdir -p ${DST_DIR}# 这是你之前修改过的行,强制添加了 MPI include 路径$(NVCC) -o $@ $(NVCUFLAGS) -I/lihongliang/zm/build/include -c $<
${DST_DIR}/%.o: %.cu ...
: 定义了如何从任意一个.cu
文件生成对应的放在${DST_DIR}
目录下的.o
文件。%
是通配符。%.cu common.h $(TEST_VERIFIABLE_HDRS)
: 定义了依赖关系,即.o
文件依赖于对应的.cu
文件、common.h
以及其他头文件。如果这些依赖文件有任何一个比.o
文件更新,就需要重新执行命令。$<
: 自动变量,代表规则中的第一个依赖(这里是.cu
文件名)。$@
: 自动变量,代表规则中的目标(这里是.o
文件名)。@
: 命令前的@
表示执行时不打印命令本身,只打印printf
的输出。mkdir -p ${DST_DIR}
: 确保目标目录存在。$(NVCC) ... -c $< -o $@
: 核心编译命令。-c
表示只编译不链接,生成目标文件。
- 编译 C++ 文件 (
timer.cc
):${DST_DIR}/timer.o: timer.cc timer.h@printf "Compiling %-35s > %s\n" $< $@@mkdir -p ${DST_DIR}$(CXX) $(CXXFLAGS) -o $@ -c timer.cc
- 类似地,定义了如何用 C++ 编译器 (
$(CXX)
通常是g++
) 和$(CXXFLAGS)
来编译timer.cc
。
- 类似地,定义了如何用 C++ 编译器 (
- 链接生成最终可执行文件 (
_perf
):${DST_DIR}/%_perf:${DST_DIR}/%.o ${DST_DIR}/common.o ${DST_DIR}/timer.o $(TEST_VERIFIABLE_OBJS)@printf "Linking %-35s > %s\n" $< $@@mkdir -p ${DST_DIR}$(NVCC) -o $@ $(NVCUFLAGS) $^ ${NVLDFLAGS}
${DST_DIR}/%_perf: ...
: 定义了如何从对应的.o
文件 (%.o
) 以及common.o
,timer.o
和其他依赖的.o
文件 ($(TEST_VERIFIABLE_OBJS)
) 来生成最终的_perf
可执行文件。$^
: 自动变量,代表规则中的所有依赖(这里是所有的.o
文件)。$(NVCC) ... $^ ${NVLDFLAGS}
: 核心链接命令。nvcc
在没有-c
时会执行链接操作。它将所有依赖的.o
文件 ($^
) 和在$(NVLDFLAGS)
中指定的库(如-lcudart
,-lrt
,-lnccl
,-lmpi
)链接起来,生成目标可执行文件$@
。
- 编译
四、 一些疑问
-
make MPI=1 CUDA_HOME=...
相当于将 Makefile 中的变量替换了?- 是的,但不完全是替换,更准确地说是“覆盖默认值”或“在执行前定义”。 Makefile 中的
?=
定义的是默认值。当你在命令行上传入VAR=value
时,make
会在解析 Makefile 之前就记住这个定义。当 Makefile 解析到VAR ?= default_value
时,因为它发现VAR
已经被命令行定义了,所以?=
的赋值操作就不会执行,从而使用了你命令行传入的值。如果 Makefile 中用的是=
或:=
,命令行传入的值仍然会覆盖它们(命令行优先级最高)。
- 是的,但不完全是替换,更准确地说是“覆盖默认值”或“在执行前定义”。 Makefile 中的
-
编译器/Make 怎么知道应该生成什么文件,放到什么地方去?
- 通过 Makefile 中的规则定义的!
- 生成什么文件: 由规则中的目标 (Target) 指定。例如,
build: ${BIN_FILES}
指定了最终要生成${BIN_FILES}
列表中的所有文件。${DST_DIR}/%.o: %.cu ...
指定了要为每个.cu
生成一个.o
文件。${DST_DIR}/%_perf: ...
指定了要为每个测试生成一个_perf
可执行文件。 - 放到什么地方: 由规则目标和命令中的路径指定。
${DST_DIR}/%.o
: 明确指定了.o
文件要放在${DST_DIR}
(也就是../build
) 目录下。${DST_DIR}/%_perf
: 明确指定了可执行文件也要放在${DST_DIR}
目录下。- 编译/链接命令中的
-o $@
选项明确告诉编译器/链接器将输出文件命名为规则的目标$@
(包含了完整的路径)。 @mkdir -p ${DST_DIR}
命令确保了目标目录在写入文件之前就存在。
- 生成什么文件: 由规则中的目标 (Target) 指定。例如,
- 通过 Makefile 中的规则定义的!
-
我在命令行输入
make
命令之后调用的是我当前目录下的makefile
文件吗?- 是的,默认情况下是这样。 当你运行
make
时,它会按照以下顺序在你当前工作目录查找 Makefile 文件:GNUmakefile
(GNU Make 特有)makefile
Makefile
- 它会使用找到的第一个文件。所以,通常我们将 Makefile 命名为
Makefile
(首字母大写)。 - 你也可以使用
-f <filename>
选项来明确指定使用哪个 Makefile 文件,例如make -f MyMakefile
。 - 对于
nccl-tests
: 当你在nccl-tests-master
目录下运行make
时,它读取的是顶层的Makefile
。这个顶层 Makefile 通常会通过make -C src ...
或include src/Makefile
的方式去调用或包含src
子目录下的Makefile
来执行实际的编译链接工作。所以,最终执行编译规则的是src/Makefile
(或者它包含的其他文件)。
- 是的,默认情况下是这样。 当你运行
五、Makefile代码
#
# Copyright (c) 2015-2022, NVIDIA CORPORATION. All rights reserved.
#
# See LICENSE.txt for license information
#CUDA_HOME ?= /usr/local/cuda
PREFIX ?= /usr/local
VERBOSE ?= 0
DEBUG ?= 0CUDA_LIB ?= $(CUDA_HOME)/lib64
CUDA_INC ?= $(CUDA_HOME)/include
NVCC ?= $(CUDA_HOME)/bin/nvcc
CUDARTLIB ?= cudartCUDA_VERSION = $(strip $(shell which $(NVCC) >/dev/null && $(NVCC) --version | grep release | sed 's/.*release //' | sed 's/\,.*//'))
CUDA_MAJOR = $(shell echo $(CUDA_VERSION) | cut -d "." -f 1)
CUDA_MINOR = $(shell echo $(CUDA_VERSION) | cut -d "." -f 2)# Better define NVCC_GENCODE in your environment to the minimal set
# of archs to reduce compile time.
ifeq ($(shell test "0$(CUDA_MAJOR)" -eq 12 -a "0$(CUDA_MINOR)" -ge 8 -o "0$(CUDA_MAJOR)" -ge 13; echo $$?),0)
# Include Blackwell support if we're using CUDA12.8 or above
NVCC_GENCODE ?= -gencode=arch=compute_80,code=sm_80 \-gencode=arch=compute_90,code=sm_90 \-gencode=arch=compute_100,code=sm_100 \-gencode=arch=compute_120,code=sm_120 \-gencode=arch=compute_120,code=compute_120
else ifeq ($(shell test "0$(CUDA_MAJOR)" -ge 12; echo $$?),0)
NVCC_GENCODE ?= -gencode=arch=compute_60,code=sm_60 \-gencode=arch=compute_61,code=sm_61 \-gencode=arch=compute_70,code=sm_70 \-gencode=arch=compute_80,code=sm_80 \-gencode=arch=compute_90,code=sm_90 \-gencode=arch=compute_90,code=compute_90
else ifeq ($(shell test "0$(CUDA_MAJOR)" -ge 11; echo $$?),0)
NVCC_GENCODE ?= -gencode=arch=compute_60,code=sm_60 \-gencode=arch=compute_61,code=sm_61 \-gencode=arch=compute_70,code=sm_70 \-gencode=arch=compute_80,code=sm_80 \-gencode=arch=compute_80,code=compute_80
else
NVCC_GENCODE ?= -gencode=arch=compute_35,code=sm_35 \-gencode=arch=compute_50,code=sm_50 \-gencode=arch=compute_60,code=sm_60 \-gencode=arch=compute_61,code=sm_61 \-gencode=arch=compute_70,code=sm_70 \-gencode=arch=compute_70,code=compute_70
endifNVCUFLAGS := -ccbin $(CXX) $(NVCC_GENCODE) -std=c++11
CXXFLAGS := -std=c++11LDFLAGS := -L${CUDA_LIB} -lcudart -lrt
NVLDFLAGS := -L${CUDA_LIB} -l${CUDARTLIB} -lrtifeq ($(DEBUG), 0)
NVCUFLAGS += -O3 -g
CXXFLAGS += -O3 -g
else
NVCUFLAGS += -O0 -G -g
CXXFLAGS += -O0 -g -ggdb3
endififneq ($(VERBOSE), 0)
NVCUFLAGS += -Xcompiler -Wall,-Wextra,-Wno-unused-parameter
else
.SILENT:
endif.PHONY: build cleanBUILDDIR ?= ../build
ifneq ($(NCCL_HOME), "")
NVCUFLAGS += -I$(NCCL_HOME)/include/
NVLDFLAGS += -L$(NCCL_HOME)/lib
endififeq ($(MPI), 1)
NVCUFLAGS += -DMPI_SUPPORT -I$(MPI_HOME)/include
NVLDFLAGS += -L$(MPI_HOME)/lib -L$(MPI_HOME)/lib64 -lmpi
endif
ifeq ($(MPI_IBM),1)
NVCUFLAGS += -DMPI_SUPPORT
NVLDFLAGS += -lmpi_ibm
endif
LIBRARIES += nccl
NVLDFLAGS += $(LIBRARIES:%=-l%)DST_DIR := $(BUILDDIR)
SRC_FILES := $(wildcard *.cu)
OBJ_FILES := $(SRC_FILES:%.cu=${DST_DIR}/%.o)
BIN_FILES_LIST := all_reduce all_gather broadcast reduce_scatter reduce alltoall scatter gather sendrecv hypercube
BIN_FILES := $(BIN_FILES_LIST:%=${DST_DIR}/%_perf)build: ${BIN_FILES}clean:rm -rf ${DST_DIR}TEST_VERIFIABLE_SRCDIR := ../verifiable
TEST_VERIFIABLE_BUILDDIR := $(BUILDDIR)/verifiable
include ../verifiable/verifiable.mk${DST_DIR}/%.o: %.cu common.h $(TEST_VERIFIABLE_HDRS)@printf "Compiling %-35s > %s\n" $< $@@mkdir -p ${DST_DIR}$(NVCC) -o $@ $(NVCUFLAGS) -I/path/to/mpi/bulid/include -c $<${DST_DIR}/timer.o: timer.cc timer.h@printf "Compiling %-35s > %s\n" $< $@@mkdir -p ${DST_DIR}$(CXX) $(CXXFLAGS) -o $@ -c timer.cc${DST_DIR}/%_perf:${DST_DIR}/%.o ${DST_DIR}/common.o ${DST_DIR}/timer.o $(TEST_VERIFIABLE_OBJS)@printf "Linking %-35s > %s\n" $< $@@mkdir -p ${DST_DIR}$(NVCC) -o $@ $(NVCUFLAGS) $^ ${NVLDFLAGS}