ELC441 · Distributed & Parallel Processing · Dr. Aya EladghamParallel MidKit

• 1986–2002: microprocessor performance rose ~50% per year; since then only ~20% per year. زمان الأداء كان بيزيد 50% في السنة، دلوقتي 20% بس — السرعة وقفت تقريباً.
• Intelligent solution: instead of a faster single processor, put multiple processors (cores) on one chip. بدل ما نعمل بروسيسور أسرع، نحط كذا core على نفس الـ chip.
• Doing more than one task at a time is not new: already used in DMA between I/O and main memory. الفكرة مش جديدة — الـ DMA بينقل داتا بين الـ I/O والميموري والـ CPU شغال في حاجة تانية.
• Adding cores does not help if programmers don't know how to use them; serial programs usually don't benefit. الكور الزيادة مالهاش لازمة لو البرنامج serial — لازم المبرمج يكتب parallel.
• We need more performance because problems grow: climate modeling, protein folding, drug discovery, energy research, data analysis. أمثلة لمشاكل محتاجة computing ضخم — احفظ 3 منهم على الأقل.
• Physics chain: smaller transistors → higher density → shorter switching delay → higher frequency → faster processor → more power → more heat → unreliable processors. تصغير الترانزستور بيزوّد السرعة بس بيزوّد الحرارة لحد ما البروسيسور يبقى مش reliable — عشان كده وقفنا نزوّد الـ frequency.
• Solution: move from single-core to multicore ("core" = CPU) → introducing parallelism. الحل: multicore، وده معناه parallelism.
• Ways to get parallel programs: rewrite serial programs; or automatic translation (very difficult, limited success); or devise an entirely new algorithm. التحويل الأوتوماتيك من serial لـ parallel صعب ونتايجه محدودة — أحياناً لازم خوارزمية جديدة خالص.

• Task: compute n values and add them. Serial: one loop sum += x. المثال: جمع n قيمة. النسخة الـ serial لوب واحد.
• With p cores (p ≪ n): each core sums about n/p values into its private my_sum (own my_first_i, my_last_i). كل core بياخد حتة من الداتا (n/p) ويجمعها في متغير خاص بيه — ده نفس اللي عملناه في Lab 1 بالـ pthreads.
• Method 1 (master): all cores send my_sum to the master core 0, which adds them. Example 8 cores: sums 8,19,7,15,7,13,12,14 → 95. Master does 7 receives + 7 additions (p − 1). الطريقة الأولى: الـ master بيستقبل من الكل ويجمع — p−1 عملية.
• Method 2 (tree): pair cores: 0+1, 2+3, …; then 0+2, 4+6; then 0+4 … Master does 3 receives + 3 additions (log₂ p). الطريقة الأحسن: شجرة — كل مرحلة عدد الـ cores الشغالة بيتقسم على 2، والـ master بيعمل log₂p بس.
• With 1000 cores: master method = 999 receives/adds; tree = 10 → improvement ≈ 100×. كل ما الـ cores تزيد الفرق بيكبر جداً: 999 مقابل 10.

• Task parallelism: partition the various tasks of the problem among the cores. كل core بيعمل شغلانة مختلفة (زي سؤال Tut1 Q4: sort / matrix / image / simulation).
• Data parallelism: partition the data among the cores; each core does similar operations on its part. نفس العملية على أجزاء مختلفة من الداتا — زي جمع الـ array في Lab 1 والـ vector add في CUDA.
• Concurrent: multiple tasks can be in progress at any instant (dealing with many tasks in the same time period). أكتر من task "شغالين" في نفس الفترة — مش لازم في نفس اللحظة بالظبط.
• Parallel: multiple tasks cooperate closely and run at the same time. tasks بتتنفذ فعلاً في نفس اللحظة ومتعاونة جداً.
• Distributed: a program cooperates with other programs, processing across multiple computers. الشغل متوزع على كذا كمبيوتر.
• Parallel and distributed programs are concurrent (not the other way around). أي parallel أو distributed هو concurrent، بس مش كل concurrent يبقى parallel.
• Conclusions: serial programs don't benefit from multicore; automatic parallelization isn't the most efficient; parallel programming = learning to coordinate the cores. الخلاصة: البرمجة الـ parallel هي إنك تنسّق بين الـ cores (communication + load balancing + synchronization).

• Application demands: more cycles/memory for scientific computing (CFD, biology, chemistry, physics) and general-purpose (video, graphics, CAD, databases, gaming). التطبيقات محتاجة computing وميموري أكتر — ده الـ driving force.
• Technology trends: physical limitations, but Moore's law still alive (more transistors). الترانزستورات لسه بتزيد، بس فيه حدود فيزيائية للسرعة.
• Architecture trends: ILP (superscalar, VLIW) is valuable but limited; higher clock → deeper pipelines, longer latencies, higher CPI. الـ ILP جوه البروسيسور الواحد وصل لحده.
• Thread-level parallelism (TLP) in multiprocessors is the most viable way forward → main motivation for chip multiprocessors (CMPs). الحل العملي: parallelism على مستوى الـ threads — ودي فكرة الـ multicore chips.
• TOP500: started 1993 (Univ. of Mannheim); benchmark = LINPACK; first #1 = Thinking Machines CM-5/1024 at Los Alamos, 59.7 GFLOPS; updated twice a year. قايمة أسرع 500 سوبر كمبيوتر، بتتقاس بـ LINPACK وبتتحدث مرتين في السنة.
• June 2026 #1: LineShine (China, ~13.8M cores, 2,198 PFlop/s Rmax); then El Capitan, Frontier, Aurora. معلومة خفيفة من السلايد — الأهم تعرف إن الأداء بقى بالـ PetaFLOPS وملايين الـ cores.

