{"id":15485,"date":"2024-12-14T17:37:57","date_gmt":"2024-12-14T08:37:57","guid":{"rendered":"http:\/\/www.gisdeveloper.co.kr\/?p=15485"},"modified":"2024-12-14T20:54:35","modified_gmt":"2024-12-14T11:54:35","slug":"webgpu%eb%a5%bc-%ec%9d%b4%ec%9a%a9%ed%95%9c-%ea%b3%84%ec%82%b0computing","status":"publish","type":"post","link":"http:\/\/www.gisdeveloper.co.kr\/?p=15485","title":{"rendered":"WebGPU\ub97c \uc774\uc6a9\ud55c \uacc4\uc0b0(Computing)"},"content":{"rendered":"

3\uac1c\uc758 \uc2e4\uc218\uac12\uc744 \uc804\ub2ec\ud558\uace0 \uc774 \uac12\uc5d0 2\ub97c \uacf1\ud55c \uacb0\uacfc\ub97c GPU\ub97c \ud1b5\ud574 \ubcd1\ub82c\ub85c \ucc98\ub9ac \uc2e4\ud589\ud574 \uadf8 \uacb0\uacfc\ub97c \uc5bb\ub294 \ucf54\ub4dc\uc785\ub2c8\ub2e4. \uba3c\uc800 GPU\uc5d0 \ub300\ud55c \uac1d\uccb4\ub97c \uc5bb\ub294 \uac83\uc740 \ube44\ub3d9\uae30\ub85c \ucc98\ub9ac\ud574\uc57c \ud558\ubbc0\ub85c \ub2e4\uc74c\uacfc \uac19\uc740 \ucf54\ub4dc\uac00 \ud544\uc694\ud569\ub2c8\ub2e4.<\/p>\n

\r\nasync function main() {\r\n  \/\/ \uc5ec\uae30\uc5d0 \uc774\ud6c4 \ubaa8\ub4e0 \ucf54\ub4dc\uac00 \uc791\uc131\ub429\ub2c8\ub2e4.\r\n}\r\n\r\nawait main();\r\n<\/pre>\n
\uc790, \uc774\uc81c GPU\uc5d0 \ub300\ud55c \uac1d\uccb4\ub97c \uc5bb\uc2b5\ub2c8\ub2e4.<\/p>\n
\r\n  const adapter = await navigator.gpu.requestAdapter();\r\n  const device = await adapter.requestDevice();\r\n<\/pre>\n
\uc2e4\ud589\ud560 \uc250\uc774\ub354 \ucf54\ub4dc\ub97c \uc791\uc131\ud569\ub2c8\ub2e4. \uc250\uc774\ub354 \ucf54\ub4dc\ub294 WGSL\ub85c \uc791\uc131\ub429\ub2c8\ub2e4.<\/p>\n
\r\n  const module = device.createShaderModule({\r\n    label: 'doubling compute module',\r\n    code: \/* wgsl *\/ `\r\n      @group(0) @binding(0) var<storage, read_write> data: array<f32>;\r\n \r\n      @compute @workgroup_size(1) fn computeSomething(\r\n        @builtin(global_invocation_id) id: vec3u\r\n      ) {\r\n        let i = id.x;\r\n        data[i] = data[i] * 2.0;\r\n      }\r\n    `,\r\n  });\r\n<\/pre>\n
\uc704\uc758 \uc250\uc774\ub354 \ucf54\ub4dc\ub97c \uc2e4\ud589\ud560 \ud30c\uc774\ud504\ub77c\uc778 \uac1d\uccb4\ub97c \uc0dd\uc131\ud569\ub2c8\ub2e4.<\/p>\n
\r\n  const pipeline = device.createComputePipeline({\r\n    label: 'x2 compute pipeline',\r\n    layout: 'auto',\r\n    compute: {\r\n      module,\r\n      entryPoint: 'computeSomething', \/\/ \uc2e4\ud589\ud560 \uc250\uc774\ub354 \ud568\uc218\r\n    },\r\n  });\r\n<\/pre>\n
\uc2e4\ud589\ud560 \uc250\uc774\ub354 \ud568\uc218\ub97c \ubcf4\uba74 1\uac1c\uc758 \uc778\uc790\ub97c \ubc1b\ub294\ub370, \uadf8 \uc778\uc790\ub97c \uc815\uc758\ud569\ub2c8\ub2e4.<\/p>\n
\r\n  const input = new Float32Array([1, 3, 5]);\r\n<\/pre>\n
\uc704\uc758 \uc778\uc790\ub294 CPU \uba54\ubaa8\ub9ac\uc5d0 \uc788\uc73c\ubbc0\ub85c \uc774\ub97c GPU \uba54\ubaa8\ub9ac\uc5d0 \ub9cc\ub4e4\uc5b4\uc11c \ubcf5\uc0ac\ud574 \uc918\uc57c \ud569\ub2c8\ub2e4.<\/p>\n
\r\n  const workBuffer = device.createBuffer({\r\n    label: 'work buffer',\r\n    size: input.byteLength,\r\n    usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_SRC | GPUBufferUsage.COPY_DST,\r\n  });\r\n\r\n  device.queue.writeBuffer(workBuffer, 0, input);\r\n<\/pre>\n
GPU\ub97c \ud1b5\ud574 \uacc4\uc0b0\ub41c \ucd5c\uc885 \uacb0\uacfc\ub97c \uc77d\uae30 \uc704\ud55c \uc0ac\ubcf8 \ubc84\ud37c\ub97c \uc0dd\uc131\ud569\ub2c8\ub2e4.<\/p>\n
\r\n  const resultBuffer = device.createBuffer({\r\n    label: 'result buffer',\r\n    size: input.byteLength,\r\n    usage: GPUBufferUsage.MAP_READ | GPUBufferUsage.COPY_DST\r\n  });\r\n<\/pre>\n
\uacc4\uc0b0\uc744 \uc704\ud574 \uc0ac\uc6a9\ub420 \uc778\uc790\uc5d0 \ub300\ud55c \ubc84\ud37c\uac00 \ubb34\uc5c7\uc778\uc9c0\ub97c \uba85\ud655\ud788 \uc9c0\uc815\ud569\ub2c8\ub2e4.<\/p>\n
\r\n  const bindGroup = device.createBindGroup({\r\n    label: 'bindGroup for work buffer',\r\n    layout: pipeline.getBindGroupLayout(0),\r\n    entries: [\r\n      { binding: 0, resource: { buffer: workBuffer } },\r\n    ],\r\n  });\r\n<\/pre>\n
GPU\uc5d0\uc11c \uc2e4\ud589\ud560 \uba85\ub839\uc744 \uc778\ucf54\ub529\ud569\ub2c8\ub2e4.<\/p>\n
\r\n  const encoder = device.createCommandEncoder({ label: 'x2 encoder' });\r\n  const pass = encoder.beginComputePass({ label: 'x2 compute pass' });\r\n\r\n  pass.setPipeline(pipeline);\r\n  pass.setBindGroup(0, bindGroup);\r\n  pass.dispatchWorkgroups(input.length);\r\n  pass.end();\r\n<\/pre>\n
\uacc4\uc0b0 \uacb0\uacfc\ub97c \uc77d\uc5b4 \ub0b4\uae30 \uc704\ud574 \ub9f5\ud551\uc6a9 \ubc84\ud37c\uc5d0 \ubcf5\uc0ac\ud558\ub294 \uba85\ub839\uc744 \uc778\ucf54\ub529\ud569\ub2c8\ub2e4.<\/p>\n
\r\n  encoder.copyBufferToBuffer(workBuffer, 0, resultBuffer, 0, resultBuffer.size);\r\n<\/pre>\n
\uc2e4\ud574 \uba85\ub839\uc758 \uc2e4\ud589\uc740 \ub2e4\uc74c \ucf54\ub4dc\ub97c \ud1b5\ud574 \uc774\ub8e8\uc5b4\uc9d1\ub2c8\ub2e4.<\/p>\n
\r\n   const commandBuffer = encoder.finish();\r\n   device.queue.submit([commandBuffer]);\r\n<\/pre>\n
\uc2e4\ud589 \uacb0\uacfc\ub97c \ucf58\uc194\uc5d0 \ucd9c\ub825\ud558\uae30 \uc704\ud55c \ucf54\ub4dc\uc785\ub2c8\ub2e4.<\/p>\n
\r\n  await resultBuffer.mapAsync(GPUMapMode.READ);\r\n  const result = new Float32Array(resultBuffer.getMappedRange());\r\n \r\n  console.log('input', input);\r\n  console.log('result', result);\r\n \r\n  resultBuffer.unmap();\r\n<\/pre>\n
\uc2e4\ud589 \uacb0\uacfc\ub97c \ubcf4\uba74 \ub2e4\uc74c\uacfc \uac19\uc2b5\ub2c8\ub2e4.<\/p>\n
\"\"<\/p>\n
\uc2e4\ud589 \uacb0\uacfc\ub97c \ubcf4\uba74 \uc785\ub825\uac12\uc5d0 2\ubc30\ub9cc\ud07c \uacf1\ud574\uc84c\ub294\ub370, \uc774\ub294 GPU\ub97c \ud1b5\ud574 \ubcd1\ub82c\ub85c \uacc4\uc0b0\ub41c \uac83\uc785\ub2c8\ub2e4.<\/p>\n","protected":false},"excerpt":{"rendered":"
3\uac1c\uc758 \uc2e4\uc218\uac12\uc744 \uc804\ub2ec\ud558\uace0 \uc774 \uac12\uc5d0 2\ub97c \uacf1\ud55c \uacb0\uacfc\ub97c GPU\ub97c \ud1b5\ud574 \ubcd1\ub82c\ub85c \ucc98\ub9ac \uc2e4\ud589\ud574 \uadf8 \uacb0\uacfc\ub97c \uc5bb\ub294 \ucf54\ub4dc\uc785\ub2c8\ub2e4. \uba3c\uc800 GPU\uc5d0 \ub300\ud55c \uac1d\uccb4\ub97c \uc5bb\ub294 \uac83\uc740 \ube44\ub3d9\uae30\ub85c \ucc98\ub9ac\ud574\uc57c \ud558\ubbc0\ub85c \ub2e4\uc74c\uacfc \uac19\uc740 \ucf54\ub4dc\uac00 \ud544\uc694\ud569\ub2c8\ub2e4. async function main() { \/\/ \uc5ec\uae30\uc5d0 \uc774\ud6c4 \ubaa8\ub4e0 \ucf54\ub4dc\uac00 \uc791\uc131\ub429\ub2c8\ub2e4. } await main(); \uc790, \uc774\uc81c GPU\uc5d0 \ub300\ud55c \uac1d\uccb4\ub97c \uc5bb\uc2b5\ub2c8\ub2e4. const adapter = await navigator.gpu.requestAdapter(); const … <\/p>\n
\ub354 \ubcf4\uae30 “WebGPU\ub97c \uc774\uc6a9\ud55c \uacc4\uc0b0(Computing)”<\/span><\/a><\/p>\n","protected":false},"author":1,"featured_media":0,"comment_status":"open","ping_status":"open","sticky":false,"template":"","format":"standard","meta":{"footnotes":""},"categories":[139],"tags":[],"class_list":["post-15485","post","type-post","status-publish","format-standard","hentry","category-webgl"],"_links":{"self":[{"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=\/wp\/v2\/posts\/15485","targetHints":{"allow":["GET"]}}],"collection":[{"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=\/wp\/v2\/posts"}],"about":[{"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=\/wp\/v2\/types\/post"}],"author":[{"embeddable":true,"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=\/wp\/v2\/users\/1"}],"replies":[{"embeddable":true,"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=%2Fwp%2Fv2%2Fcomments&post=15485"}],"version-history":[{"count":4,"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=\/wp\/v2\/posts\/15485\/revisions"}],"predecessor-version":[{"id":15487,"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=\/wp\/v2\/posts\/15485\/revisions\/15487"}],"wp:attachment":[{"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=%2Fwp%2Fv2%2Fmedia&parent=15485"}],"wp:term":[{"taxonomy":"category","embeddable":true,"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=%2Fwp%2Fv2%2Fcategories&post=15485"},{"taxonomy":"post_tag","embeddable":true,"href":"http:\/\/www.gisdeveloper.co.kr\/index.php?rest_route=%2Fwp%2Fv2%2Ftags&post=15485"}],"curies":[{"name":"wp","href":"https:\/\/api.w.org\/{rel}","templated":true}]}}