CUDA 线程是时间切片的。这是什么意思？

Question

CUDA 编程指南 4.1 的 B.10 节解释说：

[...] 设备完成运行所花费的时钟周期数 执行线程，[不同于] 的时钟周期数 设备实际花费在执行线程指令上。以前的号码 大于后者[...]

我了解第一个是完成线程执行的挂钟时间。第二次是第一次减去线程空闲的时间。当线程的指令需要等待先前指令的结果（指令依赖）、等待内存中的操作数值或在同步点等待时，线程将处于空闲状态。

指南接着说：

前一个数字大于后者，因为线程是时间 切片。

在这种情况下时间切片是什么意思？说线程是时间切片是什么意思？

请注意，该术语不会出现在指南的其他任何地方。 （如果我在这里遗漏了一些明显的上下文，请原谅我，我不是以英语为母语的人。）

Answer 1

在这种情况下，时间切片是指多个 warp 运行在多处理器 (SM) 上的事实，并且 SM 在执行过程中在 warp 之间切换以隐藏延迟。这与传统 CPU 线程中的抢占不同；也不等同于流水线。

如果你有这样的代码：

if (threadIdx.x == 0 && blockIdx.x == 0) x = clock();

// other work done by all threads

if (threadIdx.x == 0 && blockIdx.x == 0) y = clock();

如果在 SM 上运行多个 warp，那么y-x 的值将大于在线程 0 中执行的实际时间（== warp 0）。这不仅仅是因为线程 0 必须等待指令或内存访问的结果，还因为执行其他 warp 所花费的时间。

编程指南中这句话的重点是，使用 clock() 进行绝对时序或延迟测量是很棘手的。

Answer 2

当多个线程正在运行并且它们必须共享一个处理单元时，通常的处理方式是给每个线程一个固定的最大运行时间（您的时间片），然后它被抢占，另一个线程得到运行一段时间。因此，如果您的线程无法在一个时间片内完成其工作，那么它可能不得不等到再次轮到它。这需要多长时间取决于并行线程的数量、它们在做什么、调度程序的实现方式以及可用的处理资源。

Answer 3

在这种情况下，时间切片也意味着preemption。

您可以将时间片视为总可用执行时间的某个百分比。

实际上，您的线程被安排运行一段时间，但是如果需要执行其他线程，调度程序可能只会给您一个较小的时间片。