• Parallel computer (multiple processor system) = a collection of communicating processing elements that cooperate to solve large problems fast by dividing them into parallel tasks. مجموعة processors بتتواصل مع بعض وبتقسم المشكلة لـ tasks تتنفذ مع بعض.
• Can run a single parallel task (→ speedup) or many independent tasks (→ high throughput), or both. يا إما تسرّع برنامج واحد، يا إما تشغّل برامج كتير في نفس الوقت.
• Broad issues: concurrency & communication of the algorithm (dependency graphs); computing resources & allocation (number of PEs, what is mapped to each); data access, communication & synchronization; performance (maximize speedup). القضايا: الاعتماديات بين الـ tasks، توزيع الشغل على الـ PEs، التواصل والـ synchronization، والأداء.
• Goal: maximize speedup by minimizing parallelization overheads and balancing workload. الهدف: speedup عالي = overhead قليل + توزيع شغل متساوي.
• Generic architecture = general model, not a commercial machine: multiple PEs, memory system (shared or distributed), interconnection network, I/O system, parallel software. الموديل العام: processors + ميموري + شبكة توصيل + I/O + برنامج parallel.

• Instruction & data streams → Flynn's taxonomy (Duncan's taxonomy = historical extension). حسب عدد الـ instruction streams والـ data streams.
• Memory organization → shared (common address space), distributed (each node has local memory), hybrid (shared inside node + distributed between nodes). حسب الميموري: مشتركة، موزعة، أو خليط.
• Level of parallelism (grain size) → ILP (pipelining), DLP (same operation on many elements, loops), TLP (threads on multicore), task-level (different tasks). مستوى التوازي: على مستوى الـ instruction، الداتا، الـ thread، أو الـ task.
• Grain size (reference notes): fine-grained (very small tasks, ~<20 instructions, high communication/switching overhead, e.g. matrix ops, image processing), coarse-grained (large tasks, low overhead, less parallelism, e.g. web apps), medium-grained (in between, e.g. database operations). fine = tasks صغيرة جداً و overhead عالي؛ coarse = tasks كبيرة و overhead قليل.
• Processor organization → processor arrays, multiprocessors, multicomputers, clusters. حسب تنظيم البروسيسورز.
• Processor heterogeneity → homogeneous (similar PEs) vs heterogeneous (CPU + GPU, CPU + NPU, accelerators). كلهم شبه بعض، ولا أنواع مختلفة زي CPU + GPU.
• The classifications complement each other: a CPU+GPU system is MIMD (Flynn), heterogeneous, hybrid memory, exploits ILP+DLP+TLP+task, multiprocessor. نفس الجهاز يتوصف بكل التقسيمات دي مع بعض — مش بيلغوا بعض.

• Defined 1966; classifies machines by number of instruction streams and data streams. Instruction stream = order in which the PU executes instructions; data stream = data traffic between PU and memory. التقسيم على حسب: كام سلسلة أوامر وكام سلسلة داتا.
• SISD: traditional single-processor von Neumann; one Program Counter → strictly serial. Use: no parallelism needed (text processing, basic arithmetic). Example: single-core CPU. أمر واحد على داتا واحدة — الكمبيوتر العادي القديم.
• SIMD: one instruction broadcast to many PEs, each on different data, synchronously. Components: PEs + Memory Modules (MM). Examples: GPUs, vector processors. Use: multimedia, DSP, image processing, HPC. نفس الأمر على داتا كتير في نفس اللحظة — زي إنك تزوّد سطوع كل البكسلات مرة واحدة.
• MISD: many processors, different instructions on the same data stream. Rare. Advantages: fault tolerance, performance for special algorithms. Uses: cryptography (several deciphering algorithms on same ciphertext), scientific computing, DSP. أوامر مختلفة على نفس الداتا — نادر، بيُستخدم للـ fault tolerance أو تشفير.
• MIMD: independent instruction streams on separate data, asynchronous; multiple control units + PEs. The "true" parallel computing; basis of all multicore / multiprocessor systems. كل بروسيسور ليه برنامجه وداتاه — ده الـ multicore اللي في جهازك.
• Limitation: Flynn describes behavior (streams), not physical structure → MIMD needs sub-classification by memory/interconnect. Flynn بيوصف السلوك بس، مش التركيب — عشان كده بنقسم الـ MIMD تاني حسب الميموري.
• GPU ≠ MIMD automatically: thousands of units but they run the same instruction on different data → best described as SIMT (Single Instruction, Multiple Threads); threads may diverge in control flow. كتر الوحدات مش معناه MIMD — الـ GPU أقرب لـ SIMD ويتسمى SIMT، وده مثال إن تقسيم Flynn مش كفاية.

• MIMD = many processors + many memory modules + interconnection network → split by how processors access memory: tightly coupled (shared memory) vs loosely coupled (distributed memory). الـ MIMD نوعين: ميموري مشتركة أو كل بروسيسور ليه ميموري.
• Shared memory (tightly coupled): all PEs share one global address space; communicate by reading/writing shared data; no direct messages needed. الكل شايف نفس العناوين — واحد يكتب في مكان والتاني يقرا منه. = multiprocessor.
• UMA: every PE has about the same access time to all memory; typical of SMP; easy to program. زمن الوصول للميموري واحد لكل البروسيسورز — Uniform هي الكلمة المفتاح.
• NUMA: memory physically distributed but one shared address space; local = faster, remote = slower. كل بروسيسور عنده ميموري قريبة سريعة، والبعيدة أبطأ — Non-uniform.
• COMA: memory organized as caches; data migrates toward the requesting processor; special case of NUMA where distributed memories become caches. كل الميموري كأنها cache، والداتا بتتنقل لعند البروسيسور اللي بيطلبها.
• Distributed memory (loosely coupled): no global memory; each processor has local memory not directly accessible by others; communicate by message passing (e.g. MPI). كل بروسيسور له ميموري خاصة، والتواصل برسايل عبر الشبكة. = multicomputer.
• Why distributed? Shared global memory causes memory conflicts, contention, poor scalability as processors increase. الميموري المشتركة بتزحم لما البروسيسورز تكتر — التوزيع بيحل ده بس محتاج message passing.
• Multicomputer: nodes each with processor(s), local memory, I/O, OS = collection of complete computers on a network. كل node كمبيوتر كامل.
• Tightly coupled → shared memory → shared data; loosely coupled → distributed memory → message passing. احفظ السطرين دول زي ما هما.

