Tại sao kích thước bộ nhớ không đổi bị hạn chế trong CUDA?

Question

Theo "CUDA C Programming Guide", quyền truy cập bộ nhớ không đổi chỉ có lợi nếu bộ nhớ cache liên bộ đa bộ được nhấn (Mục 5.3.2.4) . Nếu không, có thể có nhiều yêu cầu bộ nhớ hơn cho một nửa sợi dọc hơn trong trường hợp đọc bộ nhớ toàn cầu kết hợp. Vậy tại sao kích thước bộ nhớ không đổi được giới hạn ở 64 KB?

Một câu hỏi khác để không hỏi hai lần. Theo như tôi hiểu, trong kiến ​​trúc Fermi, cache kết cấu được kết hợp với bộ đệm L2. Việc sử dụng kết cấu vẫn có ý nghĩa hay đọc bộ nhớ toàn cục được lưu trữ theo cùng một cách?

---

liên tục Memory (Mục 5.3.2.4)

Không gian bộ nhớ liên tục nằm trong bộ nhớ điện thoại và được lưu trữ trong bộ nhớ cache liên tục đề cập tại mục F.3.1 và F. 4.1.

Đối với các thiết bị có khả năng tính toán 1.x, yêu cầu bộ nhớ liên tục cho một sợi dọc được chia thành hai yêu cầu, một cho mỗi nửa dọc, được phát hành độc lập.

Yêu cầu sau đó được chia thành nhiều yêu cầu riêng biệt vì có các địa chỉ bộ nhớ khác nhau trong yêu cầu ban đầu, giảm thông lượng theo hệ số bằng số yêu cầu riêng biệt.

Các yêu cầu kết quả sau đó được phục vụ tại thông lượng của bộ đệm không đổi trong trường hợp có lần truy cập bộ nhớ cache hoặc thông lượng bộ nhớ thiết bị khác.

Answer 1

Kích thước bộ nhớ không đổi là 64 KB cho khả năng tính toán 1.0-3.0 thiết bị. Bộ làm việc bộ đệm chỉ là 8KB (xem Hướng dẫn lập trình CUDA v4.2 Bảng F-2).

Bộ nhớ liên tục được trình điều khiển, trình biên dịch và các biến được khai báo __device__ __constant__. Trình điều khiển sử dụng bộ nhớ liên tục để truyền thông số, kết cấu kết cấu, vv Trình biên dịch sử dụng hằng số trong nhiều hướng dẫn (xem tháo gỡ).

Các biến được đặt trong bộ nhớ không đổi có thể được đọc và ghi bằng các hàm thời gian chạy máy chủ cudaMemcpyToSymbol() và cudaMemcpyFromSymbol() (xem Hướng dẫn lập trình CUDA phần v.2.2 B.2.2). Bộ nhớ không đổi nằm trong bộ nhớ thiết bị nhưng được truy cập thông qua bộ đệm không đổi.

Trên kết cấu Fermi, hằng số, L1 và I-Cache là tất cả các cache cấp 1 trong hoặc xung quanh mỗi SM. Tất cả cache cấp 1 đều truy cập bộ nhớ thiết bị thông qua cache L2.

Giới hạn hằng số 64 KB là mỗi CUmodule là đơn vị biên dịch CUDA. Khái niệm CUmodule được ẩn trong thời gian chạy CUDA nhưng có thể truy cập được bằng API trình điều khiển CUDA.