GPU・環境構築

【CUDA】GPU間Peer-to-Peer通信とNVLink活用：エラー「cudaErrorPeerAccessUnsupported」の解決法

· 18分で読めます · 1 閲覧

問題の概要：GPU間通信のエラーとパフォーマンス低下

マルチGPU環境（例：RTX 3090, A100, H100を2枚以上搭載）で深層学習の分散学習や大規模なシミュレーションを実行する際、以下のような問題に遭遇することがあります。

  • エラーメッセージ: cudaErrorPeerAccessUnsupported: peer access is not supported between these two devices
  • GPU間でデータを転送する際の帯域幅が、PCIeバスの理論値（例：PCIe 4.0 x16で約32GB/s）よりも大幅に低い
  • NVLink対応GPUを搭載しているにもかかわらず、その高速インターフェースが活用されている実感がない

これらの問題は、GPU間の直接通信（Peer-to-Peer, P2P）が正しく設定・活用されていない場合に発生します。特に、NVLinkという専用の高速相互接続技術が利用可能な環境では、その設定がパフォーマンスに決定的な影響を与えます。

原因の解説：P2P通信の制約とNVLinkの条件

CUDAにおけるGPU間のPeer-to-Peer通信は、すべてのハードウェア構成で自動的に有効になるわけではありません。主な原因は以下の3つです。

1. ハードウェアおよびドライバの制限

P2P通信は、同じ物理マシン内の特定のGPU組み合わせでのみサポートされます。異なるアーキテクチャ（例：Tesla V100とGeForce RTX 4090）や、一部の古いGPU間ではサポートされていない場合があります。また、ドライバやCUDA Toolkitのバージョンが古いと、新しいハードウェアの機能を十分に活用できません。

2. PCIeトポロジーの問題

マザーボードのPCIeレーン割り当てや、CPUのPCIeコントローラ（NUMAノード）をまたがるGPU間では、P2P通信が禁止されることがあります。例えば、2つのGPUが異なるCPUソケットに接続されている場合、直接通信できない構成が一般的です。

3. NVLinkの未接続・未認識

NVLinkは物理的なブリッジ（NVLink Bridge）でGPUを接続する必要があります。このブリッジが正しく装着されていない、または対応していないブリッジ（世代違いなど）を使用している場合、NVLinkは有効化されず、通信は低速なPCIeバスにフォールバックします。

解決方法：ステップバイステップでの確認と設定

ステップ1: システムトポロジーとP2P対応の確認

まず、現在のシステムでどのGPU間でP2P通信が可能かを確認します。CUDAサンプルコードまたはnvidia-smiコマンドを使用します。

# nvidia-smiでトポロジーを確認
nvidia-smi topo -m

# 出力例（P2P対応状況が表示される）:
#        GPU0    GPU1    GPU2    GPU3
# GPU0   X      PHB     SOC     SOC
# GPU1   PHB    X       SOC     SOC
# GPU2   SOC    SOC     X       PHB
# GPU3   SOC    SOC     PHB     X
#
# PHB: P2Pは可能だが、パフォーマンスはPCIeホストブリッジに制限
# SOC: P2Pは可能で、同じCPUソケット上にあるため最適
# NODE: P2P不可（異なるNUMAノード）
# NV: NVLink接続（nはリンク数）

また、以下のような簡単なCUDAプログラムでP2Pアクセス可否をテストできます。

// p2p_test.cu (コンパイル: nvcc p2p_test.cu -o p2p_test)
#include <stdio.h>
#include <cuda_runtime.h>

int main() {
    int num_devices;
    cudaGetDeviceCount(&num_devices);
    printf("Number of GPUs: %dn", num_devices);

    for (int i = 0; i < num_devices; i++) {
        for (int j = 0; j < num_devices; j++) {
            if (i == j) continue;
            int can_access;
            cudaDeviceCanAccessPeer(&can_access, i, j);
            printf("GPU%d -> GPU%d : P2P Access %sn", 
                   i, j, can_access ? "OK" : "NO");
        }
    }
    return 0;
}

ステップ2: NVLink状態の確認と物理接続の見直し

nvidia-smiコマンドでNVLinkの状態を詳細に確認します。

# NVLinkの帯域幅とエラー情報を表示
nvidia-smi nvlink -c
nvidia-smi nvlink -s

# 各GPUのNVLink情報を表示（GPUインデックスを指定）
nvidia-smi -i 0 --query-gpu=nvlink_bandwidth.total --format=csv

出力にNVLinkリンクが表示されない、または帯域幅が0の場合は、以下の点を確認してください。

  1. GPUがNVLinkに対応しているか（コンシューマ向けGeForceの多くは非対応）
  2. 正しい世代のNVLinkブリッジが確実に装着されているか
  3. マザーボードのスロット間隔がブリッジの物理的形状と合っているか

ステップ3: CUDAプログラム内でのP2P有効化

ハードウェア的にP2Pが可能でも、ソフトウェア側で明示的に有効化する必要があります。以下のコードは、すべての可能なGPUペア間でP2Pアクセスを有効にする典型的な方法です。

// enable_p2p.cu
#include <cuda_runtime.h>
#include <stdio.h>