• Performance = how effectively a parallel system uses its processors: same work in less time, or more work in the same time. Depends on parallelism, workload distribution, communication, synchronization, memory behavior. الأداء = استغلال الـ cores صح: نفس الشغل في وقت أقل أو شغل أكتر في نفس الوقت.
• Execution time is the most fundamental measure: T₁ on one processor, Tₚ on p processors. Want Tₚ < T₁; ideal Tₚ = T₁ / p. T1 = الزمن بـ processor واحد، Tp = بـ p. المثالي إن الزمن يتقسم على p.
• Speedup Sₚ = T₁ / Tₚ; ideal Sₚ = p; example T₁ = 100 s, T₄ = 30 s → S₄ = 3.33×. Actual is usually lower than ideal. الـ speedup = القديم ÷ الجديد. المثالي يساوي عدد الـ processors، والحقيقي أقل.
• Efficiency Eₚ = Sₚ / p (×100%); example S₈ = 3.33 → E₈ = 41.6%; 100% = ideal utilization. الكفاءة = نصيب كل processor من الـ speedup. 100% يعني كل core شغال على آخره.
• Extra work in parallel execution: communication, synchronization (waiting at barriers), task/thread management (creation, scheduling, termination) → reduce speedup & efficiency. الـ overheads التلاتة اللي بتقلل الـ speedup — هي نفسها إجابة Lab 1 Q3.
• Memory performance: processors compete for memory bandwidth; cache hits/misses matter; multicore faces cache coherence, false sharing, memory contention. الميموري ممكن تبقى عنق الزجاجة — لما الـ threads كلها تقرا من الـ RAM مع بعض الـ bandwidth بيتشبع.

• Gene Amdahl, 1967. If only a fraction f of a program can run in parallel on N processors: f = الجزء اللي ينفع يتعمل parallel، و (1−f) الجزء اللي لازم يفضل serial.
• Speedup = 1 / [ (1 − f) + f / N ]
  الجزء الـ serial بياخد نفس الوقت، والجزء الـ parallel بس هو اللي بيتقسم على N.
• Conclusion 1: when f is small, parallel processors have little effect (f → 0 ⇒ speedup → 1). لو معظم البرنامج serial، زيادة الـ processors مش هتفرق.
• Conclusion 2: speedup is bounded by 1 / (1 − f) regardless of N (N → ∞). مهما زوّدت الـ processors، الحد الأقصى 1/(1−f). مثال: f = 0.8 ⇒ أقصى speedup = 5.
• Generalized Amdahl: an enhancement speeds up fraction f by factor k ⇒ Speedup = 1 / [(1 − f) + f / k]. نفس القانون لأي تحسين: بدل N حط k (معامل التسريع).
• Tutorial example (f = 0.8): S(8) = 1/(0.2 + 0.1) = 3.33, S(100) = 1/(0.2 + 0.008) ≈ 4.81, S(∞) = 5. 100 processor بيدوا 4.81 بس مش 100 — الـ 20% الـ serial هي اللي بتحكم.

• Design constraints: capacity (bigger is better), speed (keep up with CPU), cost → trade-off; no single technology has it all → memory hierarchy. مفيش ميموري كبيرة وسريعة ورخيصة مع بعض — فبنعمل طبقات.
• Going down the hierarchy: less cost/bit, higher capacity, greater access time, less access frequency. Registers → cache → main memory → disk → tape. كل ما تنزل: أرخص، أكبر، أبطأ، وبيتطلب أقل.
• Locality of reference: memory references cluster (loops, subroutines, arrays/tables); over a short period the CPU works with fixed clusters; over a long period the clusters change. البرنامج بيشتغل على مجموعة عناوين قريبة من بعض لفترة — ده اللي بيخلي الكاش ينفع.

• Cache = small fast memory between CPU and main memory, may be on the CPU chip, not visible to the programmer, volatile semiconductor. ميموري صغيرة سريعة بين الـ CPU والـ RAM.
• Operation: CPU requests a word → check cache → hit: deliver from cache (fast); miss: read the whole block from MM into cache, then deliver to CPU. Cache keeps tags to know which MM block is in each line. hit = لقيتها في الكاش. miss = هات البلوك كله من الـ RAM (عشان الـ locality). الـ tag بيقول البلوك ده جاي منين.
• Word transfer CPU↔cache; block transfer cache↔MM. Multi-level: L1 (fastest) → L2 → L3 → MM (slow). بين الـ CPU والكاش بننقل word، بين الكاش والـ RAM بننقل block.
• Design elements: 1. mapping function, 2. replacement algorithm, 3. read/write policies, 4. number of caches, 5. addresses, 6. size, 7. block/line size. السبع قرارات في تصميم الكاش — أول أربعة هم اللي اتشرحوا بالتفصيل.
• Mapping function = how MM blocks are mapped to cache lines: direct, associative, set-associative. الـ mapping = البلوك يتحط في أنهي line.

• Each MM block maps to exactly one cache line: line = b mod m (m = number of lines). كل بلوك ليه مكان واحد بس في الكاش = رقم البلوك mod عدد الـ lines.
• Address = Tag (s − r) | Line (r) | Word (w); w = log₂(words per block), s = log₂(blocks in MM), r = log₂(lines in cache). العنوان: آخر w bits = الكلمة جوه البلوك، بعدهم r bits = رقم الـ line، والباقي tag.
• Shortcuts: s + w = address length; tag = log₂(MM size / cache size). اختصار: الـ tag = log₂ (حجم الـ RAM ÷ حجم الكاش).
• Tiny example: 32 B MM (5-bit address), 4-B blocks (M = 8), 16-B cache (m = 4) ⇒ w = 2, r = 2, tag = 1. Address 10110 → tag 1, line 01, word 10. مثال صغير تتمرن عليه بإيدك.
• Big example: 16 MB MM (24 bits), 4-B blocks (4M blocks), 64 KB cache (16K lines) ⇒ Tag 8 | Line 14 | Word 2. Address 16339C → tag 16, line 0CE7, word 0. المثال الكبير: 8/14/2. قسّم الـ hex لـ binary وبعدين قطّع من اليمين.
• Lookup: go to the line, compare one tag. No two blocks mapping to the same line have the same tag. مقارنة واحدة بس — عشان كده سريع ورخيص.
• Pros: simple, inexpensive. Con: fixed location → if a program repeatedly accesses 2 blocks mapping to the same line, misses are very high (thrashing). العيب: بلوكين بيتخانقوا على نفس الـ line فالـ miss بيزيد.

