NCCL 2.27 ile Hızlı Çıkarım ve Dayanıklı Eğitim Nasıl Sağlanır?

Yapay zeka yükleri ölçeklendikçe, hızlı ve güvenilir GPU iletişimi sadece eğitim için değil, aynı zamanda büyük ölçekli tahminler için de kritik bir hale geliyor. NVIDIA Kolektif İletişim Kütüphanesi (NCCL), NVIDIA GPU’ları ve PCIe, NVLink, Ethernet (RoCE) ve InfiniBand (IB) gibi çeşitli bağlantıların optimize edildiği yüksek performanslı, topoloji bilgili toplu işlemler sunar: AllReduce, Broadcast, Reduce, AllGather ve ReduceScatter.

İletişim ve hesaplamanın tek çekirdekli uygulanması ile NCCL, düşük gecikmeli senkronizasyon sağlar. Bu, hem dağıtılmış eğitim hem de gerçek zamanlı tahmin senaryoları için idealdir. Geliştiriciler, NCCL’in dinamik topoloji algılama ve basit C tabanlı API sayesinde özel donanım yapılandırmaları için ayar yapmadan düğümler arasında ölçeklenebilirlik elde edebilirler.

Bu yazıda, NCCL 2.27 sürümünün en son özellikleri tanıtılmaktadır. Bu yeni özellikler, tahmin gecikmesini, eğitim direncini ve geliştirici gözlemlenebilirliğini artırmaktadır. Daha fazla bilgi ve başlangıç için, NVIDIA/nccl GitHub deposuna göz atabilirsiniz.

Yeni Performans Seviyeleri Açığa Çıkıyor

NCCL 2.27, GPU’lar arasında kolektif iletişimi iyileştiren önemli güncellemeler sunmakta; gecikmeyi, bant genişliği verimliliğini ve ölçeklenebilirlik sorunlarını ele almaktadır. Bu gelişmeler, modern yapay zeka altyapısının ihtiyaçlarıyla uyum sağlamaktadır. Artık ultra düşük gecikme süresi, gerçek zamanlı tahmin süreçleri için bir zorunluluk haline gelmiştir. Ayrıca, büyük ölçekli dağıtımların güvenilir bir şekilde çalışmasını sağlamak için sağlam hata toleransı gereklidir.

Ana sürüm öne çıkan başlıkları arasında düşük gecikmeli çekirdekler, simetrik bellek desteği, doğrudan NIC desteği ve NVLink ve InfiniBand SHARP desteği yer almaktadır.

Simetrik Bellek Desteği ile Düşük Gecikme

Bu sürüm, GPU’lar arasında aynı sanal adreslere sahip tamponların optimize edilmiş toplu işlemlerden yararlanmasına olanak tanıyan simetrik bellek desteği sunmaktadır. Bu çekirdekler, tüm mesaj boyutları için gecikmeyi önemli ölçüde azaltmaktadır; küçük mesaj boyutları için %7.6’ya kadar gecikme azaltımı sağlanmıştır.

Azaltımlar, FP32 biriktiriciler (ya da FP16 ile FP8 kullanarak NVLink Anahtarlı Sistemlerde) ile hesaplandığı için AllReduce, AllGather ve ReduceScatter işlemlerinde hem doğruluğu hem de belirleyiciliği artırmaktadır.

Simetrik bellek, tek bir NVLink alanı içinde NVLink iletişimini desteklemektedir. Bu, NVIDIA GB200 ve GB300 sistemlerinde NVL72 (72 GPU) ya da NVIDIA DGX ve HGX sistemlerinde NVL8 (8 GPU) şeklindedir. NVL8 alanlarında bile, geliştiriciler küçük ve orta mesaj boyutları için %2.5’e kadar daha yüksek performans elde edebilmektedir. Test rehberliği için NCCL-Test deposuna göz atabilirsiniz.

Doğrudan NIC Desteği

