NVIDIA CUDA எதிர் OpenAI Triton: AI க்கான GPU நிரலாக்கத்தின் நவீன ஒப்பீடு (NVIDIA CUDA vs OpenAI Triton: A Modern Comparison of GPU Programming for AI)

அயன்களைப் பொறியில் அடைக்கும் குவாண்டம் கணினி ஆய்வுக் கருவியின் விளக்கப் படம்; GPU அல்லது NVIDIA அமைப்பு அல்ல.
நிரலாக்க மாதிரிகள் (programming models), செயல்திறன் (performance) மற்றும் நடைமுறைப் பயன்பாடு ஆகியவற்றின் ஆய்வு அடிப்படையிலான மதிப்பாய்வு
ஆய்வு சரிபார்க்கப்பட்ட திகதி: 2026 நவம்பர் 15
சுருக்கம் (Abstract)
NVIDIA CUDA மற்றும் Triton ஆகியவை GPU நிரலாக்கத்துக்கான (GPU programming) போட்டி அணுகுமுறைகள் என்று அடிக்கடி விவரிக்கப்படுகின்றன. CUDA குறைந்த மட்டக் கட்டுப்பாட்டை (low-level control) வழங்குகிறது; Triton, AI கேர்னல்களை (kernels) எழுதுவதற்கு Python அடிப்படையிலான, இலகுவாக அணுகக்கூடிய வழியை வழங்குகிறது. இந்த விவரிப்பு அவற்றின் வரலாற்றின் ஒரு பகுதியைச் சொல்கிறது; ஆனால் இன்று அது போதுமானதல்ல.
CUDA என்பது நிரலாக்க மாதிரிகள், நூலகங்கள் (libraries), நிரல்மாற்றிகள் (compilers) மற்றும் இயக்கநேர வசதிகளை (runtime facilities) உள்ளடக்கிய பரந்த கணினித் தளம் ஆகும். Triton என்பது திறந்த மூலமான (open-source) மொழியும் நிரல்மாற்றியும் ஆகும்; அது திறமையான தனிப்பயன் எண்சார் கேர்னல்களில் (custom numerical kernels) கவனம் செலுத்துகிறது. நவீன CUDA, CUDA Tile மூலம் டைல் அடிப்படையிலான நிரலாக்கத்தையும் (tile-based programming) உள்ளடக்குகிறது; NVIDIA இன் CuTe DSL, வன்பொருளுக்குச் சிறப்பாக அமைந்த கேர்னல்களுக்கு (hardware-specialised kernels) இன்னொரு Python அடிப்படையிலான அணுகுமுறையை வழங்குகிறது.
இந்தக் கட்டுரை அவற்றின் அருவமாக்கல்கள் (abstractions), சூழல்தொகுதிகள் (ecosystems), வன்பொருள் ஆதரவு மற்றும் பொறியியல் சமரசங்களை (engineering trade-offs) ஒப்பிடுகிறது. ஒரே ஒரு வெற்றியாளர் உண்டு என்று கூறாமல், செயல்திறனை மதிப்பிடுவதற்கான ஒரு வழிமுறையை முன்வைக்கிறது.
முக்கிய சொற்கள் (Keywords): CUDA, Triton, GPU programming, CUDA Tile, CuTe DSL, kernel fusion, AI infrastructure.
1. இந்த ஒப்பீடு ஏன் முக்கியம்
AI செயல்திறன் என்பது செயலிகளின் எண்ணிக்கையையோ மாதிரியின் அளவையோ மட்டும் சார்ந்தது அல்ல. சக்திவாய்ந்த GPU கூட, தரவு நகர்வதற்காகக் காத்திருப்பதிலும், சிறிய கேர்னல்களை (kernels) ஏவுவதிலும், வேலைகளை ஒத்திசைப்பதிலும் (synchronising) அல்லது திறனற்ற நினைவக அமைப்பைச் (memory layout) செயலாக்குவதிலும் கணிசமான நேரத்தைச் செலவிடலாம்.
நிரலாக்க அணுகுமுறை, செயலாக்க வேகத்தை மட்டுமன்றி, உருவாக்க நேரம், கொண்டுசெல்லக்கூடிய தன்மை (portability), வழுநீக்கம் (debugging) மற்றும் நீண்டகாலப் பராமரிப்பையும் பாதிக்கிறது. எனவே பயனுள்ள கேள்வி “எந்த மொழி வேகமானது?” என்பது மட்டுமல்ல. அது:
இந்த வேலைச்சுமைக்கும் (workload), வன்பொருளுக்கும், குழுவுக்கும் சரியான, பராமரிக்கக்கூடிய, திறமையான செயலாக்கத்தை எந்த அணுகுமுறை தருகிறது?
2. CUDA மற்றும் Triton சமமான வகைகள் அல்ல
CUDA என்பது NVIDIA இன் இணைச் செயலாக்கக் கணினித் தளமும் (parallel computing platform) நிரலாக்க மாதிரியும் ஆகும். GPU செயலாக்கம், நினைவக மேலாண்மை, சாதனக் கட்டுப்பாடு, மற்றும் நூலகங்கள், கருவிகளின் பரந்த தொகுப்பு ஆகியவை அதன் எல்லைக்குள் அடங்கும். CUDA C++ என்பது அந்தத் தளத்தினுள் கேர்னல்களை எழுதும் ஒரு வழி மட்டுமே. NVIDIA CUDA Programming Guide
Triton என்பது, குறிப்பாக இயந்திரக் கற்றலுக்கு (machine learning), திறமையான எண்சார் கேர்னல்களை எழுதுவதற்கான திறந்த மூல மொழியும் நிரல்மாற்றியும் ஆகும். அதன் ஆய்வு அடித்தளங்கள், 2021 இல் OpenAI அறிவித்த Triton 1.0 க்கு முந்தியவை; Triton இன் அசல் உருவாக்குநரின் பணியின் மீதே OpenAI கட்டியெழுப்பியது. இது மூடிய OpenAI சேவை அல்ல. OpenAI’s original Triton announcement
நியாயமான பகுப்பாய்வு நான்கு ஒப்பீடுகளைப் பிரித்துப் பார்க்கிறது:
1. தனிப்பயன் கேர்னல்களை வெளிப்படுத்தும் வழிகளாக CUDA C++ எதிர் Triton.
2. டைல் சார்ந்த நிரலாக்க அணுகுமுறைகளாக CUDA Tile எதிர் Triton.
3. பரந்த CUDA சூழல்தொகுதி (ecosystem) எதிர் Triton இன் குறுகிய மொழி மற்றும் நிரல்மாற்றிச் சூழல்தொகுதி.
4. குறிப்பிட்ட வேலைச்சுமையில் அளவிடப்பட்ட குறிப்பிட்ட செயலாக்கங்கள் (implementations).
பெயர் குறிப்பு: நிரலாக்க மொழியான Triton என்பது NVIDIA Triton Inference Server இலிருந்து வேறுபட்டது.
3. CUDA C++: வெளிப்படையான இணைச் செயலாக்கம்
வழமையான CUDA C++ பொதுவாக ஒற்றை அறிவுறுத்தல், பல இழைகள் (single-instruction, multiple-thread) என்னும் SIMT மாதிரியைப் பயன்படுத்துகிறது. ஒரு கேர்னல் (kernel), இழைத் தொகுதிகளின் (thread blocks) கட்டமைப்பு (grid) முழுவதிலும் ஏவப்படுகிறது; ஒவ்வோர் இழையும் (thread) தான் செயலாக்க வேண்டிய தரவு எது என்பதைக் கணக்கிட முடியும்.
எடுத்துக்காட்டாக:
__global__ void add_cuda(
const float* a, const float* b,
float* output, int n)
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
if (i < n) {
output[i] = a[i] + b[i];
}
}நேர்மறை நீளத்துக்கான ஏவலில், ஒரு தொகுதிக்கு 256 இழைகளையும், உள்ளீடு முழுவதையும் மூடப் போதுமான தொகுதிகளையும் பயன்படுத்தலாம். தெளிவுக்காக, நினைவக ஒதுக்கீடு (allocation), வழுச் சோதனை (error checking) மற்றும் ஒத்திசைவு (synchronisation) ஆகியவை இங்கே விடப்பட்டுள்ளன.
இழைகள் எவ்வாறு ஒத்துழைக்கின்றன, நினைவகத்தை எவ்வாறு பயன்படுத்துகின்றன, எவ்வாறு ஒத்திசைகின்றன என்பதன் மீது CUDA C++ உருவாக்குநர்களுக்குக் கட்டுப்பாட்டை வழங்குகிறது. சிறப்புச் செயலாக்க உத்திகள் தேவைப்படும் வழிமுறைகளுக்கு இது பயனுள்ளது; ஆனால் உருவாக்குநர் மேலும் பல செயலாக்க விவரங்களையும் நிர்வகிக்க வேண்டியுள்ளது. CUDA programming model
4. Triton: கணிப்பைத் தொகுதிகளாக வெளிப்படுத்தல்
Triton, Python அடிப்படையிலான தொடரியலையும் (syntax) ஒரு சிறப்பு மொழியையும் பயன்படுத்துகிறது. ஒவ்வொரு உறுப்பையும் ஒரு வன்பொருள் இழைக்கு வெளிப்படையாக ஒதுக்குவதற்குப் பதிலாக, நிரலாளர் தரவுத் தொகுதி (block) மீதான செயல்பாடுகளை விவரிக்கிறார். Triton இன் நிரல்மாற்றி அந்தச் செயல்பாடுகளை GPU வளங்களுடன் பொருத்துகிறது.
import triton
import triton.language as tl
@triton.jit
def add_triton(a, b, output, n, BLOCK: tl.constexpr):
offsets = tl.program_id(0) * BLOCK + tl.arange(0, BLOCK)
valid = offsets < n
x = tl.load(a + offsets, mask=valid, other=0.0)
y = tl.load(b + offsets, mask=valid, other=0.0)
tl.store(output + offsets, x + y, mask=valid)Python தொடரியல் இருப்பதால், வழமையான விளக்கப்படும் Python (interpreted Python) ஒவ்வொரு கணிப்பையும் GPU இல் செய்கிறது என்று பொருள் அல்ல. கேர்னல் நிரல்மாற்றம் (compile) செய்யப்படுகிறது. மேலும், BLOCK என்பது ஒரு நிரல் நிகழ்வுக்கு (program instance) ஒதுக்கப்பட்ட உறுப்புகளை விவரிக்கிறது; அது வெறுமனே வேறு பெயரில் உள்ள CUDA இழை எண்ணிக்கை அல்ல. Triton vector-addition tutorial
இந்த எடுத்துக்காட்டுகள் நிரலாக்க மாதிரிகளை விளக்குகின்றன. அவை இந்தக் கட்டுரைக்காக இயக்கிப் பார்க்கப்படவில்லை, செயல்திறன் அளவீடும் (benchmark) செய்யப்படவில்லை.
5. நவீன CUDA டைல் அடிப்படையிலான நிரலாக்கத்தையும் ஆதரிக்கிறது
“CUDA இழை அடிப்படையிலானது, Triton டைல் அடிப்படையிலானது” என்ற கூற்று இப்போது முழுமையற்றது.
NVIDIA இன் தற்போதைய CUDA ஆவணங்கள், Python இடைமுகமும் CUDA Tile C++ உம் உள்ளடங்கிய CUDA Tile ஐ விவரிக்கின்றன. டைல் செயல்பாடுகள், தரவுத் தொகுதிகள் மீதான வேலையை வெளிப்படுத்த நிரலாளர்களை அனுமதிக்கின்றன; நிரல்மாற்றி அந்த வேலையை வன்பொருள் இழைகளுடன் பொருத்துகிறது. அதே பயன்பாட்டில் டைல் கேர்னல்களையும் SIMT கேர்னல்களையும் ஒருங்கே பயன்படுத்துவதையும் ஆவணங்கள் விவரிக்கின்றன. CUDA Tile C++ ஆனது CUDA Toolkit 13.3 முதல் ஆவணப்படுத்தப்பட்டுள்ளது. CUDA Tile programming guide
ஒப்பீட்டில் இது ஒரு முக்கிய மாற்றம். டைல் அடிப்படையிலான நிரலாக்கம் இனி Triton க்கு மட்டுமே உரிய தனித்துவம் அல்ல; நவீன GPU வன்பொருளின் சிக்கலுக்கான பரந்த பதிலாக அது உள்ளது.
cuTile என்பது CuTe DSL அல்ல
ஒத்த பெயர்களைக் கொண்ட இந்தத் தொழில்நுட்பங்களைக் குழப்பிக்கொள்ளக்கூடாது:
- - cuTile என்பது CUDA Tile இன் Python இடைமுகம்.
- - CuTe DSL என்பது NVIDIA இன் CUTLASS சூழல்தொகுதியுடன் தொடர்புடைய, அமைப்பு அறிந்த (layout-aware) மற்றும் வன்பொருளுக்குச் சிறப்பாக அமைந்த கேர்னல்களுக்காக வடிவமைக்கப்பட்ட Python அடிப்படையிலான மொழி.
இரண்டும் Python தொடரியலைப் பயன்படுத்துகின்றன; ஆனால் அவற்றின் நிரலாக்க மாதிரிகளும் நோக்கமும் வேறுபட்டவை. உருவாக்குநருக்கு வன்பொருள் மீது எவ்வளவு கட்டுப்பாடு உண்டு என்பதை Python தொடரியல் மட்டும் தீர்மானிப்பதில்லை. NVIDIA CuTe DSL documentation
6. Triton எதைத் தானியக்கமாக்குகிறது, பொறியாளர்கள் இன்னும் எதைத் தீர்மானிக்க வேண்டும்
டென்சர் (tensor) சார்ந்த கணிப்புகளுக்குத் தேவையான வெளிப்படையான இழை ஒருங்கிணைப்பின் (thread coordination) அளவை Triton குறைக்க முடியும். ஆனால் அது பொறியியல் முடிவுகளை நீக்குவதில்லை.
உருவாக்குநர் இன்னும் பின்வருவனவற்றைக் கருத்தில் கொள்ள வேண்டும்:
- - டைல் அளவுகளும் நிரல் அமைப்பும் (program organisation).
- - தொடர்ச்சியான எதிர் இடைவெளியுடைய (contiguous versus strided) நினைவக அணுகல்.
- - தரவு மீள்பயன்பாடும் தற்காலிக நினைவக (cache) நடத்தையும்.
- - பதிவகச் சுமையும் (register pressure) வள வரம்புகளும்.
- - உள்ளீட்டு வடிவங்களும் (input shapes) எல்லை கையாளலும் (boundary handling).
- - எண்சார் துல்லியமும் (numerical precision) ஏற்கத்தக்க பிழையும்.
Triton இன் சொந்த matrix-multiplication பயிற்சி, டைல் பரிமாணங்கள், தற்காலிக நினைவகத்தை அறிந்த ஒழுங்கு (cache-aware ordering) மற்றும் கட்டமைப்பு இசைவாக்கம் (configuration tuning) பற்றி விவாதிக்கிறது. எனவே செயல்திறன் என்பது தானாகக் கிடைக்கும் உத்தரவாதம் அல்ல; அது வடிவமைப்பும் அளவீடும் சார்ந்த பிரச்சினை. Triton matrix-multiplication tutorial
குறிப்பாக, Triton எப்போதும் நிபுணர் எழுதிய CUDA க்கு நிகராக இருக்கும், எப்போதும் உச்ச அலைக்கற்றையை (peak bandwidth) எட்டும், அல்லது ஒவ்வொரு செயல்பாட்டுக்கும் தானாகவே Tensor Cores ஐப் பயன்படுத்தும் என்று சொல்வது தவறாக வழிநடத்தும். எடுத்துக்காட்டாக, உறுப்பு வாரியான திசையன் கூட்டல் (elementwise vector addition), Triton இல் எழுதப்பட்டது என்ற காரணத்தால் மட்டும் matrix multiplication ஆகிவிடாது.
7. வன்பொருள் ஆதரவும் கொண்டுசெல்லக்கூடிய தன்மையும் (portability)
இந்த மதிப்பாய்வின் நேரத்தில், upstream Triton களஞ்சியம் (repository) Linux, compute capability 8.0 அல்லது அதற்கு மேற்பட்ட NVIDIA GPU கள், மற்றும் ROCm 6.2 அல்லது அதற்கு மேற்பட்ட AMD GPU கள் ஆகியவற்றைப் பட்டியலிடுகிறது. CPU ஆதரவு உருவாக்கத்தில் உள்ளது என்று அது குறிப்பிடுகிறது. ஆதரவு மாறக்கூடும்; பட்டியலிடப்பட்ட சாதனம் என்பதால் ஒவ்வொரு செயல்பாட்டுக்கும் தரவு வகைக்கும் ஒரே ஆதரவோ செயல்திறனோ உறுதி என்று பொருள் அல்ல. Triton compatibility information
கொண்டுசெல்லக்கூடிய தன்மையை மூன்று வகைகளாகப் பிரித்துப் பார்ப்பது உதவும்:
1. மூல நிலைக் கொண்டுசெல்லல் (Source portability): ஒத்த நிரல் இன்னொரு இலக்கில் நிரல்மாற்றம் ஆகக்கூடும்.
2. செயல்பாட்டு நிலைக் கொண்டுசெல்லல் (Functional portability): எதிர்பார்த்த முடிவைத் தருகிறது.
3. செயல்திறன் நிலைக் கொண்டுசெல்லல் (Performance portability): அந்த இலக்கிலும் திறமையாக இருக்கிறது.
இரண்டு GPU கட்டமைப்புகளில் ஓடக்கூடிய ஒரு கேர்னலுக்கு, ஒவ்வொன்றிலும் நன்றாகச் செயல்பட வெவ்வேறு டைல் அளவுகளோ இசைவாக்கமோ (tuning) தேவைப்படக்கூடும்.
CUDA அடிப்படையில் NVIDIA இன் தளம். வன்பொருள் விவரங்களை CUDA Tile அருவமாக்குவதை, அது ஒவ்வொரு தயாரிப்பாளரின் GPU களையும் இலக்காகக் கொள்கிறது என்ற வாக்குறுதியாகக் கருதக்கூடாது.
8. பரந்த CUDA சூழல்தொகுதி (ecosystem)
கையால் எழுதிய கேர்னல்களுக்கு அப்பாலும் CUDA இன் நடைமுறைப் பெறுமதி விரிந்தது.
- - cuBLAS மற்றும் cuBLASLt ஆகியவை நிலைபெற்ற நேரியல் இயற்கணித (linear-algebra) மற்றும் கட்டமைக்கக்கூடிய matrix-multiplication செயலாக்கங்களை வழங்குகின்றன. தனிப்பயன் நிரலை மதிப்பிடும்போது அவை முக்கிய ஒப்பீட்டு அடிப்படைகள் (baselines). NVIDIA cuBLAS documentation
- - cuDNN மேம்படுத்தப்பட்ட ஆழ்கற்றல் (deep-learning) செயல்பாடுகளை வழங்குகிறது; ஆதரவு செயல்பாடு, மென்பொருள் பதிப்புகள் மற்றும் வன்பொருளைப் பொறுத்தது. NVIDIA cuDNN overview
- - NCCL, AllReduce, AllGather மற்றும் ReduceScatter உள்ளிட்ட பல GPU தொடர்பாடல் செயல்பாடுகளை வழங்குகிறது. வேகமான உள்ளூர் கேர்னல், பகிரப்பட்ட தொடர்பாடல் அடுக்குக்கு (distributed communication layer) மாற்றாகாது. NCCL documentation
தேர்ந்தெடுத்த கேர்னல்களுக்கு Triton ஐப் பயன்படுத்துவது, NVIDIA நூலகங்களையோ CUDA தளத்தையோ கைவிடுவதைக் குறிக்க வேண்டியதில்லை. ஒரு பயன்பாடு, கட்டமைப்பால் (framework) உருவாக்கப்பட்ட கேர்னல்கள், வழங்குநர் நூலகங்கள் (vendor libraries), தனிப்பயன் Triton அல்லது CUDA கூறுகளை ஒன்றிணைக்கலாம்.
9. Triton மற்றும் PyTorch
Triton, கேர்னல்களை நேரடியாக எழுதும் உருவாக்குநர்களுக்கும், கட்டமைப்பு நிரல்மாற்றத்துக்கும் (framework compilation) தொடர்புடையது. torch.library.triton_op உள்ளிட்ட வழிகளில், பயனர் வரையறுத்த Triton செயல்பாடுகளை torch.compile உடன் ஒருங்கிணைக்கும் வழிகளை PyTorch ஆவணப்படுத்தியுள்ளது. PyTorch’s user-defined Triton tutorial
இருப்பினும், உற்பத்திச் சூழல் ஒருங்கிணைப்புக்கு, டென்சர்களை உள்ளீடாக ஏற்பதை விட மேலானவை தேவை. ஆதரிக்கப்படும் சாதனங்கள், அமைப்புகள் (layouts), வடிவ மாற்றங்கள், மீள்நிரல்மாற்றம் (recompilation), செயல்பாட்டுப் பதிவு (operator registration), எண்சார் சகிப்பு வரம்புகள் (numerical tolerances) மற்றும் வழங்கல் தேவைகளை (deployment requirements) ஒரு குழு ஆராய வேண்டும்.
பயிற்சிக்கு (training), தனிப்பயன் forward கேர்னல் தானாகவே முழுமையான செயல்பாடு (operator) ஆகிவிடாது; சாய்வு நடத்தையும் (gradient behaviour) செயலாக்கப்பட்டோ ஆதரிக்கப்பட்டோ சோதிக்கப்பட வேண்டும்.
10. செயல்திறனைத் தீர்மானிப்பது எது?
பயனுள்ள கருத்து மாதிரி:
அடையக்கூடிய செயல்திறன் (Attainable throughput) ≲ பின்வருவனவற்றில் சிறியது:
- - செயலியின் கணிப்பு உச்சவரம்பு (compute ceiling); மற்றும்
- - நினைவக அலைக்கற்றை (memory bandwidth) பெருக்கல் கணிதச் செறிவு (arithmetic intensity).
கணிதச் செறிவு (Arithmetic intensity) என்பது, மாற்றப்படும் தரவுடன் ஒப்பிடும்போது செய்யப்படும் பயனுள்ள கணிதத்தின் அளவு. ஒரு வேலைச்சுமை கணிப்பாலா, நினைவகத்தாலா அல்லது வேறு காரணிகளாலா அதிகம் வரையறுக்கப்படுகிறது என்பதைக் கண்டறிய NVIDIA Nsight Compute போன்ற சுயவிவரக் கருவிகள் (profiling tools) உதவும். Nsight Compute profiling guide
எளிய float32 திசையன் கூட்டலில், ஒவ்வோர் உறுப்புக்கும் ஏறத்தாழ இரண்டு நான்கு பைட் வாசிப்புகளும், ஒரு நான்கு பைட் எழுதலும், ஒரு கூட்டலும் உள்ளன. தற்காலிக நினைவக விளைவுகளைப் புறக்கணித்தால், மாற்றப்படும் ஒவ்வொரு 12 பைட்டுக்கும் ஏறத்தாழ ஒரு செயல்பாடு. எனவே இது முக்கியமாகத் தரவு நகர்வு மற்றும் மேலதிகச் சுமைக்கான (overhead) சோதனை; Tensor Core matrix-multiplication செயல்திறனுக்குப் பயனுள்ள விளக்கம் அல்ல.
Matrix multiplication, attention, reductions மற்றும் ஒழுங்கற்ற சுட்டுதல் (irregular indexing) ஆகியவற்றுக்கு வெவ்வேறு கட்டுப்பாடுகள் உள்ளன. ஒரு செயல்பாட்டின் முடிவு, அனைவருக்கும் பொருந்தும் செயல்திறன் தரவரிசையை நிறுவ முடியாது.
11. Kernel fusion மற்றும் FlashAttention
Kernel fusion, இடைநிலைத் தரவு மாற்றங்களையும் ஏவல் மேலதிகச் சுமையையும் குறைக்கச் செயல்பாடுகளை ஒன்றிணைக்கிறது. இதை CUDA சார்ந்த அல்லது Triton சார்ந்த செயலாக்கங்கள் இரண்டிலும் பயன்படுத்தலாம்.
FlashAttention என்பது IO-aware வழிமுறைக்கு ஒரு முக்கிய எடுத்துக்காட்டு. அது attention கணிப்பை டைல்களாகப் பிரிக்கிறது; GPU உயர் அலைக்கற்றை நினைவகத்தில் (high-bandwidth memory) முழு attention அணியையும் உருவாக்குவதைத் தவிர்க்கிறது. அடிப்படை உத்தி ஒரு மொழிக்கு மட்டும் உரியதல்ல. அது இடைநிலைச் சேமிப்பையும் நினைவகப் போக்குவரத்தையும் குறைக்கிறது; ஆனால் துல்லியமான அடர் attention இன் இருபடிக் கணிதத்தை (quadratic arithmetic) நீக்குவதில்லை. Original FlashAttention paper
Triton ஒரு fused-attention பயிற்சியை வழங்குகிறது. அதிகாரபூர்வ FlashAttention திட்டமும் CuTe DSL உள்ளிட்ட பிற தொழில்நுட்பங்களைப் பயன்படுத்தும் செயலாக்கங்களை ஆவணப்படுத்துகிறது. FlashAttention repository
நியாயமான ஒப்பீட்டுக்கு இது முக்கியம்: fused செய்யப்பட்ட Triton செயலாக்கத்தை, வேண்டுமென்றே fuse செய்யப்படாத CUDA செயலாக்கத்துடன் ஒப்பிடுவது, மொழி வேறுபாட்டுடன் வழிமுறை வேறுபாட்டையும் அளக்கிறது.
12. நியாயமான செயல்திறன் அளவீட்டை (benchmark) எவ்வாறு நடத்துவது
பின்வருவது முன்மொழியப்பட்ட ஆய்வு வழிமுறை; இந்தக் கட்டுரைக்காக நடத்தப்பட்ட பரிசோதனை அல்ல.
பிரதிநிதித்துவ வேலைச்சுமைகளைத் தேர்ந்தெடுத்தல்
உறுப்பு வாரிச் செயல்பாடுகள் (elementwise operations), reductions, normalisation, matrix multiplication, attention மற்றும் நோக்கம் கொண்ட பயன்பாட்டுக்குப் பிரதிநிதித்துவமான குறைந்தது ஒரு செயல்பாட்டைச் சோதிக்கவும். யதார்த்தமான உள்ளீட்டு வடிவங்களையும் எல்லை நிலைகளையும் (boundary cases) சேர்க்கவும்.
முதலில் சரியான தன்மையைச் சரிபார்த்தல்
வெளியீடுகளைப் பொருத்தமான மேற்கோளுடன் (reference) ஒப்பிடவும். டைல் அளவின் மடங்கல்லாத அளவுகள், தொடர்புடைய அமைப்புகள் மற்றும் எண்சார் துல்லியத்தைச் சோதிக்கவும். பயிற்சிக்கு, forward முடிவுகளுடன் சாய்வுகளையும் (gradients) ஒப்பிடவும்.
தொடக்க நிலையைச் சீரான நிலையிலிருந்து பிரித்தல்
நிரல்மாற்றத்தையும் autotuning ஐயும், சூடேற்றப்பட்ட செயலாக்கத்திலிருந்து (warmed execution) தனியாகத் தெரிவிக்கவும். குறிப்பாக ஒரு சேவை புதிய வடிவங்களை அடிக்கடி எதிர்கொள்ளும்போது, உற்பத்திச் சூழலில் இரண்டும் முக்கியமாகலாம்.
நம்பகமான ஒப்பீட்டு அடிப்படைகளை ஒப்பிடுதல் (Compare credible baselines)
எளிமையான கையால் எழுதிய கேர்னலை மட்டுமன்றி, தொடர்புடைய கட்டமைப்பு (framework) மற்றும் வழங்குநர் நூலகச் செயலாக்கங்களையும் சேர்க்கவும். GPU மாதிரி, மென்பொருள் பதிப்புகள், துல்லியம், கேர்னல் கட்டமைப்பு மற்றும் அளவீட்டு முறையைப் பதிவு செய்யவும்.
முழுப் பயன்பாட்டையும் அளத்தல்
வேகமான கேர்னல் என்பதால் முழுமையான (end-to-end) செயல்திறனில் அதே அளவு முன்னேற்றம் கிடைக்கும் என்று கூற முடியாது. எடுத்துக்காட்டாக, ஒரு கூறு முதலில் இயக்க நேரத்தில் 30% ஐ எடுத்து, அது இரு மடங்கு வேகமாக்கப்பட்டால், மொத்த வேக அதிகரிப்பு:
1 ÷ (0.7 + 0.3 ÷ 2) ≈ 1.18, அல்லது ஏறத்தாழ 18%.
மீதமுள்ள 70% மாறாமல் இருக்கும். அளவிடப்பட்ட தடைப்புள்ளிகளைச் (bottlenecks) சுயவிவரப்படுத்தி, படிப்படியாக மேம்படுத்தவும் NVIDIA அவ்வாறே பரிந்துரைக்கிறது. CUDA Best Practices Guide
13. வழுநீக்கமும் (debugging) பராமரிப்பும்
உயர் மட்டத் தொடரியல், தவறான masks, செல்லாத நினைவக அணுகல் அல்லது எண்சார் பிழைகளைத் தடுப்பதில்லை. assertions, interpreter மற்றும் NVIDIA Compute Sanitizer பயன்பாடு உள்ளிட்ட வழுநீக்க வசதிகளை Triton ஆவணப்படுத்துகிறது; ஒவ்வொன்றுக்கும் வரம்புகள் உள்ளன, நிரல்மாற்றம் செய்யப்பட்ட GPU செயலாக்கத்துக்கும் இன்னும் சோதனை தேவை. Triton debugging documentation
எந்த அணுகுமுறையிலும், உற்பத்திச் சூழல் கேர்னலுக்குப் பின்வருவன இருக்க வேண்டும்:
- - தெளிவான விவரக்குறிப்பும் மேற்கோள் செயலாக்கமும் (reference implementation).
- - பிரதிநிதித்துவ வடிவங்கள் முழுவதிலும் சரியான தன்மைக்கான, பின்னடைவுக்கான (regression) சோதனைகள்.
- - ஆவணப்படுத்தப்பட்ட துல்லியமும் சகிப்பு வரம்புகளும் (tolerances).
- - வெளிப்படையான வன்பொருள், மென்பொருள் தேவைகள்.
- - செயல்திறன் சோதனைகளும், பொருத்தமான இடத்தில் மாற்று வழியும் (fallback).
- - சார்புகள் மாறும்போது பராமரிப்புக்குப் பொறுப்பான ஒரு உரிமையாளர்.
சிறிய வேக முன்னேற்றம், ஆதரிக்கக் கடினமான நொய்மையான (fragile) செயலாக்கத்தை நியாயப்படுத்தாமல் போகலாம்.
14. ஓர் அமைப்பு எதைத் தேர்ந்தெடுக்க வேண்டும்?
நிறுவப்பட்ட நூலகங்களும் கட்டமைப்பு நிரல்மாற்றமும் (framework compilation) ஏற்கெனவே தேவையைப் பூர்த்தி செய்யும்போது, அவற்றிலிருந்து தொடங்கவும். தனிப்பயன் கேர்னல்களுக்குத் தொடர்ச்சியான பராமரிப்புச் செலவு உண்டு.
தனிப்பயன் டென்சர் செயல்பாடுகள், fusion மற்றும் விரைவான பரிசோதனைக்கு, குறிப்பாக Python மையமான AI குழுவில், Triton ஐக் கருத்தில் கொள்ளவும்.
வெளிப்படையான குறைந்த மட்டச் செயலாக்கக் கட்டுப்பாடு, சொந்த நிரல் (native code) ஒருங்கிணைப்பு அல்லது சிறப்பு வன்பொருள் இடைவினைகள் முக்கியமாக இருக்கும்போது CUDA C++ ஐக் கருத்தில் கொள்ளவும்.
டைல் சார்ந்த NVIDIA பணிப்பாய்வு (workflow) குழுவுக்குப் பொருந்தி, தொடர்புடைய toolkit உம் வன்பொருளும் ஆதரிக்கப்படும்போது CUDA Tile ஐக் கருத்தில் கொள்ளவும்.
பொருத்தமான, அமைப்பு அறிந்த, வன்பொருளுக்குச் சிறப்பாக அமைந்த எண்சார் கேர்னல்களுக்கு CuTe DSL ஐக் கருத்தில் கொள்ளவும். அதை cuTile இலிருந்தும் Triton இலிருந்தும் தனியாக மதிப்பிடவும்.
பல அமைப்புகளுக்குச் சிறந்த தீர்வு ஒரு கலவையாகும்: நிலையான செயல்பாடுகளுக்கு வழங்குநர் நூலகங்கள், தேர்ந்தெடுத்த தனிப்பயன் கேர்னல்களுக்கு Triton, சுயவிவரப்படுத்தல் நியாயப்படுத்தும் இடங்களில் குறைந்த மட்ட CUDA அல்லது CuTe கூறுகள்.
முடிவுரை (Conclusion)
டென்சர் சார்ந்த பல AI வேலைச்சுமைகளுக்குத் தனிப்பயன் GPU கேர்னல் உருவாக்கத்தை Triton மேலும் இலகுவாக அணுகக்கூடியதாக்கியுள்ளது. CUDA, விரிவான நூலகங்கள், கருவிகள் மற்றும் செயலாக்க வசதிகளுடன் கூடிய பரந்த தளமாகவே தொடர்கிறது.
வேறுபாடு பரிணமித்துள்ளது: நவீன CUDA டைல் அடிப்படையிலான இடைமுகங்களை வழங்குகிறது; Python அடிப்படையிலான GPU நிரலாக்கம் உயர் மட்டச் செயல்பாடுகளையோ விரிவான வன்பொருள் கட்டுப்பாட்டையோ வெளிப்படுத்த முடியும். “கைமுறை CUDA” எதிர் “தானியங்கி Triton” என்ற பழைய முரண்பாடு மிகையாக எளிமைப்படுத்தப்பட்டது.
எந்த அணுகுமுறையும் வேகமான முடிவை உறுதிப்படுத்துவதில்லை. சரியான தெரிவு வழிமுறை, வன்பொருள், உள்ளீட்டு வடிவங்கள், எண்சார் தேவைகள், பொறியியல் திறன்கள் மற்றும் பராமரிப்புச் செலவுகளைப் பொறுத்தது.
மிகவும் நம்பகமான உத்தி: வேலைச்சுமையின் தேவைகளைப் பூர்த்தி செய்யும் மிகவும் பயனுள்ள அருவமாக்கலைத் தேர்ந்தெடுத்து, சரியான தன்மையைச் சரிபார்த்து, நம்பகமான ஒப்பீட்டு அடிப்படைகளுடன் ஒப்பிட்டு, அளவீடுகள் நியாயப்படுத்தும் இடங்களில் குறைந்த மட்டக் கட்டுப்பாட்டைப் பயன்படுத்துவதாகும்.
ஆய்வுக் குறிப்பு: இது ஒரு தொழில்நுட்ப மதிப்பாய்வும் முன்மொழியப்பட்ட செயல்திறன் அளவீட்டு வழிமுறையும்; புதிய பரிசோதனைகளின் அறிக்கை அல்ல. மூல வெளியீடுகளையும், NVIDIA, OpenAI, Triton மற்றும் PyTorch இன் அதிகாரபூர்வ ஆவணங்களையும் நான் பார்வையிட்டேன். இந்தக் கட்டுரைக்காக GPU செயல்திறன் அளவீடுகளை நான் இயக்கவில்லை; SciSpace, Scite.ai, NotebookLM, Elicit அல்லது Julius AI ஐயும் பயன்படுத்தவில்லை.