• A block can load into any cache line. Address = Tag (s) | Word (w); the tag is the whole block number. البلوك يروح أي line فاضية. العنوان جزئين بس: tag و word.
• Lookup: every line's tag is compared (in parallel) → searching is expensive. Number of lines is not determined by the address format. لازم تقارن مع كل الـ tags — هاردوير غالي.
• Tiny: tag 3 bits, word 2 (10110 → tag 101, word 10). Big: Tag 22 | Word 2; 16339C → tag 058CE7. في المثال الكبير الـ tag = 22 bit.

• Cache divided into v sets, each with k lines (k-way). Block b goes to set i = b mod v, in any line of that set. حل وسط: البلوك ليه set واحدة بس جواها يختار أي line من الـ k.
• Address = Tag (s − d) | Set (d) | Word (w); v = 2ᵈ; lines m = k × v. زي الـ direct بالظبط بس بدل line bits بقى set bits.
• Number of sets = cache size / (k × line size). عدد الـ sets = حجم الكاش ÷ (عدد الـ ways × حجم الـ line) — أهم معادلة في Tut2.
• Big example (2-way): 16K lines → 8K sets ⇒ Tag 9 | Set 13 | Word 2; 16339C → tag 02C, set 0CE7. في المثال الكبير: 9/13/2.
• Special cases: k = 1 → direct; v = 1 (k = all lines) → fully associative. 1-way = direct، وset واحدة فيها كل الـ lines = fully associative.

• Direct mapping: no choice — each block has only one line; replace it. في الـ direct مفيش اختيار.
• Associative / set-associative use hardware algorithms: في الباقي لازم نختار مين يطلع:
  • LRU — replace the block not accessed for the longest time.اللي بقاله أطول وقت محدش لمسه (قراءة أو كتابة).
  • FIFO — replace the oldest block (first loaded).أقدم واحد دخل، حتى لو لسه مستخدم.
  • LFU — replace the block with the fewest hits.أقل واحد اتعمله hits.
  • Random — replace any block.أي واحد عشوائي.

• Reads dominate (every instruction fetch is a read; most instructions don't write). معظم الوصول للميموري قراءة.
• Read-hit: read word from cache. Read-miss: read-through = read word from MM and bring block to cache; no-read-through = bring block to cache, then read from cache. read-through: الكلمة بتروح للـ CPU على طول من الـ RAM وفي نفس الوقت البلوك بيتحمل.
• Write-hit — write-through: write to both cache and MM. + simple, MM always consistent, read miss never causes a write. − every write goes to MM (slow, more bandwidth). اكتب في الاتنين على طول — بسيط بس بطيء.
• Write-hit — write-back: write to cache only; update MM when the block is replaced; needs a dirty bit. + many writes = one MM write, faster. − harder, MM not always consistent, a read miss may force writing a dirty block. اكتب في الكاش بس وعلّم dirty، ولما البلوك يطلع اكتبه في الـ RAM — عشان كده بيظهر عمود Write-back? في Tut2.
• Write-miss — write-allocate: bring the block into cache, then act as write-hit. No-write-allocate: write the word to MM only, don't bring the block. allocate = هات البلوك الأول. no-allocate = اكتب في الـ RAM ومتجيبش حاجة.
• Usual pairs: write-back + write-allocate; write-through + no-write-allocate. الأزواج المعتادة — زي MS2 و MS1 في Tut2 Q4.

• Several CPUs share one MM, each with its own cache ⇒ if data changes in one cache, the copies in MM and other caches become invalid — even with write-through. لو كل CPU عنده كاش، أي تعديل في كاش بيخلي النسخ التانية قديمة — حتى مع write-through.
• Cache coherency = keeping data in all caches consistent. Solutions: bus watching with write-through (controller snoops and invalidates); hardware transparency (extra hardware updates MM and other caches); non-cacheable memory (shared portion not cached). 3 حلول سريعة — التفاصيل في محاضرة 4.
• Multi-level caches: L1 closest, fastest, smallest; then L2, L3; CPU checks L1 → L2 → MM. Higher levels used to be off-chip, not anymore. L1 الأصغر والأسرع، وبعدين L2 و L3، وكلهم بقوا جوه الـ chip.
• From Tut2 Q1: split cache = separate instruction and data caches (L1 I + L1 D); shared cache = one cache used by all cores (L3); unified = instructions and data together. split = كاش للأوامر وكاش للداتا. shared = كل الـ cores بتستخدمه. لو مكتوب 2 × يبقى private لكل core.

• Any processor can directly reference any memory location; communication is implicit through loads/stores (+ synchronization). التواصل بيحصل لوحده من خلال القراءة والكتابة في الميموري المشتركة.
• Private caches may hold multiple inconsistent copies of the same data ⇒ cache coherence problem, must be solved in hardware. الكاشات الخاصة هي سبب المشكلة.
• UMA types: bus-based SMP (2–4 processors on a shared bus/FSB, symmetric access, uses bus snooping); shared cache (CMPs, e.g. IBM Power 4: 2 cores share L2 — no coherence problem but high bandwidth needs, conflicts, longer hit latency); dancehall (no local memory, all memory equally far — not popular). 3 أنواع UMA: bus مشترك، كاش مشترك، أو dancehall (الميموري كلها بعيدة).
• NUMA (distributed shared memory): memory physically distributed among nodes, remote access slower; most popular for scalable MPPs; coherence by directory-based methods. الـ NUMA بيكبر أحسن، والـ coherence فيه بالـ directory.
• Examples: Intel Pentium Pro Quad (UMA, one shared FSB limits to 2–4 processors); AMD 8-way Opteron (NUMA, point-to-point HyperTransport, each CPU has own memory controllers); IBM Power 4 (CMP, shared L2). أمثلة من السلايدات — اربط كل واحد بنوعه.
• Interleaving (S8): split memory into banks so accesses overlap → higher bandwidth. تقسيم الميموري لـ banks عشان تبدأ قراءة جديدة قبل ما القديمة تخلص.

• Intuition: reading a location should return the latest value written by any processor. أي قراءة لازم ترجع آخر قيمة اتكتبت.
• Easy in uniprocessors (only DMA I/O issue, solved in software); in multiprocessors it is pervasive and performance-critical → basic hardware design issue. في الـ multiprocessor المشكلة في كل حتة ولازم الهاردوير يحلها.
• Caches reduce AMAT and interconnect bandwidth, but copies of a variable in many caches → a write by one processor may not be visible to others (stale values). الكاش بيسرّع، بس بيعمل نسخ قديمة (stale).
• Coherence requires: 1) write visibility to all processors, 2) in the correct order. Definition: all cached copies of shared data have the same value at all times. شرطين: الكل يشوف الكتابة، وبنفس الترتيب.
• Example (S15): P1 reads u=5, P3 reads u=5, P3 writes u=7 (write-back), then P1 reads old 5 from its cache, P2 reads old 5 from memory ⇒ incorrect execution. المثال الكلاسيكي: P3 غيّر u، والباقي لسه شايفين 5.