NCCL 2.27, GPU ölçekleme iletişiminde tam ağ bant genişliğini açığa çıkarmak için Doğrudan NIC yapılandırmalarını desteklemektedir. Seçilen NVIDIA Grace Blackwell platformlarında, CX8 NIC’leri ve NVIDIA Blackwell GPU’ları PCIe Gen6 x16 desteği sunarak 800 Gb/s ağ bant genişliği sunabilmektedir. Ancak Grace CPU’ları şu anda yalnızca PCIe Gen5 desteklediği için, throughput 400 Gb/s ile sınırlıdır.

Bunu ele almak için CX8 NIC, iki sanal PCIe ağacı sunmaktadır: Bir ağaçta, GPU PF üzerinden PCIe Gen6 x16 bağlantılarıyla doğrudan GPU PF’ye bağlanan NVIDIA CX8 NIC veri doğrudan fonksiyonu (PF) yer alırken, diğer ağaçta normal NIC PF CPU kök bağlantısına bağlıdır.

Bu yapılandırma, GPUDirect RDMA ve ilgili teknolojilerin, CPU ile GPU arasındaki bant genişliğini doyurmadan tam 800 Gb/s bant genişliği elde etmesini sağlamaktadır. Bu durum, birden fazla GPU bir CPU’yu paylaştığında özellikle önemlidir. Doğrudan NIC, yüksek hacimli tahmin ve eğitim yükleri için tam hızda ağ iletişimini mümkün kılmaktadır.

NVLink ve InfiniBand SHARP Desteği

NCCL 2.27, NVLink ve IB ağları için SHARP (Ölçeklenebilir Hiyerarşik Toplama ve Azaltma Protokolü) desteğini eklemektedir. Bu, hesaplama yoğun görevleri dışlamak için ağda azaltma işlemlerine olanak tanımaktadır. Bu yeni sürüm, NVLink Sharp ve IB Sharp kullanırken AllGather (AG) ve ReduceScatter (RS) kolektif işlemleri için SHARP desteği sunmaktadır.

Bu özellikle büyük ölçekli LLM eğitimleri için faydalıdır. AG ve RS artık AllReduce yerine tercih edildiği için hesaplama ve iletişim arasında daha iyi örtüşme sağlanmaktadır. Geleneksel halka tabanlı uygulamalar, 16 veya daha fazla SM tüketebilirken, NVLink ve IB SHARP ile bu talep 6 SM veya daha az seviyeye düşmektedir. Bu durum, model hesaplama için kaynakları serbest bırakarak genel eğitim verimliliğini artırmaktadır. Sonuç olarak, 1.000 GPU seviyesinde ve üzerindeki ölçeklenebilirlik ve performans artmaktadır.

Büyük Ölçekli Eğitimde Direnci Artırma ile NCCL Shrink

NCCL 2.27, dağıtılmış eğitimi daha dayanıklı, esnek ve verimli hale getirmek için Communicator Shrink özelliğini tanıtmaktadır. Yüzlerce veya binlerce GPU arasında çalışan eğitim görevleri, cihaz arızalarına karşı duyarlıdır. Communicator Shrink, eğitim sırasında başarısız veya gereksiz GPU’ların dinamik olarak hariç tutulmasına olanak tanımaktadır. Bu özellik iki işletim modu desteklemektedir:

• Varsayılan Mod: Planlı yeniden yapılandırmalar için kullanılır ve tüm işlemlerin tamamlanmasını sağlar.
• Hata Modu: Beklenmedik cihaz arızalarından kurtulmak için devam eden işlemleri otomatik olarak iptal eder.

NCCL Shrink, geliştiricilere:

• Eğitimde kesintiyi önlemek için dinamik olarak iletişimcileri yeniden yapılandırma imkanı sunar.
• Mümkünse kaynakları paylaşabilme yeteneği sağlar.
• Cihaz arızalarını minimum kesinti ile yönetme imkanı tanır.

NCCL Shrink özelliğinin planlı yeniden yapılandırma ve hata kurtarma senaryolarında örnek kullanımı şu şekildedir:

// Planlı yeniden yapılandırma: normal çalışırken bir rengi hariç tutma
NCCLCHECK(ncclGroupStart());
for (int i = 0; i < nGpus; i++) {
  if (i != excludedRank) {
    NCCLCHECK(ncclCommShrink(
      comm[i], &excludeRank, 1,
      &newcomm[i], NULL, NCCL_SHRINK_DEFAULT));
  }
}
NCCLCHECK(ncclGroupEnd());
// Hata kurtarma: bir cihaz arızasından sonra bir rengi hariç tutma
NCCLCHECK(ncclGroupStart());
for (int i = 0; i < nGpus; i++) {
  if (i != excludedRank) {
    NCCLCHECK(ncclCommShrink(
      comm[i], &excludeRank, 1,
      &newcomm[i], NULL, NCCL_SHRINK_ABORT));
  }
}
NCCLCHECK(ncclGroupEnd());

Geliştiricilere Yönelik Ek Özellikler

Bu sürümde geliştiricilere yönelik ek özellikler arasında simetrik bellek API’leri ve geliştirilmiş profil oluşturma yer almaktadır.

Simetrik Bellek API’leri

Simetrik bellek, NCCL 2.27’nin temel bir yeteneğidir ve yüksek performanslı, düşük gecikmeli kolektif işlemlere olanak tanır. Tüm sıralarda aynı sanal adreslerde tahsis edilen bellek tamponları, NCCL’in senkronizasyon aşamalarını azaltarak ve bant genişliği verimliliğini artırarak optimize edilmiş çekirdekler gerçekleştirmesine imkân tanır.

Bu desteği sağlamak amacıyla, kolektif simetrik bellek kaydı için bir pencere API’si sunulmaktadır:

ncclCommWindowRegister(ncclComm_t comm, void* buff, size_t size, 
ncclWindow_t* win, int winFlags);

ncclCommWindowDeregister(ncclComm_t comm, ncclWindow_t win);

• ncclCommWindowRegister, kullanıcı tarafından tahsis edilen belleği NCCL iletişimcisi ile kaydeder. Bellek, CUDA Sanal Bellek Yönetimi (VMM) API kullanılarak tahsis edilmelidir.
• winFlags içinde NCCL_WIN_COLL_SYMMETRIC bayrağının yer alması gerekmektedir; bu simetrik çekirdek optimizasyonlarını etkinleştirir.
• Tüm sıraların eşleşen ofsetlere sahip tamponlar sağlaması gerekmektedir; bu simetrik adresleme için önemlidir.
• Kayıt işlemleri, ncclCommWindowDeregister, yerel bir işlem olup yalnızca tüm ilgili kolektif işlemler tamamlandıktan sonra yapılmalıdır.

ncclCommWindowRegister kolektif ve bloklama anlamına gelir; bu nedenle, birden fazla GPU tek bir iş parçacığı tarafından yönetiliyorsa, ncclGroupStart ve ncclGroupEnd içerisinde yer almalıdır.

Eğer simetrik bellek istenmiyorsa, kullanıcılar bu özelliği NCCL_WIN_ENABLE=0 olarak ayarlayarak tamamen devre dışı bırakabilir.

Şekil 4, NCCL pencere API’si üzerinden simetrik bellek kaydının nasıl yapıldığını göstermektedir. Sanal adreslerin hizalanması ile NCCL, toplu işlemler için optimize edilmiş düşük gecikmeli çekirdeklerin gerçekleştirilmesini sağlamaktadır.

Geliştirilmiş Profil Oluşturma

NCCL 2.27, geliştiricilere ve araçlara iletişim performansını tanımlamak için daha doğru ve verimli bir enstrümantasyon sunan bir dizi geliştirme ile birlikte gelmektedir.

Proxy Olaylarının Uyumu

