From 344b507e4fb7220a24edf5f78efc1677162ac3e8 Mon Sep 17 00:00:00 2001 From: foreman Date: Wed, 4 Apr 2018 18:00:17 -0400 Subject: [PATCH] P4 to Git Change 1536925 by vsytchen@vsytchen-ocl-win10 on 2018/04/04 17:20:38 SWDEV-79445 - OCL generic changes and code clean-up 1. This change replaces the use of std::map with std::unordered_map to improve lookup/insert time. 2. Replace the use of std::make_pair and std::pair constructor with uniform initialization for cleaner code. 3. Replace the use of std::Container::iterator type with the auto keyword for cleaner code. 4. Use range based for loops where needed. ReviewBoardURL = http://ocltc.amd.com/reviews/r/14517/diff/ Affected files ... ... //depot/stg/opencl/drivers/opencl/api/hip/hip_platform.cpp#4 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_context.cpp#58 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_d3d10.cpp#16 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_d3d10_amd.hpp#9 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_d3d11.cpp#24 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_d3d11_amd.hpp#13 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_d3d9.cpp#34 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_d3d9_amd.hpp#17 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_gl.cpp#57 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_pipe.cpp#7 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_program.cpp#46 edit ... //depot/stg/opencl/drivers/opencl/api/opencl/amdocl/cl_svm.cpp#23 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/appprofile.hpp#14 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/cpu/cpuprogram.cpp#72 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/cpu/cpuvirtual.cpp#27 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/device.cpp#216 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/device.hpp#297 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuappprofile.cpp#13 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpubinary.cpp#59 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpucompiler.cpp#158 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpudevice.cpp#587 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpukernel.cpp#322 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuprintf.cpp#46 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuprogram.cpp#237 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuprogram.hpp#70 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuresource.cpp#242 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuvirtual.cpp#415 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/gpu/gpuvirtual.hpp#143 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palappprofile.cpp#3 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palcompiler.cpp#22 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/paldevice.cpp#79 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palprintf.cpp#9 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palprogram.cpp#59 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palresource.cpp#60 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palvirtual.cpp#84 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/pal/palvirtual.hpp#46 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/CMakeLists.txt#11 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/pro/prodevice.cpp#4 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/pro/prodevice.hpp#5 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocbinary.hpp#6 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/roccompiler.cpp#42 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/roccounters.cpp#3 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocprintf.cpp#10 edit ... //depot/stg/opencl/drivers/opencl/runtime/device/rocm/rocprogram.cpp#81 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/command.cpp#81 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/command.hpp#89 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/commandqueue.cpp#24 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/context.cpp#49 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/context.hpp#29 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/memory.cpp#129 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/memory.hpp#102 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/perfctr.hpp#7 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/program.cpp#91 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/program.hpp#43 edit ... //depot/stg/opencl/drivers/opencl/runtime/platform/sampler.hpp#9 edit ... //depot/stg/opencl/drivers/opencl/runtime/utils/flags.cpp#17 edit [ROCm/clr commit: 1aade00b0992a0384b4d62174a350679e53ec698] --- .../opencl/api/opencl/amdocl/cl_context.cpp | 5 +-- .../clr/opencl/api/opencl/amdocl/cl_d3d10.cpp | 14 +++---- .../opencl/api/opencl/amdocl/cl_d3d10_amd.hpp | 3 +- .../clr/opencl/api/opencl/amdocl/cl_d3d11.cpp | 16 ++++---- .../opencl/api/opencl/amdocl/cl_d3d11_amd.hpp | 3 +- .../clr/opencl/api/opencl/amdocl/cl_d3d9.cpp | 25 +++++------- .../opencl/api/opencl/amdocl/cl_d3d9_amd.hpp | 3 +- .../clr/opencl/api/opencl/amdocl/cl_gl.cpp | 24 +++++------ .../clr/opencl/api/opencl/amdocl/cl_pipe.cpp | 5 +-- .../opencl/api/opencl/amdocl/cl_program.cpp | 40 ++++++++----------- .../clr/opencl/api/opencl/amdocl/cl_svm.cpp | 20 +++++----- 11 files changed, 66 insertions(+), 92 deletions(-) diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_context.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_context.cpp index 8a1cfa96d8..007ac7d4d2 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_context.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_context.cpp @@ -296,9 +296,8 @@ RUNTIME_ENTRY(cl_int, clGetContextInfo, *not_null(param_value_size_ret) = valueSize; if (param_value != NULL) { cl_device_id* device_list = (cl_device_id*)param_value; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - *device_list++ = const_cast(as_cl(*it)); + for (const auto& it : devices) { + *device_list++ = const_cast(as_cl(it)); } } return CL_SUCCESS; diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_d3d10.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_d3d10.cpp index 0003cf2419..074f5cd736 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_d3d10.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_d3d10.cpp @@ -125,7 +125,7 @@ RUNTIME_ENTRY(cl_int, clGetDeviceIDsFromD3D10KHR, break; } - std::vector::iterator it = compatible_devices.begin(); + auto it = compatible_devices.cbegin(); cl_uint compatible_count = std::min(num_entries, (cl_uint)compatible_devices.size()); while (compatible_count--) { @@ -298,9 +298,8 @@ RUNTIME_ENTRY_RET(cl_mem, clCreateImageFromD3D10Resource, const std::vector& devices = as_amd(context)->devices(); bool supportPass = false; bool sizePass = false; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().imageSupport_) { + for (const auto& it : devices) { + if (it->info().imageSupport_) { supportPass = true; } } @@ -1104,9 +1103,8 @@ int D3D10Object::initD3D10Object(const Context& amdContext, ID3D10Resource* pRes ScopedLock sl(resLock_); // Check if this ressource has already been used for interop - std::vector>::iterator it; - for (it = resources_.begin(); it != resources_.end(); ++it) { - if ((*it).first == (void*)pRes && (*it).second == subres) { + for (const auto& it : resources_) { + if (it.first == (void*)pRes && it.second == subres) { return CL_INVALID_D3D10_RESOURCE_KHR; } } @@ -1349,7 +1347,7 @@ int D3D10Object::initD3D10Object(const Context& amdContext, ID3D10Resource* pRes return CL_INVALID_IMAGE_FORMAT_DESCRIPTOR; } } - resources_.push_back(std::make_pair(pRes, subres)); + resources_.push_back({pRes, subres}); return CL_SUCCESS; } diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_d3d10_amd.hpp b/projects/clr/opencl/api/opencl/amdocl/cl_d3d10_amd.hpp index fdacf1a5d3..79f1de7d66 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_d3d10_amd.hpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_d3d10_amd.hpp @@ -163,9 +163,8 @@ public: if(pD3D10ResOrig_) pD3D10ResOrig_->Release(); if(pQuery_) pQuery_->Release(); // Check if this resource has already been used for interop - std::vector>::iterator it; if(resources_.size()) { - for(it = resources_.begin(); it != resources_.end(); ++it) { + for(auto& it = resources_.cbegin(); it != resources_.cend(); it++) { if(((pD3D10ResOrig_ && (*it).first == (void*) pD3D10ResOrig_) || ((*it).first == (void*) pD3D10Res_)) && (*it).second == subRes_) { diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_d3d11.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_d3d11.cpp index eed8ebac51..65b0c2224d 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_d3d11.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_d3d11.cpp @@ -127,7 +127,7 @@ RUNTIME_ENTRY(cl_int, clGetDeviceIDsFromD3D11KHR, break; } - std::vector::iterator it = compatible_devices.begin(); + auto it = compatible_devices.cbegin(); cl_uint compatible_count = std::min(num_entries, (cl_uint)compatible_devices.size()); while (compatible_count--) { @@ -300,9 +300,8 @@ RUNTIME_ENTRY_RET(cl_mem, clCreateImageFromD3D11Resource, const std::vector& devices = as_amd(context)->devices(); bool supportPass = false; bool sizePass = false; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().imageSupport_) { + for (const auto& it : devices) { + if (it->info().imageSupport_) { supportPass = true; } } @@ -613,10 +612,9 @@ int D3D11Object::initD3D11Object(const Context& amdContext, ID3D11Resource* pRes ScopedLock sl(resLock_); // Check if this ressource has already been used for interop - std::vector>>::iterator it; - for (it = resources_.begin(); it != resources_.end(); ++it) { - if ((*it).first == (void*)pRes && (*it).second.first == subres && - (*it).second.second == plane) { + for (const auto& it : resources_) { + if (it.first == (void*)pRes && it.second.first == subres && + it.second.second == plane) { return CL_INVALID_D3D11_RESOURCE_KHR; } } @@ -889,7 +887,7 @@ int D3D11Object::initD3D11Object(const Context& amdContext, ID3D11Resource* pRes return CL_INVALID_IMAGE_FORMAT_DESCRIPTOR; } } - resources_.push_back(std::make_pair(pRes, std::make_pair(subres, plane))); + resources_.push_back({pRes, {subres, plane}}); return CL_SUCCESS; } diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_d3d11_amd.hpp b/projects/clr/opencl/api/opencl/amdocl/cl_d3d11_amd.hpp index 2cccd12c96..f2ba1ce726 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_d3d11_amd.hpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_d3d11_amd.hpp @@ -171,9 +171,8 @@ public: if(pD3D11ResOrig_) pD3D11ResOrig_->Release(); if(pQuery_) pQuery_->Release(); // Check if this resource has already been used for interop - std::vector>>::iterator it; if(resources_.size()) { - for(it = resources_.begin(); it != resources_.end(); ++it) { + for(auto& it = resources_.cbegin(); it != resources_.cend(); it++) { if(((pD3D11ResOrig_ && (*it).first == (void*) pD3D11ResOrig_) || ((*it).first == (void*) pD3D11Res_)) && (*it).second.first == subRes_ diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_d3d9.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_d3d9.cpp index 06b33f909e..d63658a809 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_d3d9.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_d3d9.cpp @@ -101,7 +101,7 @@ RUNTIME_ENTRY(cl_int, clGetDeviceIDsFromDX9MediaAdapterKHR, break; } - std::vector::iterator it = compatible_devices.begin(); + auto it = compatible_devices.cbegin(); cl_uint compatible_count = std::min(num_entries, (cl_uint)compatible_devices.size()); while (compatible_count--) { @@ -183,9 +183,8 @@ RUNTIME_ENTRY_RET(cl_mem, clCreateFromDX9MediaSurfaceKHR, const std::vector& devices = as_amd(context)->devices(); bool supportPass = false; bool sizePass = false; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().imageSupport_) { + for (const auto& it : devices) { + if (it->info().imageSupport_) { supportPass = true; } } @@ -484,9 +483,8 @@ int D3D9Object::initD3D9Object(const Context& amdContext, return CL_INVALID_DX9_MEDIA_ADAPTER_KHR; // Not supported yet } - std::vector>::iterator it; - for (it = resources_.begin(); it != resources_.end(); ++it) { - if ((*it).first.surfInfo.resource == cl_surf_info->resource && (*it).first.surfPlane == plane) { + for (const auto& it : resources_) { + if (it.first.surfInfo.resource == cl_surf_info->resource && it.first.surfPlane == plane) { return CL_INVALID_D3D9_RESOURCE_KHR; } } @@ -519,12 +517,11 @@ int D3D9Object::initD3D9Object(const Context& amdContext, // first check if the format is NV12 or YV12, which we need special handling if (NULL == shared_handle) { bool found = false; - std::vector>::iterator it; - for (it = resources_.begin(); it != resources_.end(); ++it) { - if ((*it).first.surfInfo.resource == cl_surf_info->resource && - (*it).first.surfPlane != plane) { - obj.handleShared_ = (*it).second.surfInfo.shared_handle; - obj.pD3D9Res_ = (*it).second.surfInfo.resource; + for (const auto& it : resources_) { + if (it.first.surfInfo.resource == cl_surf_info->resource && + it.first.surfPlane != plane) { + obj.handleShared_ = it.second.surfInfo.shared_handle; + obj.pD3D9Res_ = it.second.surfInfo.resource; obj.pD3D9Res_->AddRef(); obj.objDesc_ = obj.objDescOrig_; found = true; @@ -567,7 +564,7 @@ int D3D9Object::initD3D9Object(const Context& amdContext, TD3D9RESINFO d3d9ObjShared = {{obj.pD3D9Res_, obj.handleShared_}, plane}; if (errcode == CL_SUCCESS) { - resources_.push_back(std::make_pair(d3d9ObjOri, d3d9ObjShared)); + resources_.push_back({d3d9ObjOri, d3d9ObjShared}); } return errcode; diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_d3d9_amd.hpp b/projects/clr/opencl/api/opencl/amdocl/cl_d3d9_amd.hpp index 4424addb55..2bf5bd2f34 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_d3d9_amd.hpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_d3d9_amd.hpp @@ -239,9 +239,8 @@ public: if(pD3D9Aux_) pD3D9Aux_->Release(); if(pQuery_) pQuery_->Release(); //if the resouce is being used - std::vector>::iterator it; if(resources_.size()) { - for(it = resources_.begin(); it != resources_.end(); ++it) { + for(auto& it = resources_.cbegin(); it != resources_.cend(); it++) { if( surfInfo_.resource && ((*it).first.surfInfo.resource == surfInfo_.resource) && ((*it).first.surfPlane == surfPlane_)) { diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_gl.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_gl.cpp index 8a367cbe4f..5bc3cec800 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_gl.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_gl.cpp @@ -199,9 +199,8 @@ RUNTIME_ENTRY_RET(cl_mem, clCreateFromGLTexture, const std::vector& devices = as_amd(context)->devices(); bool supportPass = false; bool sizePass = false; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().imageSupport_) { + for (const auto& it : devices) { + if (it->info().imageSupport_) { supportPass = true; } } @@ -285,9 +284,8 @@ RUNTIME_ENTRY_RET(cl_mem, clCreateFromGLTexture2D, const std::vector& devices = as_amd(context)->devices(); bool supportPass = false; bool sizePass = false; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().imageSupport_) { + for (const auto& it : devices) { + if (it->info().imageSupport_) { supportPass = true; } } @@ -366,9 +364,8 @@ RUNTIME_ENTRY_RET(cl_mem, clCreateFromGLTexture3D, const std::vector& devices = as_amd(context)->devices(); bool supportPass = false; bool sizePass = false; - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().imageSupport_) { + for (const auto& it : devices) { + if (it->info().imageSupport_) { supportPass = true; } } @@ -914,9 +911,8 @@ RUNTIME_ENTRY(cl_int, clGetGLContextInfoKHR, if (param_value != NULL) { cl_device_id* deviceList = (cl_device_id*)param_value; - std::vector::const_iterator it; - for (it = compatible_devices.begin(); it != compatible_devices.end(); ++it) { - *deviceList++ = as_cl(*it); + for (const auto& it : compatible_devices) { + *deviceList++ = as_cl(it); } } @@ -1616,8 +1612,8 @@ cl_mem clCreateFromGLBufferAMD(Context& amdContext, cl_mem_flags flags, GLuint b // We should come up with a more elegant solution to handle this. assert(amdContext.devices().size() == 1); - std::vector::const_iterator itr = amdContext.devices().begin(); - amd::Device& dev = *(*itr); + const auto it = amdContext.devices().cbegin(); + const amd::Device& dev = *(*it); if (dev.type() != CL_DEVICE_TYPE_CPU) { device::Memory* mem = pBufferGL->getDeviceMemory(dev); diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_pipe.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_pipe.cpp index 4b91fc19a0..6cc17882fd 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_pipe.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_pipe.cpp @@ -81,10 +81,9 @@ RUNTIME_ENTRY_RET(cl_mem, clCreatePipe, size_t size = sizeof(struct clk_pipe_t) + pipe_packet_size * pipe_max_packets; const std::vector& devices = as_amd(context)->devices(); - std::vector::const_iterator it; bool sizePass = false; - for (it = devices.begin(); it != devices.end(); ++it) { - if (((*it)->info().maxMemAllocSize_ >= size)) { + for (const auto& it : devices) { + if (it->info().maxMemAllocSize_ >= size) { sizePass = true; break; } diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_program.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_program.cpp index d1f4ac0eb8..be4eec8227 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_program.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_program.cpp @@ -24,9 +24,8 @@ static amd::Program* createProgram(cl_context context, cl_uint num_devices, // Add programs for all devices in the context. if (device_list == NULL) { const std::vector& devices = as_amd(context)->devices(); - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if (program->addDeviceProgram(**it) == CL_OUT_OF_HOST_MEMORY) { + for (const auto& it : devices) { + if (program->addDeviceProgram(*it) == CL_OUT_OF_HOST_MEMORY) { *not_null(errcode_ret) = CL_OUT_OF_HOST_MEMORY; program->release(); return NULL; @@ -142,9 +141,8 @@ RUNTIME_ENTRY_RET(cl_program, clCreateProgramWithSource, // Add programs for all devices in the context. const std::vector& devices = as_amd(context)->devices(); - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if (program->addDeviceProgram(**it) == CL_OUT_OF_HOST_MEMORY) { + for (const auto& it : devices) { + if (program->addDeviceProgram(*it) == CL_OUT_OF_HOST_MEMORY) { *not_null(errcode_ret) = CL_OUT_OF_HOST_MEMORY; program->release(); return (cl_program)0; @@ -204,9 +202,8 @@ RUNTIME_ENTRY_RET(cl_program, clCreateProgramWithIL, // Add programs for all devices in the context. const std::vector& devices = as_amd(context)->devices(); - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if (program->addDeviceProgram(**it, il, length) == CL_OUT_OF_HOST_MEMORY) { + for (const auto& it : devices) { + if (program->addDeviceProgram(*it, il, length) == CL_OUT_OF_HOST_MEMORY) { *not_null(errcode_ret) = CL_OUT_OF_HOST_MEMORY; program->release(); return (cl_program)0; @@ -368,9 +365,8 @@ RUNTIME_ENTRY_RET(cl_program, clCreateProgramWithAssemblyAMD, // Add programs for all devices in the context. const std::vector& devices = as_amd(context)->devices(); - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if (program->addDeviceProgram(**it) == CL_OUT_OF_HOST_MEMORY) { + for (const auto& it : devices) { + if (program->addDeviceProgram(*it) == CL_OUT_OF_HOST_MEMORY) { *not_null(errcode_ret) = CL_OUT_OF_HOST_MEMORY; program->release(); return (cl_program)0; @@ -964,9 +960,8 @@ RUNTIME_ENTRY(cl_int, clGetProgramInfo, *not_null(param_value_size_ret) = valueSize; if (param_value != NULL) { cl_device_id* device_list = (cl_device_id*)param_value; - amd::Program::devicelist_t::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - *device_list++ = const_cast(as_cl(*it)); + for (const auto& it : devices) { + *device_list++ = const_cast(as_cl(it)); } if (param_value_size > valueSize) { ::memset(static_cast
(param_value) + valueSize, '\0', @@ -991,9 +986,8 @@ RUNTIME_ENTRY(cl_int, clGetProgramInfo, *not_null(param_value_size_ret) = valueSize; if (param_value != NULL) { size_t* binary_sizes = (size_t*)param_value; - amd::Program::devicelist_t::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - *binary_sizes++ = amdProgram->getDeviceProgram(**it)->binary().second; + for (const auto& it : devices) { + *binary_sizes++ = amdProgram->getDeviceProgram(*it)->binary().second; } if (param_value_size > valueSize) { ::memset(static_cast
(param_value) + valueSize, '\0', @@ -1014,9 +1008,8 @@ RUNTIME_ENTRY(cl_int, clGetProgramInfo, *not_null(param_value_size_ret) = valueSize; if (param_value != NULL) { char** binaries = (char**)param_value; - amd::Program::devicelist_t::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - const device::Program::binary_t& binary = amdProgram->getDeviceProgram(**it)->binary(); + for (const auto& it : devices) { + const device::Program::binary_t& binary = amdProgram->getDeviceProgram(*it)->binary(); // If an entry value in the array is NULL, // then runtime should skip copying the program binary if (*binaries != NULL) { @@ -1376,9 +1369,8 @@ RUNTIME_ENTRY(cl_int, clCreateKernelsInProgram, (cl_program program, cl_uint num const amd::Program::symbols_t& symbols = as_amd(program)->symbols(); cl_kernel* result = kernels; - amd::Program::symbols_t::const_iterator it; - for (it = symbols.begin(); it != symbols.end(); ++it) { - amd::Kernel* kernel = new amd::Kernel(*as_amd(program), it->second, it->first); + for (const auto& it : symbols) { + amd::Kernel* kernel = new amd::Kernel(*as_amd(program), it.second, it.first); if (kernel == NULL) { while (--result >= kernels) { as_amd(*result)->release(); diff --git a/projects/clr/opencl/api/opencl/amdocl/cl_svm.cpp b/projects/clr/opencl/api/opencl/amdocl/cl_svm.cpp index 8cc2907606..a0c3ecee5b 100644 --- a/projects/clr/opencl/api/opencl/amdocl/cl_svm.cpp +++ b/projects/clr/opencl/api/opencl/amdocl/cl_svm.cpp @@ -74,7 +74,7 @@ static bool validateMapFlags(cl_map_flags flags) { * \param size is the size in bytes of the SVM buffer to be allocated. * * \param alignment is the minimum alignment in bytes that is required for the - * newly created buffer’s memory region. It must be a power of two up to the + * newly created buffer’s memory region. It must be a power of two up to the * largest data type supported by the OpenCL device. For the full profile, the * largest data type is long16. For the embedded profile, it is long16 if the * device supports 64-bit integers; otherwise it is int16. If alignment is 0, a @@ -127,25 +127,24 @@ RUNTIME_ENTRY_RET_NOERRCODE(void*, clSVMAlloc, (cl_context context, cl_svm_mem_f cl_device_svm_capabilities combinedSvmCapabilities = 0; const cl_uint hostAddressBits = LP64_SWITCH(32, 64); cl_uint minContextAlignment = std::numeric_limits::max(); - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - cl_device_svm_capabilities svmCapabilities = (*it)->info().svmCapabilities_; + for (const auto& it : devices) { + cl_device_svm_capabilities svmCapabilities = it->info().svmCapabilities_; if (svmCapabilities == 0) { continue; } combinedSvmCapabilities |= svmCapabilities; - if ((*it)->info().maxMemAllocSize_ >= size) { + if (it->info().maxMemAllocSize_ >= size) { sizePass = true; } - if ((*it)->info().addressBits_ < hostAddressBits) { + if (it->info().addressBits_ < hostAddressBits) { LogWarning("address mode mismatch between host and device"); return NULL; } // maximum alignment for a device is given in bits. - cl_uint baseAlignment = (*it)->info().memBaseAddrAlign_ >> 3; + cl_uint baseAlignment = it->info().memBaseAddrAlign_ >> 3; if (alignment > baseAlignment) { LogWarning("invalid parameter \"alignment\""); return NULL; @@ -877,7 +876,7 @@ RUNTIME_EXIT * (clEnqueueNDRangeKernel) until the argument value is changed by a call to * clSetKernelArgSVMPointer for \a kernel. The SVM pointer can only be used for * arguments that are declared to be a pointer to global or constant memory. - * The SVM pointer value must be aligned according to the argument’s type. For + * The SVM pointer value must be aligned according to the argument�s type. For * example, if the argument is declared to be global float4 *p, the SVM pointer * value passed for p must be at a minimum aligned to a float4. The SVM pointer * value specified as the argument value can be the pointer returned by @@ -985,9 +984,8 @@ RUNTIME_ENTRY(cl_int, clSetKernelExecInfo, (cl_kernel kernel, cl_kernel_exec_inf const amd::Context* amdContext = &amdKernel->program().context(); bool foundFineGrainedSystemDevice = false; const std::vector& devices = amdContext->devices(); - std::vector::const_iterator it; - for (it = devices.begin(); it != devices.end(); ++it) { - if ((*it)->info().svmCapabilities_ & CL_DEVICE_SVM_FINE_GRAIN_SYSTEM) { + for (const auto it : devices) { + if (it->info().svmCapabilities_ & CL_DEVICE_SVM_FINE_GRAIN_SYSTEM) { foundFineGrainedSystemDevice = true; break; }