• Write-through with update: new value written to cache and memory, and broadcast to all other caches holding copies (they get updated). اكتب في الـ RAM وابعت القيمة الجديدة لكل الكاشات اللي معاها نسخة.
• Write-through with invalidation: new value written to cache and memory; all other copies are invalidated. اكتب في الـ RAM وامسح (invalidate) النسخ التانية.
• Write-back (ownership): to write, a processor must become exclusive owner; its cache becomes the home of the block; all other copies including memory's are invalidated; owner writes freely; on another processor's read, the owner sends the data to it and to memory. في الـ write-back: لازم تبقى المالك الوحيد الأول، وأي حد يطلب البلوك المالك يبعتهوله ويحدّث الـ RAM.
• Write-invalidate vs write-update (Tut2 Q6): invalidate = one message on first write, later writes local; reader takes a coherence miss. Update = every write broadcast → readers hit, but more traffic if writes are many and reads are rare. invalidate أوفر لو واحد بيكتب كتير والتاني بيقرا قليل. update أحسن لو ناس كتير بتقرا كتير.

• Shared caches: no private caches → no problem, but limits scalability (only CMPs). مفيش كاش خاص = مفيش مشكلة، بس مش بيكبر.
• Non-cacheable data (software): don't cache shared writable data (locks, queues, critical-section data); compiler tags data; not the best. متخليش الداتا المشتركة تتكيّش أصلاً.
• Cache flushing: flush whenever a synchronization primitive executes; slow without special hardware. امسح الكاش عند كل synchronization — بطيء.
• The first three are not scalable. Main hardware solutions: directory protocol and snoopy (sniffing) protocols. الحلول الحقيقية: directory أو snooping.

• A central directory tracks which caches hold copies of each memory block and coordinates coherence actions. جدول مركزي عارف مين معاه نسخة من كل بلوك.
• Write: processor asks the directory for permission → directory tells the other holders to invalidate (and memory copy) → grants permission. Later readers get the updated value. عشان تكتب: استأذن الـ directory، هو يبعت invalidate للي معاهم نسخ بس، وبعدين يديك الإذن.
• Read: processor sends a read request → directory checks status, decides if memory or another cache supplies data → processor gets the current value; directory updates the sharer list. عشان تقرا: الـ directory يقرر مين يديك الداتا (الـ RAM ولا كاش فيه آخر نسخة) ويضيفك للـ sharers.
• Example (S21): P1, P2 share A; P3 wants to write ⇒ answer B: identify P1, P2 and invalidate their copies before granting write. السؤال اللي في السلايد — الإجابة B.
• Directory stores per block: state (uncached / shared / exclusive-modified) + sharer bit-vector or owner ID. Tracks blocks, not bytes. لكل بلوك: حالته + مين معاه. وبيتابع على مستوى البلوك عشان التخزين ميبقاش ضخم.

• Bus transactions are visible to all processors; responsibility is distributed; each cache controller snoops the bus and reacts (changes state) — watching others' interactions too. كل كاش بيتصنت على الـ bus، ولو شاف حد بيكتب في بلوك عنده يتصرف.
• Types: write-update (broadcast new block to all holders) and MESI (an invalidate protocol: invalidate all remote copies on a local write). نوعين: update (ابعت القيمة) أو MESI (امسح النسخ التانية).
• E — Exclusive: only I have it, clean (= memory); read freely; can go to M silently. أنا الوحيد اللي معايا، ونفس قيمة الـ RAM.
• M — Modified: I wrote it, still the only owner, dirty (memory stale); must write back / supply on others' request. أنا عدّلت ومحدش معاه غيري — الـ RAM قديمة.
• When another processor requests a block I hold in M ⇒ I write it back; the requester loads it; both go to S. لو حد طلبه وأنا M: أكتبه في الـ RAM ونبقى إحنا الاتنين Shared.
• S — Shared: more than one processor has it; stays S even if one copy is evicted. أكتر من كاش معاهم نسخة نضيفة.
• Writing a Shared block ⇒ writer goes to M, all others go I (Invalid). لو حد كتب في S: هو يبقى M والباقي I.

• Heterogeneous computing: use the best match for the job — latency cores, throughput cores, DSP, configurable logic, HW IPs (e.g. mobile SoC). كل نوع شغل على الوحدة المناسبة ليه.
• CPU = latency-oriented: large caches (turn long memory latency into short cache latency), sophisticated control (branch prediction, data forwarding), powerful ALUs (reduced operation latency). الـ CPU مصمم يخلّص الأمر الواحد بأسرع ما يمكن.
• GPU = throughput-oriented: small caches (to boost memory throughput), simple control (no branch prediction, no data forwarding), many energy-efficient ALUs, long latency but heavily pipelined → needs a massive number of threads to tolerate latency. الـ GPU مصمم يخلّص كمية شغل ضخمة: آلاف الـ threads، وأي thread مستني ميموري يشتغل غيره.
• Winning applications use both: CPU for sequential parts (10+× faster there), GPU for parallel parts (10+× faster there). الجزء الـ serial على الـ CPU، والجزء الـ parallel على الـ GPU.

