hello_CUDA.ipynb

{
  "nbformat": 4,
  "nbformat_minor": 0,
  "metadata": {
    "colab": {
      "name": "Untitled1.ipynb의 사본",
      "version": "0.3.2",
      "provenance": [],
      "collapsed_sections": [],
      "include_colab_link": true
    },
    "kernelspec": {
      "name": "python3",
      "display_name": "Python 3"
    },
    "accelerator": "GPU"
  },
  "cells": [
    {
      "cell_type": "markdown",
      "metadata": {
        "id": "view-in-github",
        "colab_type": "text"
      },
      "source": [
        "<a href=\"https://colab.research.google.com/github/jeonggunlee/CUDATeaching/blob/master/hello_CUDA.ipynb\" target=\"_parent\"><img src=\"https://colab.research.google.com/assets/colab-badge.svg\" alt=\"Open In Colab\"/></a>"
      ]
    },
    {
      "metadata": {
        "id": "1CZGtc9NKiDI",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "# **GPU 기반의 고성능 CUDA 프로그래밍**\n",
        "## 한림대학교 소프트웨어 융합대학 이정근\n",
        "## 2019년 2월\n",
        "\n",
        "안녕하세요. 반갑습니다. 한림대학교 소프트웨어융합대학 이정근 교수라고 합니다.\n",
        "이번 GPU CUDA 교육에 참여해 주셔서 감사합니다. 본 강의는 이론과 실습으로 이루어져 있으며, 실습은 Colab을 통해서 진행될 예정입니다.\n",
        "\n",
        "우선 교육에 참여하신 모든 선생님들께서는 Github 및 Colab 계정을 만들어 주시면 감사드리겠습니다.\n",
        "\n",
        "* Github 주소: www.github.com\n",
        "* Colab 주소: colab.research.google.com\n",
        "* Google Drive 연동 : https://github.com/jeonggunlee/CUDATeaching/blob/master/colab_gdrive.ipynb\n",
        "\n",
        "*  *  *\n",
        "\n",
        "Colab에서 CUDA Coding을 실습하기 위해서는 몇가지 사항을 알아야합니다.\n",
        "\n",
        "* [코드 셀]에서 command-line 명령어 실행시키기\n",
        "   - !ls : 현재 디렉토리의 내용을 보여준다.\n",
        "   - %cd *dir*: *dir* 디렉토리로 이동한다.\n",
        "   - %pwd: 현재 위치한 디렉토리 위치를 보여준다.\n",
        "   - !git: git 명령어를 실행시킨다.\n",
        "\n",
        "* 또한 GPU 장치와 관련되어 다음 명령어를 확인해주시기 바랍니다.\n",
        "   - !nvidia-smi: 현재 사용하고 있는 GPU의 스펙과 작동 상황을 보여준다.\n",
        "   - !nvcc: Nvidia CUDA Compiler를 실행시킨다.\n",
        "   \n",
        "위의 명령어들을 실행 시켜 보시기 바랍니다. \n",
        "\n",
        "아래의 내용은 youtube 동영상을 통하여 확인할 수 있습니다. [colab을 이용한 GPU CUDA 프로그래밍 ](https://youtu.be/pT38R3jXwe0)\n"
      ]
    },
    {
      "metadata": {
        "id": "Pb77KWtoALVn",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "처음 명령어로 pwd 명령어를 실행시켜보도록 하겠습니다. pwd는 사용자가 위치한 디렉토리를 보여줍니다."
      ]
    },
    {
      "metadata": {
        "id": "ZHD_AWXq-N_m",
        "colab_type": "code",
        "outputId": "227af3e1-aee4-4171-d235-faf79b6eda44",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        }
      },
      "cell_type": "code",
      "source": [
        "!pwd"
      ],
      "execution_count": 1,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "/content\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "d3PXHN9YAV39",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "현재의 디렉토리가 content에 위치하고 있네요. 기본적으로 colab을 사용하는 사용자는 항상 위와 같은 위치를 초기 디렉토리로 사용하게 됩니다.\n",
        "다음으로 Unix/Linux 계열의 가장 기본적인 명령인 ls를 사용해보도록 하겠습니다. ls 명령어는 현재 디렉토리에 있는 화일 또는 디렉토리를 보여줍니다."
      ]
    },
    {
      "metadata": {
        "id": "DEYYd1IaNw9b",
        "colab_type": "code",
        "outputId": "fafd9aaa-eb48-4be5-aef2-c058d3229acc",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        }
      },
      "cell_type": "code",
      "source": [
        "!ls"
      ],
      "execution_count": 2,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "sample_data\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "8wqbULRBAwpm",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "위에서 보는 바와 같이 sample_data라는 디렉토리가 하위 디렉토리에 있다는 것을 알 수 있습니다."
      ]
    },
    {
      "metadata": {
        "id": "YilM6khOA4c-",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "다음으로 우리가 사용하고 있는 시스템의 프로세서 및 GPU에 대해서 알아보도록 하지요.\n",
        "우선 CPU에 대한 정보를 얻기 위해서 Proc file 시스템의 cpuinfo를 cat 명령어로 보도록 해보지요."
      ]
    },
    {
      "metadata": {
        "id": "23SOyndQ-Te1",
        "colab_type": "code",
        "outputId": "da406273-f24e-4498-b186-a18fa9b68562",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 1045
        }
      },
      "cell_type": "code",
      "source": [
        "!cat /proc/cpuinfo"
      ],
      "execution_count": 3,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "processor\t: 0\n",
            "vendor_id\t: GenuineIntel\n",
            "cpu family\t: 6\n",
            "model\t\t: 79\n",
            "model name\t: Intel(R) Xeon(R) CPU @ 2.20GHz\n",
            "stepping\t: 0\n",
            "microcode\t: 0x1\n",
            "cpu MHz\t\t: 2200.000\n",
            "cache size\t: 56320 KB\n",
            "physical id\t: 0\n",
            "siblings\t: 2\n",
            "core id\t\t: 0\n",
            "cpu cores\t: 1\n",
            "apicid\t\t: 0\n",
            "initial apicid\t: 0\n",
            "fpu\t\t: yes\n",
            "fpu_exception\t: yes\n",
            "cpuid level\t: 13\n",
            "wp\t\t: yes\n",
            "flags\t\t: fpu vme de pse tsc msr pae mce cx8 apic sep mtrr pge mca cmov pat pse36 clflush mmx fxsr sse sse2 ss ht syscall nx pdpe1gb rdtscp lm constant_tsc rep_good nopl xtopology nonstop_tsc cpuid tsc_known_freq pni pclmulqdq ssse3 fma cx16 pcid sse4_1 sse4_2 x2apic movbe popcnt aes xsave avx f16c rdrand hypervisor lahf_lm abm 3dnowprefetch invpcid_single pti ssbd ibrs ibpb stibp fsgsbase tsc_adjust bmi1 hle avx2 smep bmi2 erms invpcid rtm rdseed adx smap xsaveopt arat arch_capabilities\n",
            "bugs\t\t: cpu_meltdown spectre_v1 spectre_v2 spec_store_bypass l1tf\n",
            "bogomips\t: 4400.00\n",
            "clflush size\t: 64\n",
            "cache_alignment\t: 64\n",
            "address sizes\t: 46 bits physical, 48 bits virtual\n",
            "power management:\n",
            "\n",
            "processor\t: 1\n",
            "vendor_id\t: GenuineIntel\n",
            "cpu family\t: 6\n",
            "model\t\t: 79\n",
            "model name\t: Intel(R) Xeon(R) CPU @ 2.20GHz\n",
            "stepping\t: 0\n",
            "microcode\t: 0x1\n",
            "cpu MHz\t\t: 2200.000\n",
            "cache size\t: 56320 KB\n",
            "physical id\t: 0\n",
            "siblings\t: 2\n",
            "core id\t\t: 0\n",
            "cpu cores\t: 1\n",
            "apicid\t\t: 1\n",
            "initial apicid\t: 1\n",
            "fpu\t\t: yes\n",
            "fpu_exception\t: yes\n",
            "cpuid level\t: 13\n",
            "wp\t\t: yes\n",
            "flags\t\t: fpu vme de pse tsc msr pae mce cx8 apic sep mtrr pge mca cmov pat pse36 clflush mmx fxsr sse sse2 ss ht syscall nx pdpe1gb rdtscp lm constant_tsc rep_good nopl xtopology nonstop_tsc cpuid tsc_known_freq pni pclmulqdq ssse3 fma cx16 pcid sse4_1 sse4_2 x2apic movbe popcnt aes xsave avx f16c rdrand hypervisor lahf_lm abm 3dnowprefetch invpcid_single pti ssbd ibrs ibpb stibp fsgsbase tsc_adjust bmi1 hle avx2 smep bmi2 erms invpcid rtm rdseed adx smap xsaveopt arat arch_capabilities\n",
            "bugs\t\t: cpu_meltdown spectre_v1 spectre_v2 spec_store_bypass l1tf\n",
            "bogomips\t: 4400.00\n",
            "clflush size\t: 64\n",
            "cache_alignment\t: 64\n",
            "address sizes\t: 46 bits physical, 48 bits virtual\n",
            "power management:\n",
            "\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "kQLO9xU5r0FT",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "메모리 및 디스크 사용량을 알아볼까요 ?\n",
        "\n",
        "1.   메모리 사용량: !cat /proc/meminfo\n",
        "2.   디스크 사용량: !df -h\n",
        "\n"
      ]
    },
    {
      "metadata": {
        "id": "1uixdR9Wr4Ve",
        "colab_type": "code",
        "outputId": "f14b913d-af9d-44d4-c3d1-0665dcfd475f",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 167
        }
      },
      "cell_type": "code",
      "source": [
        "!df -h"
      ],
      "execution_count": 4,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "Filesystem      Size  Used Avail Use% Mounted on\n",
            "overlay         359G   20G  321G   6% /\n",
            "tmpfs           6.4G     0  6.4G   0% /dev\n",
            "tmpfs           6.4G     0  6.4G   0% /sys/fs/cgroup\n",
            "tmpfs           6.4G  8.0K  6.4G   1% /var/colab\n",
            "/dev/sda1       365G   24G  342G   7% /opt/bin\n",
            "shm             6.0G     0  6.0G   0% /dev/shm\n",
            "tmpfs           6.4G     0  6.4G   0% /sys/firmware\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "LPZ2wmn4BKPG",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "자 다음엔 GPU의 스펙을 살펴보도록 하지요. GPU에 대한 정보를 얻기 위해서는 nvidia-smi 명령어를 사용합니다."
      ]
    },
    {
      "metadata": {
        "id": "hlMXU3Gx-ffG",
        "colab_type": "code",
        "outputId": "70665697-9283-4654-9246-1196c64c1253",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 335
        }
      },
      "cell_type": "code",
      "source": [
        "!nvidia-smi"
      ],
      "execution_count": 5,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "Wed Jan 30 02:51:03 2019       \n",
            "+-----------------------------------------------------------------------------+\n",
            "| NVIDIA-SMI 396.44                 Driver Version: 396.44                    |\n",
            "|-------------------------------+----------------------+----------------------+\n",
            "| GPU  Name        Persistence-M| Bus-Id        Disp.A | Volatile Uncorr. ECC |\n",
            "| Fan  Temp  Perf  Pwr:Usage/Cap|         Memory-Usage | GPU-Util  Compute M. |\n",
            "|===============================+======================+======================|\n",
            "|   0  Tesla K80           Off  | 00000000:00:04.0 Off |                    0 |\n",
            "| N/A   46C    P8    30W / 149W |      0MiB / 11441MiB |      0%      Default |\n",
            "+-------------------------------+----------------------+----------------------+\n",
            "                                                                               \n",
            "+-----------------------------------------------------------------------------+\n",
            "| Processes:                                                       GPU Memory |\n",
            "|  GPU       PID   Type   Process name                             Usage      |\n",
            "|=============================================================================|\n",
            "|  No running processes found                                                 |\n",
            "+-----------------------------------------------------------------------------+\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "bm9xIN3xBTHO",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "CPU와 GPU에 대한 정보를 확인 한 후에는 우리가 수행하고자 하는 코드들을 github로 부터 가져와서 시행해보도록 하겠습니다.\n",
        "\n",
        "저같은 경우는 https://github.com/jeonggunlee/CUDATeaching 의 내용을 git 명령어를 통하여 clone 해오도록 하겠습니다.\n"
      ]
    },
    {
      "metadata": {
        "id": "GXFwkU-32_Qr",
        "colab_type": "code",
        "outputId": "9559134f-1642-441f-e7d5-3ee032e5abc0",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 148
        }
      },
      "cell_type": "code",
      "source": [
        "!git clone https://github.com/jeonggunlee/CUDATeaching"
      ],
      "execution_count": 6,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "Cloning into 'CUDATeaching'...\n",
            "remote: Enumerating objects: 53, done.\u001b[K\n",
            "remote: Counting objects:   1% (1/53)   \u001b[K\rremote: Counting objects:   3% (2/53)   \u001b[K\rremote: Counting objects:   5% (3/53)   \u001b[K\rremote: Counting objects:   7% (4/53)   \u001b[K\rremote: Counting objects:   9% (5/53)   \u001b[K\rremote: Counting objects:  11% (6/53)   \u001b[K\rremote: Counting objects:  13% (7/53)   \u001b[K\rremote: Counting objects:  15% (8/53)   \u001b[K\rremote: Counting objects:  16% (9/53)   \u001b[K\rremote: Counting objects:  18% (10/53)   \u001b[K\rremote: Counting objects:  20% (11/53)   \u001b[K\rremote: Counting objects:  22% (12/53)   \u001b[K\rremote: Counting objects:  24% (13/53)   \u001b[K\rremote: Counting objects:  26% (14/53)   \u001b[K\rremote: Counting objects:  28% (15/53)   \u001b[K\rremote: Counting objects:  30% (16/53)   \u001b[K\rremote: Counting objects:  32% (17/53)   \u001b[K\rremote: Counting objects:  33% (18/53)   \u001b[K\rremote: Counting objects:  35% (19/53)   \u001b[K\rremote: Counting objects:  37% (20/53)   \u001b[K\rremote: Counting objects:  39% (21/53)   \u001b[K\rremote: Counting objects:  41% (22/53)   \u001b[K\rremote: Counting objects:  43% (23/53)   \u001b[K\rremote: Counting objects:  45% (24/53)   \u001b[K\rremote: Counting objects:  47% (25/53)   \u001b[K\rremote: Counting objects:  49% (26/53)   \u001b[K\rremote: Counting objects:  50% (27/53)   \u001b[K\rremote: Counting objects:  52% (28/53)   \u001b[K\rremote: Counting objects:  54% (29/53)   \u001b[K\rremote: Counting objects:  56% (30/53)   \u001b[K\rremote: Counting objects:  58% (31/53)   \u001b[K\rremote: Counting objects:  60% (32/53)   \u001b[K\rremote: Counting objects:  62% (33/53)   \u001b[K\rremote: Counting objects:  64% (34/53)   \u001b[K\rremote: Counting objects:  66% (35/53)   \u001b[K\rremote: Counting objects:  67% (36/53)   \u001b[K\rremote: Counting objects:  69% (37/53)   \u001b[K\rremote: Counting objects:  71% (38/53)   \u001b[K\rremote: Counting objects:  73% (39/53)   \u001b[K\rremote: Counting objects:  75% (40/53)   \u001b[K\rremote: Counting objects:  77% (41/53)   \u001b[K\rremote: Counting objects:  79% (42/53)   \u001b[K\rremote: Counting objects:  81% (43/53)   \u001b[K\rremote: Counting objects:  83% (44/53)   \u001b[K\rremote: Counting objects:  84% (45/53)   \u001b[K\rremote: Counting objects:  86% (46/53)   \u001b[K\rremote: Counting objects:  88% (47/53)   \u001b[K\rremote: Counting objects:  90% (48/53)   \u001b[K\rremote: Counting objects:  92% (49/53)   \u001b[K\rremote: Counting objects:  94% (50/53)   \u001b[K\rremote: Counting objects:  96% (51/53)   \u001b[K\rremote: Counting objects:  98% (52/53)   \u001b[K\rremote: Counting objects: 100% (53/53)   \u001b[K\rremote: Counting objects: 100% (53/53), done.\u001b[K\n",
            "remote: Compressing objects: 100% (53/53), done.\u001b[K\n",
            "remote: Total 192 (delta 30), reused 0 (delta 0), pack-reused 139\u001b[K\n",
            "Receiving objects: 100% (192/192), 21.18 MiB | 31.39 MiB/s, done.\n",
            "Resolving deltas: 100% (80/80), done.\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "Z3IR5nKLBoCn",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "Git clone이 완료된 후에 올바로 clone이 되었는지 확인하기 위하여 ls 명령어를 수행해보독 하겠습니다."
      ]
    },
    {
      "metadata": {
        "id": "mPhAWjWE-2Hd",
        "colab_type": "code",
        "outputId": "838766c6-0d65-43f8-9bce-cc12a18e15db",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        }
      },
      "cell_type": "code",
      "source": [
        "!ls"
      ],
      "execution_count": 7,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "CUDATeaching  sample_data\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "qegzQU75BvPV",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "이후, cd 명령어를 이용하여 CUDATeaching 디렉토리로 들어가보도록 하겠습니다. 주의 할점은 시스템 명령어를 시행할때 ! 또는 %를 붙여주어야 한다는 것입니다."
      ]
    },
    {
      "metadata": {
        "id": "pbnP0p_j-5k3",
        "colab_type": "code",
        "outputId": "7f118bdd-5285-4e95-bd8f-8eb66c47c470",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        }
      },
      "cell_type": "code",
      "source": [
        "%cd CUDATeaching"
      ],
      "execution_count": 8,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "/content/CUDATeaching\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "gvgc2IbkB7-Z",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "다시 CUDATeaching 디렉토리로 들어온 후에 ls 명령어를 통하여 디렉토리에 어떤 화일이 있는지 살펴보도록 하겠습니다."
      ]
    },
    {
      "metadata": {
        "id": "xFj7PCl0_Btq",
        "colab_type": "code",
        "outputId": "66f624bf-8d1c-428c-8df6-8b15101afa69",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 55
        }
      },
      "cell_type": "code",
      "source": [
        "!ls"
      ],
      "execution_count": 9,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "'00_병렬 컴퓨팅 소개.pdf'   colab_gdrive.ipynb\t hello_CUDA.ipynb\n",
            " 01_cuda_lab\t\t    gpu4cloud.md\t README.md\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "u-YhEY8mq_v2",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "위 화일 및 디렉토리 중,01_cuda_lab에 들어가 보도록 하겠습니다. 01_cuda_lab에는 실습을 위한 cuda 소스 코드들이 있습니다."
      ]
    },
    {
      "metadata": {
        "id": "TQjjUR9-_F7c",
        "colab_type": "code",
        "outputId": "3f8e48b5-d15a-4cd2-d851-8c3e97744568",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        }
      },
      "cell_type": "code",
      "source": [
        "%cd 01_cuda_lab"
      ],
      "execution_count": 11,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "/content/CUDATeaching/01_cuda_lab\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "8CUELaLGCXdu",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "01_cuda_lab에 들어온후에 ls 명령어를 통하여 어떤 화일들이 있는지 확인해보록 하겠습니다."
      ]
    },
    {
      "metadata": {
        "id": "JJ0EtSxX_J2l",
        "colab_type": "code",
        "outputId": "f9609019-1863-4474-c8f8-06a8f977faf7",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 148
        }
      },
      "cell_type": "code",
      "source": [
        "!ls"
      ],
      "execution_count": 12,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "00_reduction_all.cu\t     06_matmul_tile3_noBankConflict.cu\tbankConflict.cu\n",
            "00_reduction_final.cu\t     07_matmul_tile4_unroll.cu\t\tclock.cu\n",
            "01_reduction.cu\t\t     1hello.cu\t\t\t\tCUDA_lab.txt\n",
            "02_reduction.cu\t\t     2Dindex.cu\t\t\t\tglobalMemory.cu\n",
            "03_matmul.cu\t\t     2index.cu\t\t\t\tREADME.md\n",
            "04_matmul_tile.cu\t     async_streams.cu\t\t\ttranspose.cu\n",
            "05_matmul_tile2_mem_coel.cu  atomicAdd.cu\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "2S5tXs76CfYd",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "가장 기본 적인 코드로 1hello.cu 를 컴파일 하고 실행시켜보도록 하겠습니다.\n",
        "1hello.cu의 내용은 아래와 같습니다.\n",
        "\n",
        "```cuda\n",
        "#include <stdio.h>\n",
        "\n",
        "__global__ void helloCUDA(void)\n",
        "{\n",
        "  printf(\"Hello thread %d in block %d\\n\", threadIdx.x, blockIdx.x);\n",
        "}\n",
        "\n",
        "int main()\n",
        "{\n",
        "  helloCUDA<<<3, 4>>>();\n",
        "  cudaDeviceReset();\n",
        "  return 0;\n",
        "}\n",
        "```\n"
      ]
    },
    {
      "metadata": {
        "id": "HDA89RC4DIZu",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "1hello.cu를 컴파일하기 위하여 nvdia GPU 컴파일러 명령어인 nvcc를 사용합니다.\n",
        "nvcc 컴파일러가 어디에 위치하는지 먼저 살펴볼까요 ?"
      ]
    },
    {
      "metadata": {
        "id": "4gXSXck4on4X",
        "colab_type": "code",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        },
        "outputId": "90c61b4a-c45a-47bf-b566-1bb45885619e"
      },
      "cell_type": "code",
      "source": [
        "!which nvcc"
      ],
      "execution_count": 13,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "/usr/local/cuda/bin/nvcc\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "PmW53_jZovgS",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "위의 명령어를 통하여 nvcc 컴파일러가 /usr/local/cuda/bin/에 위치하고 있음을 알 수 있습니다.\n",
        "다음은 nvidia GPU 카드가 설치 되었는지 확인해보는 작업을 진행해보겠습니다.\n"
      ]
    },
    {
      "metadata": {
        "id": "tpBm7vw9pNTy",
        "colab_type": "code",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 92
        },
        "outputId": "cc207fcf-bf2b-46e1-f80d-52c8eb53fda0"
      },
      "cell_type": "code",
      "source": [
        "!ls -l /dev/nv*"
      ],
      "execution_count": 14,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "crw-rw-rw- 1 root root 195,   0 Jan 30 02:41 /dev/nvidia0\n",
            "crw-rw-rw- 1 root root 195, 255 Jan 30 02:41 /dev/nvidiactl\n",
            "crw-rw-rw- 1 root root 248,   0 Jan 30 02:41 /dev/nvidia-uvm\n",
            "crw-rw-rw- 1 root root 248,   1 Jan 30 02:41 /dev/nvidia-uvm-tools\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "A9_xbMo7pTXO",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "위의 nvidia0 및 nvidiactrl 등을 통하여 nvidia GPU 장치가 설치되어 있음을 알 수 있습니다.\n",
        "\n",
        "자 그럼 장치와 컴파일러가 잘 설치되어 있음을 확인하였으니, nvcc 명령어를 이용하여 CUDA 코드를 컴파일을 해보도록 하겠습니다.\n",
        "\n",
        "명령어의 가장 단순한 컴파일 형식은 다음과 같습니다.\n",
        "   * nvcc cuda_code.cu -o executable_output\n",
        "   * 실예] nvcc ./1hello.cu -o 1hello\n",
        "   "
      ]
    },
    {
      "metadata": {
        "id": "M8uy4sE0_NCt",
        "colab_type": "code",
        "colab": {}
      },
      "cell_type": "code",
      "source": [
        "!nvcc ./1hello.cu"
      ],
      "execution_count": 0,
      "outputs": []
    },
    {
      "metadata": {
        "id": "ZqBlBr27DPH_",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "이후, 컴파일된 실행화일인 a.out을 실행시키면 해당 결과가 프린트 됩니다."
      ]
    },
    {
      "metadata": {
        "id": "QgT2HwkN_Zvd",
        "colab_type": "code",
        "outputId": "6a5f87df-a134-4a4a-a8c6-8c60842d0697",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 241
        }
      },
      "cell_type": "code",
      "source": [
        "!./a.out"
      ],
      "execution_count": 16,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "Hello thread 0 in block 2\n",
            "Hello thread 1 in block 2\n",
            "Hello thread 2 in block 2\n",
            "Hello thread 3 in block 2\n",
            "Hello thread 0 in block 1\n",
            "Hello thread 1 in block 1\n",
            "Hello thread 2 in block 1\n",
            "Hello thread 3 in block 1\n",
            "Hello thread 0 in block 0\n",
            "Hello thread 1 in block 0\n",
            "Hello thread 2 in block 0\n",
            "Hello thread 3 in block 0\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "IiUlkF5gpYEP",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "위의 코드를 수정하고 싶다면 어떻게 하면 좋을까요 ?\n",
        "\" %%writefile hello_cuda.cu \"를 이용해서 진행하면 좋을 것 같습니다.\n"
      ]
    },
    {
      "metadata": {
        "id": "kdLQnzbsppGF",
        "colab_type": "code",
        "outputId": "90da7af7-328a-42fe-e86d-fda6be04e1f0",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 36
        }
      },
      "cell_type": "code",
      "source": [
        "%%writefile hello_cuda.cu\n",
        "# include <stdio.h>\n",
        "\n",
        "__global__ void helloCUDA(void)\n",
        "{\n",
        "  printf(\"Hello thread %d in block %d\\n\", threadIdx.x, blockIdx.x);\n",
        "}\n",
        "\n",
        "int main()\n",
        "{\n",
        "  helloCUDA<<<2, 3>>>();\n",
        "  cudaDeviceReset();\n",
        "  return 0;\n",
        "}\n"
      ],
      "execution_count": 17,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "Writing hello_cuda.cu\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "vQnWq_MnqGiy",
        "colab_type": "code",
        "colab": {}
      },
      "cell_type": "code",
      "source": [
        "!nvcc ./hello_cuda.cu -o hello_cuda"
      ],
      "execution_count": 0,
      "outputs": []
    },
    {
      "metadata": {
        "id": "oP3eh9amqYt3",
        "colab_type": "code",
        "outputId": "0cbdd38b-e6a3-45b1-baf6-a64d29b6fbf9",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 167
        }
      },
      "cell_type": "code",
      "source": [
        "!ls"
      ],
      "execution_count": 19,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "00_reduction_all.cu\t\t   07_matmul_tile4_unroll.cu  clock.cu\n",
            "00_reduction_final.cu\t\t   1hello.cu\t\t      CUDA_lab.txt\n",
            "01_reduction.cu\t\t\t   2Dindex.cu\t\t      globalMemory.cu\n",
            "02_reduction.cu\t\t\t   2index.cu\t\t      hello_cuda\n",
            "03_matmul.cu\t\t\t   a.out\t\t      hello_cuda.cu\n",
            "04_matmul_tile.cu\t\t   async_streams.cu\t      README.md\n",
            "05_matmul_tile2_mem_coel.cu\t   atomicAdd.cu\t\t      transpose.cu\n",
            "06_matmul_tile3_noBankConflict.cu  bankConflict.cu\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "mSaBx4qHqeyw",
        "colab_type": "code",
        "outputId": "9bbb4fad-f092-45a7-da71-06da14848446",
        "colab": {
          "base_uri": "https://localhost:8080/",
          "height": 241
        }
      },
      "cell_type": "code",
      "source": [
        "!./a.out"
      ],
      "execution_count": 20,
      "outputs": [
        {
          "output_type": "stream",
          "text": [
            "Hello thread 0 in block 2\n",
            "Hello thread 1 in block 2\n",
            "Hello thread 2 in block 2\n",
            "Hello thread 3 in block 2\n",
            "Hello thread 0 in block 1\n",
            "Hello thread 1 in block 1\n",
            "Hello thread 2 in block 1\n",
            "Hello thread 3 in block 1\n",
            "Hello thread 0 in block 0\n",
            "Hello thread 1 in block 0\n",
            "Hello thread 2 in block 0\n",
            "Hello thread 3 in block 0\n"
          ],
          "name": "stdout"
        }
      ]
    },
    {
      "metadata": {
        "id": "WJZLln-6DVf2",
        "colab_type": "text"
      },
      "cell_type": "markdown",
      "source": [
        "오늘 colaboratory를 활용하여 간단히 GPU CUDA Programming을 하는 방법을 보여드렸습니다. 잘 활용하시면 좋을 것 같네요!\n",
        "감사합니다.\n"
      ]
    }
  ]
}