int enableP2P(int num_gpus) {
    for (int i = 0; i < num_gpus; i++) {
        cudaSetDevice(i);
        for (int j = 0; j < num_gpus; j++) {
            if (i == j) continue;
            
            int can_access;
            cudaError_t err = cudaDeviceCanAccessPeer(&can_access, i, j);
            if (err != cudaSuccess) {
                printf("Failed to query P2P access between GPU%d and GPU%d: %sn",
                       i, j, cudaGetErrorString(err));
                continue;
            }
            
            if (can_access) {
                err = cudaDeviceEnablePeerAccess(j, 0); // フラグは0
                if (err == cudaSuccess) {
                    printf("Enabled P2P access GPU%d -> GPU%dn", i, j);
                } else if (err == cudaErrorPeerAccessAlreadyEnabled) {
                    printf("P2P access already enabled GPU%d -> GPU%dn", i, j);
                } else {
                    printf("Failed to enable P2P access GPU%d -> GPU%d: %sn",
                           i, j, cudaGetErrorString(err));
                }
            } else {
                printf("P2P access not supported GPU%d -> GPU%dn", i, j);
            }
        }
    }
    return 0;
}

この設定後、cudaMemcpyPeerやcudaMemcpyPeerAsync関数を使用して、GPU間で直接データを転送できるようになります。

ステップ4: フレームワークでの設定（PyTorch例）

深層学習フレームワークを使用する場合、多くの場合は内部でP2Pを自動的に扱いますが、明示的に環境変数を設定することで最適化できる場合があります。

# PyTorchでNVLinkを最大限活用するための環境変数設定例
export NCCL_DEBUG=INFO
export NCCL_P2P_DISABLE=0  # P2Pを明示的に有効化（デフォルト）
export NCCL_P2P_LEVEL=NVL  # NVLink経由の通信を優先
export NCCL_IB_DISABLE=1   # InfiniBandを使用しない場合（クラスタ環境でない場合）

# PyTorchスクリプト内での確認
import torch
print(f"Available GPUs: {torch.cuda.device_count()}")
for i in range(torch.cuda.device_count()):
    for j in range(torch.cuda.device_count()):
        if i != j:
            can_access = torch.cuda.can_device_access_peer(i, j)
            print(f"GPU{i} can access GPU{j}: {can_access}")

コード例・コマンド例

ベンチマーク：P2P転送速度の測定

以下のコードで、PCIe経由とNVLink経由（設定されている場合）の実際の転送速度を比較できます。

// p2p_bandwidth.cu (簡略版)
#include <cuda_runtime.h>
#include <stdio.h>
#include <chrono>

void benchmarkP2P(int src_gpu, int dst_gpu, size_t size) {
    cudaSetDevice(src_gpu);
    float* src_data;
    cudaMalloc(&src_data, size);
    
    cudaSetDevice(dst_gpu);
    float* dst_data;
    cudaMalloc(&dst_data, size);
    
    // ウォームアップ
    cudaMemcpyPeer(dst_data, dst_gpu, src_data, src_gpu, size);
    
    auto start = std::chrono::high_resolution_clock::now();
    int iterations = 100;
    for (int i = 0; i < iterations; i++) {
        cudaMemcpyPeer(dst_data, dst_gpu, src_data, src_gpu, size);
    }
    cudaDeviceSynchronize();
    auto end = std::chrono::high_resolution_clock::now();
    
    std::chrono::duration<double> elapsed = end - start;
    double bandwidth = (size * iterations * 2) / (elapsed.count() * 1e9); // GB/s
    printf("GPU%d -> GPU%d Bandwidth: %.2f GB/sn", 
           src_gpu, dst_gpu, bandwidth);
    
    cudaFree(src_data);
    cudaFree(dst_data);
}

トラブルシューティングコマンド集

# 基本的なシステム情報確認
nvidia-smi
nvidia-smi topo -m

# NVLinkの詳細状態（A100, H100等）
nvidia-smi nvlink -i 0 -c   # GPU0のNVLinkカウンタを表示
nvidia-smi nvlink -i 0 -g   # GPU0のNVLinkエラーカウンタをリセット

# CUDAドライバとランタイムのバージョン確認
cat /usr/local/cuda/version.txt
nvcc --version

# PCIeリンク速度確認（Gen3 x16なら8.0GT/s、Gen4なら16.0GT/s）
lspci -vv | grep -i nvidia -A 20 | grep LnkSta

まとめ・補足情報

CUDA Peer-to-Peer通信とNVLinkの活用は、マルチGPUシステムのパフォーマンスを最大限引き出すための重要な技術です。要点をまとめます。

  1. 事前確認が重要: nvidia-smi topo -mでハードウェア的なP2P対応を必ず確認しましょう。これにより、ソフトウェア設定以前の問題を早期に発見できます。
  2. NVLinkは物理接続が命: 対応GPUと正しいブリッジの装着が必須です。特に、サーバー向けGPU（A100, H100）とワークステーション向けGPU（RTX 6000 Ada）では、NVLinkの形態（NVLink Bridge vs NVLink Connector）が異なる場合があるので注意が必要です。
  3. ソフトウェアでの明示的有効化: ハードウェアが対応していても、CUDAではcudaDeviceEnablePeerAccess()を呼び出して初めてP2P通信が可能になります。フレームワークを使用する場合も、環境変数などで最適化を図りましょう。
  4. パフォーマンスベンチマークの実施: 理論値と実際の転送速度には乖離が生じることがあります。実際のワークロードに近い条件でベンチマークを実施し、システムが期待通りに動作しているかを検証してください。

最後に、一般的な注意点として、P2Pアクセスを有効にするとGPUメモリがロックされ、他のプロセスからの使用が制限される場合があります。また、すべてのアプリケーションがP2P通信の恩恵を受けるわけではなく、GPU間で頻繁に大規模データを交換するワークロード（大規模モデルの並列訓練、分子動力学シミュレーションなど）で最も効果を発揮します。自身のユースケースに合わせて、適切な設定を行うことが重要です。

この記事は役に立ちましたか？