• Objective: data parallelism + CUDA C basics: hierarchical thread organization, launching parallel execution, thread index → data index mapping. 3 حاجات المحاضرة عايزاها منك.
• Vector addition: C[i] = A[i] + B[i] — every element independent ⇒ perfect data parallelism. كل عنصر مستقل عن التاني، فكل thread ياخد عنصر.
• Execution model: serial parts in host C code; parallel parts in device SPMD kernel code launched as KernelA<<<nBlk, nTid>>>(args). البرنامج بيتنقل بين كود serial على الـ CPU و kernel parallel على الـ GPU.
• NVCC splits the program: host code → host C compiler/linker; device code (PTX) → device just-in-time compiler. الـ nvcc بيفصل كود الـ CPU عن كود الـ GPU.
• ISA = contract between hardware and software (the set of instructions hardware can execute). A program = instructions in memory operating on data. الـ ISA هي "العقد" بين الهاردوير والسوفتوير.
• A thread is a "virtualized" von Neumann processor (its own PC, IR, registers). كل thread كأنه بروسيسور صغير افتراضي.

• A kernel is executed by a grid (array) of threads; all run the same kernel code (SPMD); each uses its indices to compute addresses and make control decisions. كل الـ threads بتنفذ نفس الكود، والفرق بس في الـ index.
• The grid is divided into thread blocks. Threads within a block cooperate via shared memory, atomic operations, barrier synchronization; threads in different blocks do not interact. جوه البلوك يقدروا يتعاونوا، بين البلوكات لأ.
• blockIdx and threadIdx can be 1D, 2D or 3D → simplifies addressing for images, volumes, PDEs. الـ indices ممكن تبقى 2D أو 3D عشان الصور والمصفوفات.
• i = blockIdx.x * blockDim.x + threadIdx.x
  أهم سطر في الكورس: رقم البلوك × حجم البلوك + رقم الـ thread جوه البلوك.

• Host (CPU) memory and device (GPU) memory are separate; data must be copied explicitly (over PCIe). ميموري الـ CPU غير ميموري الـ GPU — لازم تنقل بإيدك.
• Device code can R/W per-thread registers and all-shared global memory; host code can transfer data to/from global memory. الـ kernel بيشتغل على global memory والـ registers.
• cudaMalloc(void **ptr, size_in_bytes) — allocate in device global memory (address of a pointer + bytes). cudaFree(ptr) — free. cudaMalloc بياخد عنوان الـ pointer (عشان كده &d_A) والحجم بالـ bytes.
• cudaMemcpy(dst, src, bytes, direction) — 4 parameters; direction: cudaMemcpyHostToDevice / DeviceToHost. الترتيب: المقصد الأول، بعدين المصدر، الحجم، الاتجاه.
• Steps of vecAdd host function: 1. allocate d_A, d_B, d_C + copy A, B to device; 2. launch kernel; 3. copy C back + free. 3 خطوات — نفس ترتيب Lab 2.
• In practice check errors: cudaError_t err = cudaMalloc(...); if (err != cudaSuccess) printf(cudaGetErrorString(err) ...). كل API بترجع cudaError_t — اتأكد إنها cudaSuccess.

void vecAdd(float* h_A, float* h_B, float* h_C, int n) {
  int size = n * sizeof(float);  float *d_A, *d_B, *d_C;
  cudaMalloc((void**)&d_A, size); cudaMemcpy(d_A, h_A, size, cudaMemcpyHostToDevice);
  cudaMalloc((void**)&d_B, size); cudaMemcpy(d_B, h_B, size, cudaMemcpyHostToDevice);
  cudaMalloc((void**)&d_C, size);
  vecAddKernel<<<ceil(n/256.0), 256>>>(d_A, d_B, d_C, n);   // launch
  cudaMemcpy(h_C, d_C, size, cudaMemcpyDeviceToHost);
  cudaFree(d_A); cudaFree(d_B); cudaFree(d_C);
}

__global__
void vecAddKernel(float* A, float* B, float* C, int n) {
  int i = threadIdx.x + blockDim.x * blockIdx.x;
  if (i < n) C[i] = A[i] + B[i];     // boundary check
}
// host side
dim3 DimGrid((n-1)/256 + 1, 1, 1);   // = ceil(n/256.0)
dim3 DimBlock(256, 1, 1);
vecAddKernel<<<DimGrid, DimBlock>>>(d_A, d_B, d_C, n);

• Each thread does one pair-wise addition. if (i < n) is needed because the last block may have extra threads. الـ if عشان آخر بلوك ممكن يبقى فيه threads زيادة عن n.
• Number of blocks = ceil(n / 256.0) = (n − 1)/256 + 1 (integer) → "enough threads to cover all elements". الصيغتين بيدوا نفس النتيجة. مثلاً n = 1000 ⇒ 4 blocks = 1024 thread، 24 منهم مش بيعملوا حاجة.
• __global__: kernel, executed on device, callable from host, must return void. __device__: on device, from device. __host__: on host, from host. __device__ and __host__ can be used together. global = الـ CPU بينادي والـ GPU بينفّذ. device = جوه الـ GPU بس. host = دالة CPU عادية.

• !nvidia-smi shows the GPU (Tesla T4) and driver/CUDA version → confirms a GPU runtime. أمر بيعرض كارت الـ GPU — لو الـ runtime CPU بس هيطلع error إنه مش لاقي driver.
• %%writefile add_arrays.cu saves the cell as a file; !nvcc add_arrays.cu -o add_arrays compiles; !./add_arrays runs → 11 22 33 44 55. اكتب الملف، اعمله compile بالـ nvcc، شغّله.
• Example uses threads = 256; blocks = (N + threads − 1) / threads;, then cudaGetLastError() and cudaDeviceSynchronize() to catch errors. (N + T − 1)/T هي نفسها ceil(N/T) — صيغة تالتة لنفس الحاجة.