Daha önce, NCCL hem ncclProfileProxyOp hem de ncclProfileProxyStep olaylarını, ağ proxy iş parçacığının ilerlemesini izlemede kullanıyordu. Bu olaylar farklı ayrıntı seviyeleri sağlarken, aynı zamanda birçok enstrümantasyon noktasını da çoğaltıyordu. 2.27 sürümünde, NCCL, gereksiz ProxyOp durumlarını kaldırarak bu modeli sadeleştirmiştir. Ayrıca, ncclProfilerProxyOpInProgress_v4 durumu, iletişim ilerlemesini izlerken netlik sağlamakta ve çoğaltmayı azaltmaktadır.

Buna ek olarak, yeni bir ProxyStep olayı durumu olarak ncclProfilerProxyStepPeerWait_v4 tanıtılmıştır. Bu, bir gönderici sırasının alıcıdan bir “göndermeye hazır” sinyali için bekleme süresini yansıtır ve önceki işlevselliği entegre ederek çoğaltmayı en aza indirger.

GPU Çekirdek Olayı Doğruluğu

Zamanlama doğruluğunu artırmak için NCCL artık yerel GPU zaman damgası iletimini desteklemektedir. Bu, host tarafı olay zamanlamasına dayanmak yerine, GPU’nun kendi iç global zamanlayıcısını kullanarak başlangıç ve bitiş zaman damgalarını kaydetmesine olanak tanır. Bu, profil araçlarının GPU’dan doğrudan doğru kernel çalışma süreleri elde etmesini sağlamaktadır. Ancak, zaman damgalarını CPU zamanına dönüştüren geliştiricilerin kalibrasyon veya enterpolasyon uygulamaları gerekebilir.

Ağ Eklentisi Olay Güncellemeleri

NCCL profil arayüzü artık recordEventState mekanizmasını ağ tanımlı olaylar için desteklemektedir. Bu yeni mekanizma, profilin devam eden işlemlerin durumunu güncellemesini sağlar. Bu, yeniden iletim sinyalleri veya yoğunluk ipuçları gibi gerçek zamanlı ağ geri bildirimlerini performans zaman çizelgelerine entegre etmek için faydalıdır.

Diğer Geliştirmeler

• Profil oluşturma başlangıcı: NCCL şimdi, profil oluşturma başlangıcında iletişimci meta verilerini rapor etmektedir, bu bilgiler ad, kimlik, düğüm sayısı, sıra sayısı ve hata ayıklama seviyesini içermektedir.
• Kanal raporlaması: Rapor edilen kanal sayısı, teorik limitlerden ziyade gerçek kullanımı yansıtmaktadır. Bu, nokta-nokta (P2P) işlemleri de kapsamaktadır.
• İletişimci etiketleme: ncclConfig_t yapısı, profil işlemleri ile belirli iletişimciler arasındaki korelasyonu geliştirmek amacıyla iletişimci adlarını içerecek şekilde genişletilmiştir.

Tüm bu güncellemeler, NCCL profilci eklenti arayüzünün detaylarını artırmakta ve geliştiricilere büyük ölçekli yapay zeka iş yüklerini tanıma ve ayarlama için gerekli bilgileri sunmaktadır.

Profil eklentisiyle ilgili daha fazla bilgiye NCCL Profiler Eklentisi üzerinden ulaşabilirsiniz.

Geleceğe Yönelik Destek

Geleceğe yönelik destek şunları içermektedir:

• Veri merkezi arası iletişim: Coğrafi olarak dağıtılmış veri merkezleri arasında kolektif işlemleri erken destekleme.
• Çoklu NIC eklentisi görünürlüğü: Birden fazla ağ yapılandırmasını aynı anda kullanma yeteneği sağlar.

NCCL 2.27 ile Başlayın

NCCL 2.27’nin yeni yeteneklerini keşfedin ve daha düşük gecikme, daha fazla hataya dayanıklılık ve daha derin izlenebilirlik ile dağıtılmış tahmin ve eğitim iş akışlarınızı geliştirin.

Detaylı belgeler, kaynak kodu ve ihtiyacınız olan destek için NVIDIA/nccl GitHub deposunu ziyaret edin. NCCL’in mimarinize nasıl uyum sağlayacağı hakkında daha fazla bilgi için NCCL dokümantasyonuna göz atabilirsiniz.