• Host (CPU): prepares data, allocates memory, launches the kernel. Device (GPU): executes many threads in parallel; the kernel is the function those threads run. الـ host بيجهّز ويشغّل، الـ device بينفّذ.
• Grid = all blocks of one kernel launch; block = group of threads that can cooperate and share block resources; thread = executes the kernel for one piece of work. grid ← blocks ← threads.
• blockIdx.x = block index; blockDim.x = threads per block; threadIdx.x = thread index inside its block; if (i < N) prevents out-of-range access. 3 متغيرات built-in لازم تعرف كل واحد بيرجع إيه.
• Worked example <<<2, 4>>>: blockIdx 1, blockDim 4, threadIdx 2 ⇒ i = 1×4 + 2 = 6 → this thread computes C[6]. Indices 0–7 cover 8 elements. المثال اللي في Lab 2 بالظبط.
• <<<numberOfBlocks, threadsPerBlock>>> describes logical work, not the number of physical GPU cores; the GPU schedules blocks onto hardware. عدد الـ threads اللي بتطلبه مش لازم يساوي عدد الـ cores الحقيقية.
• Workflow: 1 prepare input on host → 2 copy to GPU → 3 launch kernel → 4 threads compute → 5 copy results back. الخمس خطوات — ممكن يطلب ترتيبهم أو يحط سطر في مكان غلط.

• Copy directions: HostToDevice (CPU→GPU), DeviceToHost, DeviceToDevice, HostToHost. 4 اتجاهات.
• cudaError_t cudaMalloc(void **devPtr, size_t size): returns success/failure; devPtr is a pointer to a pointer that receives the GPU address; size in bytes. ليه void**؟ عشان الدالة محتاجة تغيّر قيمة الـ pointer بتاعك نفسه.

• Blocks and threads can be 2D/3D: e.g. grid of 2×2 blocks, each 4×2×2 threads. Mapping block/thread indices to data indices (rows, columns). في الصور بنستخدم x للأعمدة و y للصفوف.
• Row-major layout in C: a 2D matrix is stored row after row ⇒ element (Row, Col) is at Row*Width + Col; e.g. M[2][1] in a 4-wide matrix → 2×4+1 = 9. المصفوفة في الميموري صف ورا صف، فالعنوان = الصف × العرض + العمود.
• Row = blockIdx.y*blockDim.y + threadIdx.y; Col = blockIdx.x*blockDim.x + threadIdx.x; guard if (Row < m && Col < n). نفس معادلة الـ 1D مرتين: مرة لـ y (Row) ومرة لـ x (Col).
• 76×62 picture (n = 76 columns, m = 62 rows) with 16×16 blocks ⇒ grid = ⌈76/16⌉ × ⌈62/16⌉ = 5 × 4 = 20 blocks = 5120 threads for 4712 pixels ⇒ 408 idle threads (blocked by the if). البلوكات اللي على الطرف (مناطق 2 و 3 و 4 في الرسمة) فيها threads برا الصورة — الـ if بتمنعهم.

__global__ void PictureKernel(float* d_Pin, float* d_Pout, int n, int m) {
  int Row = blockIdx.y*blockDim.y + threadIdx.y;
  int Col = blockIdx.x*blockDim.x + threadIdx.x;
  if ((Row < m) && (Col < n))
    d_Pout[Row*n + Col] = 2*d_Pin[Row*n + Col];   // scale every pixel by 2
}
// host: dim3 DimGrid(ceil(n/16.0), ceil(m/16.0), 1); dim3 DimBlock(16, 16, 1);

• Host version: 3 nested loops (i, j, k): P[i*Width+j] = Σₖ M[i*Width+k] * N[k*Width+j]. في الـ CPU: لوبين على العنصر ولوب تالت للضرب والجمع.
• GPU: each thread computes one element P[Row][Col] (removes the i, j loops; keeps the k loop). كل thread مسؤول عن خانة واحدة في الناتج — اللوب الداخلي بس هو اللي فاضل.
• Each 2D block computes a TILE_WIDTH × TILE_WIDTH tile ⇒ grid of (Width/TILE_WIDTH)² blocks, each with TILE_WIDTH² threads. Width=4, TILE=2 ⇒ 4 blocks × 4 threads. Width=8, TILE=2 ⇒ 16 blocks × 4. Width=8, TILE=4 ⇒ 4 blocks × 16.
• Block (0,1) with TILE 2: Col = 1*2 + threadIdx.x ∈ {2,3}, Row = 0*2 + threadIdx.y ∈ {0,1} ⇒ computes P[0..1][2..3]. blockIdx.x بيحدد العمود، blockIdx.y بيحدد الصف.

__global__ void MatrixMulKernel(float* M, float* N, float* P, int Width) {
  int Row = blockIdx.y*blockDim.y + threadIdx.y;
  int Col = blockIdx.x*blockDim.x + threadIdx.x;
  if ((Row < Width) && (Col < Width)) {
    float Pvalue = 0;
    for (int k = 0; k < Width; ++k)
      Pvalue += M[Row*Width + k] * N[k*Width + Col];
    P[Row*Width + Col] = Pvalue;
  }
}
// host
dim3 dimGrid(Width/TILE_WIDTH, Width/TILE_WIDTH, 1);
dim3 dimBlock(TILE_WIDTH, TILE_WIDTH, 1);
MatrixMulKernel<<<dimGrid, dimBlock>>>(Md, Nd, Pd, Width);

• Hardware may assign blocks to any processor at any time ⇒ a kernel scales to any number of parallel processors; blocks execute in any order relative to each other. نفس الكود يشتغل على GPU صغير (بلوكين في المرة) أو كبير (4 في المرة) من غير تعديل — عشان كده البلوكات مينفعش تعتمد على بعض.
• Threads are assigned to Streaming Multiprocessors (SMs) in block granularity; SM maintains thread/block indices and schedules execution. الـ SM بياخد بلوك كامل مرة واحدة، مش threads متفرقة.
• Up to 8 blocks per SM as resources allow; a Fermi SM takes up to 1536 threads (e.g. 256 × 6 blocks, or 512 × 3 blocks). الحدين مع بعض: 8 بلوكات و 1536 thread — أيهما يخلص الأول.

#include <pthread.h>   // POSIX threads
#define N 100000000L        // 10^8 elements
long long *array;          // shared by all threads (global)

typedef struct { long start; long end; long long result; } ThreadData;

void *sum_part(void *arg) {                 // the thread function
  ThreadData *data = (ThreadData *)arg;
  data->result = 0;
  for (long i = data->start; i < data->end; i++)
    data->result += array[i];                 // private partial sum (my_sum)
  return NULL;
}

int main(int argc, char *argv[]) {
  int num_threads = (argc > 1) ? atoi(argv[1]) : 1;   // ./parallel_sum 4
  array = malloc(N * sizeof(long long));
  for (long i = 0; i < N; i++) array[i] = 1;       // expected total = N
  pthread_t threads[num_threads];  ThreadData data[num_threads];
  long chunk = N / num_threads;                 // floor(N/p)
  double start = get_time();                    // timing starts AFTER init
  for (int i = 0; i < num_threads; i++) {
    data[i].start = i * chunk;
    data[i].end = (i == num_threads - 1) ? N : (i + 1) * chunk;  // last thread takes the remainder
    pthread_create(&threads[i], NULL, sum_part, &data[i]);       // fork
  }
  long long total = 0;
  for (int i = 0; i < num_threads; i++) {
    pthread_join(threads[i], NULL);              // wait (join)
    total += data[i].result;                     // master combines = serial part
  }
  double end = get_time();
  printf("Threads: %2d | Total: %lld | Time: %.6f s\n", num_threads, total, end - start);
}

• Domain decomposition: chunk = ⌊N/p⌋; thread i gets [i·chunk, (i+1)·chunk); the last thread gets the remainder up to N. كل thread بياخد حتة متصلة من الـ array، والأخير بياخد الباقي لو N مش بتقبل القسمة على p.
• Each thread writes to its own data[i].result ⇒ no race condition, no mutex needed. مفيش thread بيكتب في مكان thread تاني ⇒ مفيش lock.
• pthread_create(&tid, attr, func, arg) starts a thread; pthread_join(tid, NULL) waits for it. Master combines results = the global sum from Lecture 1 (master method, p − 1 additions). create = ابدأ، join = استنى. الجمع النهائي هو الطريقة الأولى من محاضرة 1.
• clock_gettime(CLOCK_MONOTONIC): high-resolution wall-clock; timing excludes malloc + initialization, includes thread creation/join. الوقت بيتقاس للحساب + إنشاء الـ threads بس، مش للـ init.
• Compile: gcc -O2 parallel_sum.c -o parallel_sum -pthread; run ./parallel_sum p; check cores with nproc / lscpu. -pthread عشان الـ linking، -O2 للتحسين.

شغّلت نفس الكود بتاعك على جهاز فيه 2 core بس عشان تشوف الشكل الحقيقي. أرقام جهازك هتختلف (أكيد عندك cores أكتر) — بس طريقة الحساب واحدة.

• Best Tp = minimum of the trials (least disturbed run). بناخد أقل زمن لأنه أقل run اتأثر بحاجات تانية.
• Speedup grows until p = number of cores (2), then flattens/drops; efficiency keeps falling. بعد عدد الـ cores الحقيقي الـ speedup بيقف والكفاءة بتقع — ده "المنحنى الحقيقي" اللي الدكتورة كانت عايزاكم تشوفوه.

#include <cuda_runtime.h>
#define N 8

__global__ void vectorAdd(const int *A, const int *B, int *C, int n) {
  int i = blockIdx.x * blockDim.x + threadIdx.x;
  if (i < n) { C[i] = A[i] + B[i]; }
}

int main() {
  int A[N], B[N], C[N];
  int *d_A = NULL, *d_B = NULL, *d_C = NULL;
  size_t size = N * sizeof(int);
  for (int i = 0; i < N; i++) { A[i] = (i + 1) * 10; B[i] = (i + 1); }  // 1. init host
  cudaMalloc((void**)&d_A, size); cudaMalloc((void**)&d_B, size); cudaMalloc((void**)&d_C, size); // 2
  cudaMemcpy(d_A, A, size, cudaMemcpyHostToDevice);                      // 3. H → D
  cudaMemcpy(d_B, B, size, cudaMemcpyHostToDevice);
  dim3 threadsPerBlock(4);  dim3 blocksPerGrid(2);                    // 4. launch config
  vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C, N);
  cudaDeviceSynchronize();                                               // wait for GPU
  cudaMemcpy(C, d_C, size, cudaMemcpyDeviceToHost);                      // 5. D → H
  for (int i = 0; i < N; i++) printf("Index %d: %d + %d = %d\n", i, A[i], B[i], C[i]); // 6
  cudaFree(d_A); cudaFree(d_B); cudaFree(d_C);                           // 7. free
}

• Output: Index i: (i+1)·10 + (i+1) = 11, 22, 33, 44, 55, 66, 77, 88. كل نتيجة = 11 × (i+1).
• Compile/run: !nvcc -O2 vector_add_basic.cu -o vector_add_basic then !./vector_add_basic (capital letter O). Host code → C++ compiler; device code → PTX/SASS. nvcc بيعمل compile، و ./ بيشغّل.
• Objectives: host vs device memory (copies over PCIe), 1D global index, adapting <<<grid, block>>> to N, boundary protection if (i < N). 4 أهداف اللاب — هما نفسهم محاور الأسئلة.

السؤال الأصلي

القاعدة: لازم إجمالي الـ threads (blocks × threads) ≥ N. لو أقل ⇒ ناقص عناصر. لو أكتر ⇒ تمام بشرط وجود الـ if.